cuda/src/full_chain_algorithm.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 // Local include(s).
0009 #include "traccc/examples/cuda/full_chain_algorithm.hpp"
0010
0011 // Project include(s).
0012 #include "traccc/cuda/utils/make_magnetic_field.hpp"
0013 #include "traccc/seeding/detail/track_params_estimation_config.hpp"
0014
0015 // CUDA include(s).
0016 #include <cuda_runtime_api.h>
0017
0018 // System include(s).
0019 #include <iostream>
0020 #include <stdexcept>
0021
0022 /// Helper macro for checking the return value of CUDA function calls
0023 #define CUDA_ERROR_CHECK(EXP)                                            \
0024   do {                                                                   \
0025     const cudaError_t errorCode = EXP;                                   \
0026     if (errorCode != cudaSuccess) {                                      \
0027       throw std::runtime_error(std::string("Failed to run " #EXP " (") + \
0028                                cudaGetErrorString(errorCode) + ")");     \
0029     }                                                                    \
0030   } while (false)
0031
0032 namespace traccc::cuda {
0033
0034 full_chain_algorithm::full_chain_algorithm(
0035     vecmem::memory_resource& host_mr,
0036     const clustering_config& clustering_config,
0037     const seedfinder_config& finder_config,
0038     const spacepoint_grid_config& grid_config,
0039     const seedfilter_config& filter_config,
0040     const gbts_seedfinder_config& gbts_config,
0041     const track_params_estimation_config& track_params_estimation_config,
0042     const finding_algorithm::config_type& finding_config,
0043     const fitting_algorithm::config_type& fitting_config,
0044     const detector_design_description::host& det_descr,
0045     const detector_conditions_description::host& det_cond,
0046     const magnetic_field& field, host_detector* detector,
0047     std::unique_ptr<const traccc::Logger> logger, bool useGBTS)
0048     : messaging(logger->clone()),
0049       m_host_mr(host_mr),
0050       m_pinned_host_mr(),
0051       m_cached_pinned_host_mr(m_pinned_host_mr),
0052       m_vecmem_stream{},
0053       m_stream{m_vecmem_stream.stream()},
0054       m_device_mr(),
0055       m_cached_device_mr(m_device_mr),
0056       m_copy(m_stream.cudaStream()),
0057       m_field_vec{0.f, 0.f, finder_config.bFieldInZ},
0058       m_field(make_magnetic_field(field)),
0059       m_det_descr(det_descr),
0060       m_det_cond(det_cond),
0061       m_device_det_descr(
0062           [&]() {
0063             // number of elements in the detector design description
0064             std::vector<unsigned int> sizes(det_descr.size());
0065             for (std::size_t i = 0; i < det_descr.size(); ++i) {
0066               auto this_design = det_descr.at(i);
0067               // now for each element, set the size to the largest size of
0068               // that element across all modules
0069               sizes[i] = std::max(static_cast<unsigned int>(
0070                                       ((this_design.bin_edges_x()).size())),
0071                                   static_cast<unsigned int>(
0072                                       ((this_design.bin_edges_y()).size())));
0073             }
0074             return sizes;
0075           }(),
0076           m_device_mr, &m_host_mr, vecmem::data::buffer_type::resizable),
0077       m_device_det_cond(
0078           static_cast<detector_conditions_description::buffer::size_type>(
0079               m_det_cond.get().size()),
0080           m_device_mr),
0081       m_detector(detector),
0082       m_clusterization({m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0083                        m_stream, clustering_config),
0084       m_measurement_sorting({m_cached_device_mr, &m_cached_pinned_host_mr},
0085                             m_copy, m_stream,
0086                             logger->cloneWithSuffix("MeasSortingAlg")),
0087       m_spacepoint_formation({m_cached_device_mr, &m_cached_pinned_host_mr},
0088                              m_copy, m_stream,
0089                              logger->cloneWithSuffix("SpFormationAlg")),
0090       m_seeding(finder_config, grid_config, filter_config,
0091                 {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0092                 m_stream, logger->cloneWithSuffix("SeedingAlg")),
0093       m_gbts_seeding(gbts_config,
0094                      {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0095                      m_stream, logger->cloneWithSuffix("GbtsAlg")),
0096       m_track_parameter_estimation(
0097           track_params_estimation_config,
0098           {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy, m_stream,
0099           logger->cloneWithSuffix("TrackParEstAlg")),
0100       m_finding(finding_config, {m_cached_device_mr, &m_cached_pinned_host_mr},
0101                 m_copy, m_stream, logger->cloneWithSuffix("TrackFindingAlg")),
0102       m_fitting(fitting_config, {m_cached_device_mr, &m_cached_pinned_host_mr},
0103                 m_copy, m_stream, logger->cloneWithSuffix("TrackFittingAlg")),
0104       m_clustering_config(clustering_config),
0105       m_finder_config(finder_config),
0106       m_grid_config(grid_config),
0107       m_filter_config(filter_config),
0108       m_gbts_config(gbts_config),
0109       m_track_params_estimation_config(track_params_estimation_config),
0110       m_finding_config(finding_config),
0111       m_fitting_config(fitting_config),
0112       usingGBTS(useGBTS) {
0113   // Tell the user what device is being used.
0114   int device = 0;
0115   CUDA_ERROR_CHECK(cudaGetDevice(&device));
0116   cudaDeviceProp props;
0117   CUDA_ERROR_CHECK(cudaGetDeviceProperties(&props, device));
0118   std::cout << "Using CUDA device: " << props.name << " [id: " << device
0119             << ", bus: " << props.pciBusID << ", device: " << props.pciDeviceID
0120             << "]" << std::endl;
0121
0122   // Copy the detector (description) to the device.
0123   m_copy.setup(m_device_det_descr)->wait();
0124   m_copy(vecmem::get_data(m_det_descr.get()), m_device_det_descr)->wait();
0125   m_copy(vecmem::get_data(m_det_cond.get()), m_device_det_cond)->wait();
0126   if (m_detector != nullptr) {
0127     m_device_detector =
0128         traccc::buffer_from_host_detector(*m_detector, m_device_mr, m_copy);
0129   }
0130 }
0131
0132 full_chain_algorithm::full_chain_algorithm(const full_chain_algorithm& parent)
0133     : messaging(parent.logger().clone()),
0134       m_host_mr(parent.m_host_mr),
0135       m_pinned_host_mr(),
0136       m_cached_pinned_host_mr(m_pinned_host_mr),
0137       m_vecmem_stream{},
0138       m_stream{m_vecmem_stream.stream()},
0139       m_device_mr(),
0140       m_cached_device_mr(m_device_mr),
0141       m_copy(m_stream.cudaStream()),
0142       m_field_vec(parent.m_field_vec),
0143       m_field(parent.m_field),
0144       m_det_descr(parent.m_det_descr),
0145       m_det_cond(parent.m_det_cond),
0146       m_device_det_descr(
0147           [&]() {
0148             // number of elements in the detector design description
0149             std::vector<unsigned int> sizes(parent.m_det_descr.get().size());
0150             for (std::size_t i = 0; i < parent.m_det_descr.get().size(); ++i) {
0151               auto this_design = parent.m_det_descr.get().at(i);
0152               // now for each element, set the size to the largest size of
0153               // that element across all modules
0154               sizes[i] = std::max(static_cast<unsigned int>(
0155                                       ((this_design.bin_edges_x()).size())),
0156                                   static_cast<unsigned int>(
0157                                       ((this_design.bin_edges_y()).size())));
0158             }
0159             return sizes;
0160           }(),
0161           m_device_mr, &m_host_mr, vecmem::data::buffer_type::resizable),
0162       m_device_det_cond(
0163           static_cast<detector_conditions_description::buffer::size_type>(
0164               m_det_cond.get().size()),
0165           m_device_mr),
0166       m_detector(parent.m_detector),
0167       m_clusterization({m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0168                        m_stream, parent.m_clustering_config),
0169       m_measurement_sorting({m_cached_device_mr, &m_cached_pinned_host_mr},
0170                             m_copy, m_stream,
0171                             parent.logger().cloneWithSuffix("MeasSortingAlg")),
0172       m_spacepoint_formation({m_cached_device_mr, &m_cached_pinned_host_mr},
0173                              m_copy, m_stream,
0174                              parent.logger().cloneWithSuffix("SpFormationAlg")),
0175       m_seeding(parent.m_finder_config, parent.m_grid_config,
0176                 parent.m_filter_config,
0177                 {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0178                 m_stream, parent.logger().cloneWithSuffix("SeedingAlg")),
0179       m_gbts_seeding(parent.m_gbts_config,
0180                      {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0181                      m_stream, parent.logger().cloneWithSuffix("GbtsAlg")),
0182       m_track_parameter_estimation(
0183           parent.m_track_params_estimation_config,
0184           {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy, m_stream,
0185           parent.logger().cloneWithSuffix("TrackParamEstAlg")),
0186       m_finding(parent.m_finding_config,
0187                 {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0188                 m_stream, parent.logger().cloneWithSuffix("TrackFindingAlg")),
0189       m_fitting(parent.m_fitting_config,
0190                 {m_cached_device_mr, &m_cached_pinned_host_mr}, m_copy,
0191                 m_stream, parent.logger().cloneWithSuffix("TrackFittingAlg")),
0192       m_clustering_config(parent.m_clustering_config),
0193       m_finder_config(parent.m_finder_config),
0194       m_grid_config(parent.m_grid_config),
0195       m_filter_config(parent.m_filter_config),
0196       m_gbts_config(parent.m_gbts_config),
0197       m_track_params_estimation_config(parent.m_track_params_estimation_config),
0198       m_finding_config(parent.m_finding_config),
0199       m_fitting_config(parent.m_fitting_config),
0200       usingGBTS(parent.usingGBTS) {
0201   m_copy.setup(m_device_det_descr)->wait();
0202   m_copy(vecmem::get_data(m_det_descr.get()), m_device_det_descr)->wait();
0203   m_copy(vecmem::get_data(m_det_cond.get()), m_device_det_cond)->wait();
0204   if (m_detector != nullptr) {
0205     m_device_detector =
0206         traccc::buffer_from_host_detector(*m_detector, m_device_mr, m_copy);
0207   }
0208 }
0209
0210 full_chain_algorithm::~full_chain_algorithm() = default;
0211
0212 full_chain_algorithm::output_type full_chain_algorithm::operator()(
0213     const edm::silicon_cell_collection::host& cells) const {
0214   // Create device copy of input collections
0215   edm::silicon_cell_collection::buffer cells_buffer(
0216       static_cast<unsigned int>(cells.size()), m_cached_device_mr);
0217   m_copy(vecmem::get_data(cells), cells_buffer)->ignore();
0218
0219   // Run the clusterization (asynchronously).
0220   const auto unsorted_measurements =
0221       m_clusterization(cells_buffer, m_device_det_descr, m_device_det_cond);
0222   const measurement_sorting_algorithm::output_type measurements =
0223       m_measurement_sorting(unsorted_measurements);
0224
0225   // If we have a Detray detector, run the seeding, track finding and fitting.
0226   if (m_detector != nullptr) {
0227     // Run the seed-finding (asynchronously).
0228     const spacepoint_formation_algorithm::output_type spacepoints =
0229         m_spacepoint_formation(m_device_detector, measurements);
0230
0231     triplet_seeding_algorithm::output_type seeds;
0232     if (usingGBTS) {
0233       seeds = m_gbts_seeding(spacepoints, measurements);
0234     } else {
0235       seeds = m_seeding(spacepoints);
0236     }
0237     const seed_parameter_estimation_algorithm::output_type track_params =
0238         m_track_parameter_estimation(m_field, measurements, spacepoints, seeds);
0239
0240     // Run the track finding (asynchronously).
0241     const finding_algorithm::output_type track_candidates =
0242         m_finding(m_device_detector, m_field, measurements, track_params);
0243
0244     // Copy a limited amount of result data back to the host.
0245     const auto host_tracks =
0246         m_copy.to(track_candidates.tracks, m_cached_pinned_host_mr, nullptr,
0247                   vecmem::copy::type::device_to_host);
0248     output_type result{m_host_mr};
0249     vecmem::copy host_copy;
0250     host_copy(host_tracks, result)->wait();
0251     return result;
0252
0253   }
0254   // If not, copy the measurements back to the host, and return a dummy
0255   // object.
0256   else {
0257     // Copy the measurements back to the host.
0258     edm::measurement_collection::host measurements_host(m_host_mr);
0259     m_copy(measurements, measurements_host)->wait();
0260
0261     // Return an empty object.
0262     return output_type{m_host_mr};
0263   }
0264 }
0265
0266 bound_track_parameters_collection_types::host full_chain_algorithm::seeding(
0267     const edm::silicon_cell_collection::host& cells) const {
0268   // Create device copy of input collections
0269   edm::silicon_cell_collection::buffer cells_buffer(
0270       static_cast<unsigned int>(cells.size()), m_cached_device_mr);
0271   m_copy(vecmem::get_data(cells), cells_buffer)->ignore();
0272
0273   // Run the clusterization (asynchronously).
0274   const auto unsorted_measurements =
0275       m_clusterization(cells_buffer, m_device_det_descr, m_device_det_cond);
0276   const measurement_sorting_algorithm::output_type measurements =
0277       m_measurement_sorting(unsorted_measurements);
0278
0279   // If we have a Detray detector, run the seeding, track finding and fitting.
0280   if (m_detector != nullptr) {
0281     // Run the seed-finding (asynchronously).
0282     const spacepoint_formation_algorithm::output_type spacepoints =
0283         m_spacepoint_formation(m_device_detector, measurements);
0284
0285     triplet_seeding_algorithm::output_type seeds;
0286     if (usingGBTS) {
0287       seeds = m_gbts_seeding(spacepoints, measurements);
0288     } else {
0289       seeds = m_seeding(spacepoints);
0290     }
0291     const seed_parameter_estimation_algorithm::output_type track_params =
0292         m_track_parameter_estimation(m_field, measurements, spacepoints, seeds);
0293
0294     // Copy a limited amount of result data back to the host.
0295     const auto host_seeds = m_copy.to(track_params, m_cached_pinned_host_mr,
0296                                       vecmem::copy::type::device_to_host);
0297     bound_track_parameters_collection_types::host result{&m_host_mr};
0298     vecmem::copy host_copy;
0299     host_copy(host_seeds, result)->wait();
0300     return result;
0301
0302   }
0303   // If not, copy the measurements back to the host, and return a dummy
0304   // object.
0305   else {
0306     // Copy the measurements back to the host.
0307     edm::measurement_collection::host measurements_host(m_host_mr);
0308     m_copy(measurements, measurements_host)->wait();
0309
0310     // Return an empty object.
0311     return {};
0312   }
0313 }
0314
0315 }  // namespace traccc::cuda