tests/cuda/test_kalman_fitter_telescope.cpp

0001 /** TRACCC library, part of the ACTS project (R&D line)
0002  *
0003  * (c) 2022-2026 CERN for the benefit of the ACTS project
0004  *
0005  * Mozilla Public License Version 2.0
0006  */
0007
0008 // Project include(s).
0009 #include "traccc/bfield/construct_const_bfield.hpp"
0010 #include "traccc/bfield/magnetic_field_types.hpp"
0011 #include "traccc/cuda/fitting/kalman_fitting_algorithm.hpp"
0012 #include "traccc/edm/track_container.hpp"
0013 #include "traccc/geometry/host_detector.hpp"
0014 #include "traccc/io/detector.hpp"
0015 #include "traccc/io/utils.hpp"
0016 #include "traccc/performance/details/is_same_object.hpp"
0017 #include "traccc/resolution/fitting_performance_writer.hpp"
0018 #include "traccc/simulation/event_generators.hpp"
0019 #include "traccc/simulation/simulator.hpp"
0020 #include "traccc/utils/memory_resource.hpp"
0021 #include "traccc/utils/ranges.hpp"
0022 #include "traccc/utils/seed_generator.hpp"
0023
0024 // Test include(s).
0025 #include "tests/kalman_fitting_telescope_test.hpp"
0026
0027 // VecMem include(s).
0028 #include <vecmem/memory/cuda/device_memory_resource.hpp>
0029 #include <vecmem/memory/cuda/managed_memory_resource.hpp>
0030 #include <vecmem/memory/host_memory_resource.hpp>
0031 #include <vecmem/utils/cuda/async_copy.hpp>
0032 #include <vecmem/utils/cuda/copy.hpp>
0033 #include <vecmem/utils/cuda/stream_wrapper.hpp>
0034
0035 // GTest include(s).
0036 #include <gtest/gtest.h>
0037
0038 // System include(s).
0039 #include <filesystem>
0040 #include <string>
0041
0042 using namespace traccc;
0043
0044 // This defines the local frame test suite
0045 TEST_P(KalmanFittingTelescopeTests, Run) {
0046   // Get the parameters
0047   const std::string name = std::get<0>(GetParam());
0048   const std::array<scalar, 3u> origin = std::get<1>(GetParam());
0049   const std::array<scalar, 3u> origin_stddev = std::get<2>(GetParam());
0050   const std::array<scalar, 2u> mom_range = std::get<3>(GetParam());
0051   const std::array<scalar, 2u> eta_range = std::get<4>(GetParam());
0052   const std::array<scalar, 2u> theta_range = eta_to_theta_range(eta_range);
0053   const std::array<scalar, 2u> phi_range = std::get<5>(GetParam());
0054   const traccc::pdg_particle<scalar> ptc = std::get<6>(GetParam());
0055   const unsigned int n_truth_tracks = std::get<7>(GetParam());
0056   const unsigned int n_events = std::get<8>(GetParam());
0057   const bool random_charge = std::get<9>(GetParam());
0058
0059   // Performance writer
0060   traccc::fitting_performance_writer::config fit_writer_cfg;
0061   fit_writer_cfg.file_path = "performance_track_fitting_" + name + ".root";
0062   traccc::fitting_performance_writer fit_performance_writer(
0063       fit_writer_cfg, traccc::getDefaultLogger("FittingPerformanceWriter",
0064                                                traccc::Logging::Level::INFO));
0065
0066   /*****************************
0067    * Build a telescope geometry
0068    *****************************/
0069
0070   // Memory resources used by the application.
0071   vecmem::host_memory_resource host_mr;
0072   vecmem::cuda::device_memory_resource device_mr;
0073   traccc::memory_resource mr{device_mr, &host_mr};
0074   vecmem::cuda::managed_memory_resource mng_mr;
0075
0076   // Read back detector file
0077   const std::string path = name + "/";
0078   detray::io::detector_reader_config reader_cfg{};
0079   reader_cfg.add_file(path + "telescope_detector_geometry.json")
0080       .add_file(path + "telescope_detector_homogeneous_material.json");
0081   auto [host_det, names] =
0082       detray::io::read_detector<host_detector_type>(mng_mr, reader_cfg);
0083
0084   traccc::host_detector polymorphic_detector;
0085   polymorphic_detector.set<detector_traits>(std::move(host_det));
0086
0087   const auto field = traccc::construct_const_bfield(std::get<13>(GetParam()));
0088
0089   /***************************
0090    * Generate simulation data
0091    ***************************/
0092
0093   // Track generator
0094   using generator_type =
0095       detray::random_track_generator<traccc::free_track_parameters<>,
0096                                      uniform_gen_t>;
0097   generator_type::configuration gen_cfg{};
0098   gen_cfg.n_tracks(n_truth_tracks);
0099   gen_cfg.origin(origin);
0100   gen_cfg.origin_stddev(origin_stddev);
0101   gen_cfg.phi_range(phi_range[0], phi_range[1]);
0102   gen_cfg.theta_range(theta_range[0], theta_range[1]);
0103   gen_cfg.mom_range(mom_range[0], mom_range[1]);
0104   gen_cfg.randomize_charge(random_charge);
0105   generator_type generator(gen_cfg);
0106
0107   // Smearing value for measurements
0108   traccc::measurement_smearer<traccc::default_algebra> meas_smearer(
0109       smearing[0], smearing[1]);
0110
0111   using writer_type = traccc::smearing_writer<
0112       traccc::measurement_smearer<traccc::default_algebra>>;
0113
0114   typename writer_type::config smearer_writer_cfg{meas_smearer};
0115   traccc::seed_generator<host_detector_type>::config seed_cfg{};
0116   seed_cfg.initial_sigmas = stddevs;
0117
0118   // Run simulator
0119   const std::string full_path = io::data_directory() + path;
0120   std::filesystem::create_directories(full_path);
0121   auto sim = traccc::simulator<host_detector_type, b_field_t, generator_type,
0122                                writer_type>(
0123       ptc, n_events, polymorphic_detector.as<detector_traits>(),
0124       field.as_field<traccc::const_bfield_backend_t<traccc::scalar>>(),
0125       std::move(generator), std::move(smearer_writer_cfg), full_path);
0126   sim.run();
0127
0128   /***************
0129    * Run fitting
0130    ***************/
0131
0132   // Stream object
0133   vecmem::cuda::stream_wrapper vecmem_stream;
0134   traccc::cuda::stream_wrapper stream{vecmem_stream.stream()};
0135
0136   // Copy objects
0137   vecmem::cuda::async_copy copy{stream.cudaStream()};
0138
0139   const traccc::detector_buffer detector_buffer =
0140       traccc::buffer_from_host_detector(polymorphic_detector, device_mr, copy);
0141
0142   // Seed generator
0143   seed_generator<host_detector_type> sg(
0144       polymorphic_detector.as<detector_traits>(), seed_cfg);
0145
0146   // Fitting algorithm object
0147   traccc::cuda::kalman_fitting_algorithm::config_type fit_cfg;
0148   fit_cfg.ptc_hypothesis = ptc;
0149   fit_cfg.min_pT = 100.f * traccc::unit<float>::MeV;
0150   traccc::cuda::kalman_fitting_algorithm device_fitting(fit_cfg, mr, copy,
0151                                                         stream);
0152
0153   // Iterate over events
0154   for (std::size_t i_evt = 0; i_evt < n_events; i_evt++) {
0155     // Event map
0156     traccc::event_data evt_data(path, i_evt, host_mr);
0157
0158     // Truth Track Candidates
0159     traccc::edm::measurement_collection::host measurements(host_mr);
0160     traccc::edm::track_container<traccc::default_algebra>::host
0161         track_candidates{host_mr};
0162     evt_data.generate_truth_candidates(track_candidates, measurements, sg,
0163                                        host_mr);
0164     track_candidates.measurements = vecmem::get_data(measurements);
0165
0166     // n_trakcs = 100
0167     ASSERT_EQ(track_candidates.tracks.size(), n_truth_tracks);
0168
0169     // track candidates buffer
0170     traccc::edm::measurement_collection::buffer measurements_buffer =
0171         copy.to(track_candidates.measurements, mr.main, mr.host,
0172                 vecmem::copy::type::host_to_device);
0173     traccc::edm::track_container<traccc::default_algebra>::buffer
0174         track_candidates_buffer{
0175             copy.to(vecmem::get_data(track_candidates.tracks), mr.main, mr.host,
0176                     vecmem::copy::type::host_to_device),
0177             {},
0178             measurements_buffer};
0179
0180     // Run fitting
0181     auto track_states_cuda_buffer =
0182         device_fitting(detector_buffer, field, track_candidates_buffer);
0183
0184     traccc::edm::track_container<traccc::default_algebra>::host
0185         track_states_cuda{host_mr};
0186     copy(track_states_cuda_buffer.tracks, track_states_cuda.tracks,
0187          vecmem::copy::type::device_to_host)
0188         ->wait();
0189     copy(track_states_cuda_buffer.states, track_states_cuda.states,
0190          vecmem::copy::type::device_to_host)
0191         ->wait();
0192
0193     const std::size_t n_fitted_tracks =
0194         count_successfully_fitted_tracks(track_states_cuda.tracks);
0195
0196     ASSERT_EQ(track_states_cuda.tracks.size(), n_truth_tracks);
0197     ASSERT_EQ(track_states_cuda.tracks.size(), n_fitted_tracks);
0198
0199     for (std::size_t i_trk = 0; i_trk < n_truth_tracks; i_trk++) {
0200       consistency_tests(track_states_cuda.tracks.at(i_trk),
0201                         track_states_cuda.states);
0202
0203       ndf_tests(track_states_cuda.tracks.at(i_trk), track_states_cuda.states,
0204                 measurements);
0205
0206       fit_performance_writer.write(
0207           track_states_cuda.tracks.at(i_trk), track_states_cuda.states,
0208           measurements, polymorphic_detector.as<detector_traits>(), evt_data);
0209     }
0210   }
0211
0212   fit_performance_writer.finalize();
0213
0214   /********************
0215    * Pull value test
0216    ********************/
0217
0218   static const std::vector<std::string> pull_names{
0219       "pull_d0", "pull_z0", "pull_phi", "pull_theta", "pull_qop"};
0220   pull_value_tests(fit_writer_cfg.file_path, pull_names);
0221
0222   /********************
0223    * P-value test
0224    ********************/
0225
0226   p_value_tests(fit_writer_cfg.file_path);
0227
0228   /********************
0229    * Success rate test
0230    ********************/
0231
0232   float success_rate = static_cast<float>(n_success) /
0233                        static_cast<float>(n_truth_tracks * n_events);
0234
0235   ASSERT_FLOAT_EQ(success_rate, 1.00f);
0236 }
0237
0238 INSTANTIATE_TEST_SUITE_P(
0239     CUDAKalmanFitTelescopeValidation, KalmanFittingTelescopeTests,
0240     ::testing::Values(
0241         std::make_tuple(
0242             "cuda_telescope_1_GeV_0_phi", std::array<scalar, 3u>{0.f, 0.f, 0.f},
0243             std::array<scalar, 3u>{0.f, 0.f, 0.f},
0244             std::array<scalar, 2u>{1.f, 1.f}, std::array<scalar, 2u>{0.f, 0.f},
0245             std::array<scalar, 2u>{0.f, 0.f}, traccc::muon<scalar>(), 100, 100,
0246             false, 20.f, 9u, 20.f, vector3{0, 0, 2 * traccc::unit<scalar>::T}),
0247         std::make_tuple("cuda_telescope_10_GeV_0_phi",
0248                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0249                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0250                         std::array<scalar, 2u>{10.f, 10.f},
0251                         std::array<scalar, 2u>{0.f, 0.f},
0252                         std::array<scalar, 2u>{0.f, 0.f},
0253                         traccc::muon<scalar>(), 100, 100, false, 20.f, 9u, 20.f,
0254                         vector3{0, 0, 2 * traccc::unit<scalar>::T}),
0255         std::make_tuple("cuda_telescope_100_GeV_0_phi",
0256                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0257                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0258                         std::array<scalar, 2u>{100.f, 100.f},
0259                         std::array<scalar, 2u>{0.f, 0.f},
0260                         std::array<scalar, 2u>{0.f, 0.f},
0261                         traccc::muon<scalar>(), 100, 100, false, 20.f, 9u, 20.f,
0262                         vector3{0, 0, 2 * traccc::unit<scalar>::T}),
0263         std::make_tuple("cuda_telescope_1_GeV_0_phi_antimuon",
0264                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0265                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0266                         std::array<scalar, 2u>{1.f, 1.f},
0267                         std::array<scalar, 2u>{0.f, 0.f},
0268                         std::array<scalar, 2u>{0.f, 0.f},
0269                         traccc::antimuon<scalar>(), 100, 100, false, 20.f, 9u,
0270                         20.f, vector3{0, 0, 2 * traccc::unit<scalar>::T}),
0271         std::make_tuple("cuda_telescope_1_GeV_0_phi_random_charge",
0272                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0273                         std::array<scalar, 3u>{0.f, 0.f, 0.f},
0274                         std::array<scalar, 2u>{1.f, 1.f},
0275                         std::array<scalar, 2u>{0.f, 0.f},
0276                         std::array<scalar, 2u>{0.f, 0.f},
0277                         traccc::muon<scalar>(), 100, 100, true, 20.f, 9u, 20.f,
0278                         vector3{0, 0, 2 * traccc::unit<scalar>::T})));