Track/test/TrackSoAHeterogeneous_test.cpp

0001 /**
0002    Simple test for the pixelTrack::TrackSoA data structure
0003    which inherits from PortableDeviceCollection.
0004
0005    Creates an instance of the class (automatically allocates
0006    memory on device), passes the view of the SoA data to
0007    the CUDA kernels which:
0008    - Fill the SoA with data.
0009    - Verify that the data written is correct.
0010
0011    Then, the SoA data are copied back to Host, where
0012    a temporary host-side view (tmp_view) is created using
0013    the same Layout to access the data on host and print it.
0014  */
0015
0016 #include <cstdint>
0017 #include "CUDADataFormats/Track/interface/TrackSoAHeterogeneousDevice.h"
0018 #include "CUDADataFormats/Track/interface/TrackSoAHeterogeneousHost.h"
0019 #include "HeterogeneousCore/CUDAUtilities/interface/requireDevices.h"
0020 #include "HeterogeneousCore/CUDAUtilities/interface/cudaCheck.h"
0021
0022 #include "Geometry/CommonTopologies/interface/SimplePixelTopology.h"
0023
0024 namespace testTrackSoA {
0025
0026   template <typename TrackerTraits>
0027   void runKernels(TrackSoAView<TrackerTraits> &tracks_view, cudaStream_t stream);
0028 }
0029
0030 int main() {
0031   cms::cudatest::requireDevices();
0032
0033   cudaStream_t stream;
0034   cudaCheck(cudaStreamCreateWithFlags(&stream, cudaStreamNonBlocking));
0035
0036   // Inner scope to deallocate memory before destroying the stream
0037   {
0038     // Instantiate tracks on device. PortableDeviceCollection allocates
0039     // SoA on device automatically.
0040     TrackSoAHeterogeneousDevice<pixelTopology::Phase1> tracks_d(stream);
0041     testTrackSoA::runKernels<pixelTopology::Phase1>(tracks_d.view(), stream);
0042
0043     // Instantate tracks on host. This is where the data will be
0044     // copied to from device.
0045     TrackSoAHeterogeneousHost<pixelTopology::Phase1> tracks_h(stream);
0046
0047     cudaCheck(cudaMemcpyAsync(
0048         tracks_h.buffer().get(), tracks_d.const_buffer().get(), tracks_d.bufferSize(), cudaMemcpyDeviceToHost, stream));
0049     cudaCheck(cudaStreamSynchronize(stream));
0050
0051     // Print results
0052     std::cout << "pt"
0053               << "\t"
0054               << "eta"
0055               << "\t"
0056               << "chi2"
0057               << "\t"
0058               << "quality"
0059               << "\t"
0060               << "nLayers"
0061               << "\t"
0062               << "hitIndices off" << std::endl;
0063
0064     for (int i = 0; i < 10; ++i) {
0065       std::cout << tracks_h.view()[i].pt() << "\t" << tracks_h.view()[i].eta() << "\t" << tracks_h.view()[i].chi2()
0066                 << "\t" << (int)tracks_h.view()[i].quality() << "\t" << (int)tracks_h.view()[i].nLayers() << "\t"
0067                 << tracks_h.view().hitIndices().off[i] << std::endl;
0068     }
0069   }
0070   cudaCheck(cudaStreamDestroy(stream));
0071
0072   return 0;
0073 }