Fixes, rebases, and clenups

CUDADataFormats/Track/interface/PixelTrackUtilities.h

@@ -24,7 +24,7 @@ template <typename TrackerTraits>
 struct trackSoA
 {
   static constexpr int32_t S = TrackerTraits::maxNumberOfTuples;
-  static constexpr int32_t H = TrackerTraits::maxHitsOnTrack;
+  static constexpr int32_t H = TrackerTraits::avgHitsPerTrack;
   // Aliases in order to not confuse the GENERATE_SOA_LAYOUT
   // macro with weird colons and angled brackets.
   using Vector5f = Eigen::Matrix<float, 5, 1>;

CUDADataFormats/Track/interface/TrackSoAHeterogeneousHost.h

@@ -9,7 +9,9 @@
 template <typename TrackerTraits>
 class TrackSoAHeterogeneousHost : public cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>> {
 public:
-  TrackSoAHeterogeneousHost() = default;
+
+  static constexpr int32_t S = TrackerTraits::maxNumberOfTuples; //TODO: this could be made configurable at runtime
+  explicit TrackSoAHeterogeneousHost() : cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>>(S) {}
 
   using cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>>::view;
   using cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>>::const_view;
@@ -18,7 +20,7 @@ class TrackSoAHeterogeneousHost : public cms::cuda::PortableHostCollection<Track
 
   // Constructor which specifies the SoA size
   explicit TrackSoAHeterogeneousHost(cudaStream_t stream)
-      : cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>>(TrackerTraits::maxNumberOfTuples, stream) {}
+      : cms::cuda::PortableHostCollection<TrackLayout<TrackerTraits>>(S, stream) {}
 };
 
 namespace pixelTrack {

CUDADataFormats/TrackingRecHit/interface/TrackingRecHitSoADevice.h

@@ -64,7 +64,7 @@ class TrackingRecHitSoADevice : public cms::cuda::PortableDeviceCollection<HitLa
     auto ret = cms::cuda::make_host_unique<float[]>(4 * nHits(), stream);
     size_t rowSize = sizeof(float) * nHits();
     // printf("nModules=%d \n", nModules());
-    printf("nHits=%d \n", nHits());
+    // printf("nHits=%d \n", nHits());
     cudaCheck(cudaMemcpyAsync(ret.get(), view().xLocal(), rowSize * 4, cudaMemcpyDeviceToHost, stream));
     // cudaCheck(cudaMemcpyAsync(ret.get() + rowSize , view().yLocal() , rowSize, cudaMemcpyDeviceToHost, stream));
     // cudaCheck(cudaMemcpyAsync(ret.get() + size_t(rowSize * 2), view().xerrLocal() , rowSize, cudaMemcpyDeviceToHost, stream));

CUDADataFormats/TrackingRecHit/interface/TrackingRecHitSoAHost.h

@@ -30,9 +30,25 @@ class TrackingRecHitSoAHost : public cms::cuda::PortableHostCollection<HitLayout
 
   // This SoA Host is used basically only for DQM
   // so we  just need a slim constructor
+  explicit TrackingRecHitSoAHost(uint32_t nHits) : cms::cuda::PortableHostCollection<HitLayout<TrackerTraits>>(nHits) {}
+
   explicit TrackingRecHitSoAHost(uint32_t nHits, cudaStream_t stream)
       : cms::cuda::PortableHostCollection<HitLayout<TrackerTraits>>(nHits, stream) {}
 
+  explicit TrackingRecHitSoAHost(uint32_t nHits,
+                                 int32_t offsetBPIX2,
+                                 ParamsOnGPU const* cpeParams,
+                                 uint32_t const* hitsModuleStart)
+         : cms::cuda::PortableHostCollection<HitLayout<TrackerTraits>>(nHits),
+           nHits_(nHits),
+           cpeParams_(cpeParams),
+           offsetBPIX2_(offsetBPIX2) {
+       view().nHits() = nHits;
+       std::copy(hitsModuleStart, hitsModuleStart + TrackerTraits::numberOfModules + 1, view().hitsModuleStart().begin());
+       memcpy(&(view().cpeParams()), cpeParams, sizeof(ParamsOnGPU));
+       view().offsetBPIX2() = offsetBPIX2;
+     }
+
   explicit TrackingRecHitSoAHost(uint32_t nHits,
                                  int32_t offsetBPIX2,
                                  ParamsOnGPU const* cpeParams,

CUDADataFormats/Vertex/interface/ZVertexSoAHeterogeneousHost.h

@@ -9,7 +9,8 @@
 template <int32_t S>
 class ZVertexSoAHeterogeneousHost : public cms::cuda::PortableHostCollection<ZVertexSoAHeterogeneousLayout<>> {
 public:
-  ZVertexSoAHeterogeneousHost() = default;  // Required for cms::cuda::Product
+
+  explicit ZVertexSoAHeterogeneousHost() : cms::cuda::PortableHostCollection<ZVertexSoAHeterogeneousLayout<>>(S) {}
 
   // Constructor which specifies the SoA size and CUDA stream
   explicit ZVertexSoAHeterogeneousHost(cudaStream_t stream)

DQM/SiPixelHeterogeneous/plugins/SiPixelCompareTrackSoA.cc

@@ -1,8 +1,8 @@
 // -*- C++ -*-
-// Package:    SiPixelPhase1CompareTrackSoA
-// Class:      SiPixelPhase1CompareTrackSoA
+// Package:    SiPixelCompareTrackSoA
+// Class:      SiPixelCompareTrackSoA
 //
-/**\class SiPixelPhase1CompareTrackSoA SiPixelPhase1CompareTrackSoA.cc
+/**\class SiPixelCompareTrackSoA SiPixelCompareTrackSoA.cc
 */
 //
 // Author: Suvankar Roy Chowdhury
@@ -63,19 +63,20 @@ namespace {
   }
 }  // namespace
 
-class SiPixelPhase1CompareTrackSoA : public DQMEDAnalyzer {
+template <typename T>
+class SiPixelCompareTrackSoA : public DQMEDAnalyzer {
 public:
-  using TrackSoAPhase1 = TrackSoAHeterogeneousHost<pixelTopology::Phase1>;
+  using PixelTrackSoA = TrackSoAHeterogeneousHost<T>;
 
-  explicit SiPixelPhase1CompareTrackSoA(const edm::ParameterSet&);
-  ~SiPixelPhase1CompareTrackSoA() override = default;
+  explicit SiPixelCompareTrackSoA(const edm::ParameterSet&);
+  ~SiPixelCompareTrackSoA() override = default;
   void bookHistograms(DQMStore::IBooker& ibooker, edm::Run const& iRun, edm::EventSetup const& iSetup) override;
   void analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) override;
   static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
 
 private:
-  const edm::EDGetTokenT<TrackSoAPhase1> tokenSoATrackCPU_;
-  const edm::EDGetTokenT<TrackSoAPhase1> tokenSoATrackGPU_;
+  const edm::EDGetTokenT<PixelTrackSoA> tokenSoATrackCPU_;
+  const edm::EDGetTokenT<PixelTrackSoA> tokenSoATrackGPU_;
   const std::string topFolderName_;
   const bool useQualityCut_;
   const pixelTrackSoA::Quality minQuality_;
@@ -115,9 +116,10 @@ class SiPixelPhase1CompareTrackSoA : public DQMEDAnalyzer {
 // constructors
 //
 
-SiPixelPhase1CompareTrackSoA::SiPixelPhase1CompareTrackSoA(const edm::ParameterSet& iConfig)
-    : tokenSoATrackCPU_(consumes(iConfig.getParameter<edm::InputTag>("pixelTrackSrcCPU"))),
-      tokenSoATrackGPU_(consumes(iConfig.getParameter<edm::InputTag>("pixelTrackSrcGPU"))),
+template <typename T>
+SiPixelCompareTrackSoA<T>::SiPixelCompareTrackSoA(const edm::ParameterSet& iConfig)
+    : tokenSoATrackCPU_(consumes<PixelTrackSoA>(iConfig.getParameter<edm::InputTag>("pixelTrackSrcCPU"))),
+      tokenSoATrackGPU_(consumes<PixelTrackSoA>(iConfig.getParameter<edm::InputTag>("pixelTrackSrcGPU"))),
       topFolderName_(iConfig.getParameter<std::string>("topFolderName")),
       useQualityCut_(iConfig.getParameter<bool>("useQualityCut")),
       minQuality_(pixelTrackSoA::qualityByName(iConfig.getParameter<std::string>("minQuality"))),
@@ -126,12 +128,13 @@ SiPixelPhase1CompareTrackSoA::SiPixelPhase1CompareTrackSoA(const edm::ParameterS
 //
 // -- Analyze
 //
-void SiPixelPhase1CompareTrackSoA::analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) {
-  using helper = tracksUtilities<pixelTopology::Phase1>;
+template <typename T>
+void SiPixelCompareTrackSoA<T>::analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) {
+  using helper = tracksUtilities<T>;
   const auto& tsoaHandleCPU = iEvent.getHandle(tokenSoATrackCPU_);
   const auto& tsoaHandleGPU = iEvent.getHandle(tokenSoATrackGPU_);
   if (not tsoaHandleCPU or not tsoaHandleGPU) {
-    edm::LogWarning out("SiPixelPhase1CompareTrackSoA");
+    edm::LogWarning out("SiPixelCompareTrackSoA");
     if (not tsoaHandleCPU) {
       out << "reference (cpu) tracks not found; ";
     }
@@ -170,7 +173,6 @@ void SiPixelPhase1CompareTrackSoA::analyze(const edm::Event& iEvent, const edm::
 
   //Now loop over CPU tracks//nested loop for loose gPU tracks
   for (int32_t it = 0; it < maxTracksCPU; ++it) {
-
     float chi2CPU = tsoaCPU.view()[it].chi2();
     int nHitsCPU = helper::nHits(tsoaCPU.view(), it);
 
@@ -246,9 +248,10 @@ void SiPixelPhase1CompareTrackSoA::analyze(const edm::Event& iEvent, const edm::
 //
 // -- Book Histograms
 //
-void SiPixelPhase1CompareTrackSoA::bookHistograms(DQMStore::IBooker& iBook,
-                                                  edm::Run const& iRun,
-                                                  edm::EventSetup const& iSetup) {
+template <typename T>
+void SiPixelCompareTrackSoA<T>::bookHistograms(DQMStore::IBooker& iBook,
+                                               edm::Run const& iRun,
+                                               edm::EventSetup const& iSetup) {
   iBook.cd();
   iBook.setCurrentFolder(topFolderName_);
 
@@ -289,7 +292,8 @@ void SiPixelPhase1CompareTrackSoA::bookHistograms(DQMStore::IBooker& iBook,
 
 }
 
-void SiPixelPhase1CompareTrackSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+template<typename T>
+void SiPixelCompareTrackSoA<T>::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
   // monitorpixelTrackSoA
   edm::ParameterSetDescription desc;
   desc.add<edm::InputTag>("pixelTrackSrcCPU", edm::InputTag("pixelTracksSoA@cpu"));
@@ -300,4 +304,9 @@ void SiPixelPhase1CompareTrackSoA::fillDescriptions(edm::ConfigurationDescriptio
   desc.add<double>("deltaR2cut", 0.04);
   descriptions.addWithDefaultLabel(desc);
 }
+
+using SiPixelPhase1CompareTrackSoA = SiPixelCompareTrackSoA<pixelTopology::Phase1>;
+using SiPixelPhase2CompareTrackSoA = SiPixelCompareTrackSoA<pixelTopology::Phase2>;
+
 DEFINE_FWK_MODULE(SiPixelPhase1CompareTrackSoA);
+DEFINE_FWK_MODULE(SiPixelPhase2CompareTrackSoA);

DQM/SiPixelHeterogeneous/plugins/SiPixelCompareVertexSoA.cc

@@ -1,8 +1,8 @@
 // -*- C++ -*-
-// Package:    SiPixelPhase1CompareVertexSoA
-// Class:      SiPixelPhase1CompareVertexSoA
+// Package:    SiPixelCompareVertexSoA
+// Class:      SiPixelCompareVertexSoA
 //
-/**\class SiPixelPhase1CompareVertexSoA SiPixelPhase1CompareVertexSoA.cc
+/**\class SiPixelCompareVertexSoA SiPixelCompareVertexSoA.cc 
 */
 //
 // Author: Suvankar Roy Chowdhury
@@ -21,11 +21,11 @@
 #include "CUDADataFormats/Vertex/interface/ZVertexSoAHeterogeneousHost.h"
 #include "DataFormats/BeamSpot/interface/BeamSpot.h"
 
-class SiPixelPhase1CompareVertexSoA : public DQMEDAnalyzer {
+class SiPixelCompareVertexSoA : public DQMEDAnalyzer {
 public:
   using IndToEdm = std::vector<uint16_t>;
-  explicit SiPixelPhase1CompareVertexSoA(const edm::ParameterSet&);
-  ~SiPixelPhase1CompareVertexSoA() override = default;
+  explicit SiPixelCompareVertexSoA(const edm::ParameterSet&);
+  ~SiPixelCompareVertexSoA() override = default;
   void bookHistograms(DQMStore::IBooker& ibooker, edm::Run const& iRun, edm::EventSetup const& iSetup) override;
   void analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) override;
   static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
@@ -54,21 +54,21 @@ class SiPixelPhase1CompareVertexSoA : public DQMEDAnalyzer {
 //
 
 // Note tokenSoAVertexGPU_ contains data copied from device to host, hence is a HostCollection
-SiPixelPhase1CompareVertexSoA::SiPixelPhase1CompareVertexSoA(const edm::ParameterSet& iConfig)
-    : tokenSoAVertexCPU_(consumes(iConfig.getParameter<edm::InputTag>("pixelVertexSrcCPU"))),
-      tokenSoAVertexGPU_(consumes(iConfig.getParameter<edm::InputTag>("pixelVertexSrcGPU"))),
+SiPixelCompareVertexSoA::SiPixelCompareVertexSoA(const edm::ParameterSet& iConfig)
+    : tokenSoAVertexCPU_(consumes<zVertex::ZVertexSoAHost>(iConfig.getParameter<edm::InputTag>("pixelVertexSrcCPU"))),
+      tokenSoAVertexGPU_(consumes<zVertex::ZVertexSoAHost>(iConfig.getParameter<edm::InputTag>("pixelVertexSrcGPU"))),
       tokenBeamSpot_(consumes<reco::BeamSpot>(iConfig.getParameter<edm::InputTag>("beamSpotSrc"))),
       topFolderName_(iConfig.getParameter<std::string>("topFolderName")),
       dzCut_(iConfig.getParameter<double>("dzCut")) {}
 
 //
 // -- Analyze
 //
-void SiPixelPhase1CompareVertexSoA::analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) {
+void SiPixelCompareVertexSoA::analyze(const edm::Event& iEvent, const edm::EventSetup& iSetup) {
   const auto& vsoaHandleCPU = iEvent.getHandle(tokenSoAVertexCPU_);
   const auto& vsoaHandleGPU = iEvent.getHandle(tokenSoAVertexGPU_);
   if (not vsoaHandleCPU or not vsoaHandleGPU) {
-    edm::LogWarning out("SiPixelPhase1CompareTrackSoA");
+    edm::LogWarning out("SiPixelCompareVertexSoA");
     if (not vsoaHandleCPU) {
       out << "reference (cpu) tracks not found; ";
     }
@@ -87,7 +87,7 @@ void SiPixelPhase1CompareVertexSoA::analyze(const edm::Event& iEvent, const edm:
   auto bsHandle = iEvent.getHandle(tokenBeamSpot_);
   float x0 = 0., y0 = 0., z0 = 0., dxdz = 0., dydz = 0.;
   if (!bsHandle.isValid()) {
-    edm::LogWarning("PixelVertexProducer") << "No beamspot found. returning vertexes with (0,0,Z) ";
+    edm::LogWarning("SiPixelCompareVertexSoA") << "No beamspot found. returning vertexes with (0,0,Z) ";
   } else {
     const reco::BeamSpot& bs = *bsHandle;
     x0 = bs.x0();
@@ -150,29 +150,29 @@ void SiPixelPhase1CompareVertexSoA::analyze(const edm::Event& iEvent, const edm:
 //
 // -- Book Histograms
 //
-void SiPixelPhase1CompareVertexSoA::bookHistograms(DQMStore::IBooker& ibooker,
-                                                   edm::Run const& iRun,
-                                                   edm::EventSetup const& iSetup) {
+void SiPixelCompareVertexSoA::bookHistograms(DQMStore::IBooker& ibooker,
+                                             edm::Run const& iRun,
+                                             edm::EventSetup const& iSetup) {
   ibooker.cd();
   ibooker.setCurrentFolder(topFolderName_);
 
   // FIXME: all the 2D correlation plots are quite heavy in terms of memory consumption, so a as soon as DQM supports either TH2I or THnSparse
   // these should be moved to a less resource consuming format
-  hnVertex_ = ibooker.book2I("nVertex", "# of Vertex;CPU;GPU", 101, -0.5, 100.5, 101, -0.5, 100.5);
+  hnVertex_ = ibooker.book2I("nVertex", "# of Vertices;CPU;GPU", 101, -0.5, 100.5, 101, -0.5, 100.5);
   hx_ = ibooker.book2I("vx", "Vertez x - Beamspot x;CPU;GPU", 50, -0.1, 0.1, 50, -0.1, 0.1);
   hy_ = ibooker.book2I("vy", "Vertez y - Beamspot y;CPU;GPU", 50, -0.1, 0.1, 50, -0.1, 0.1);
   hz_ = ibooker.book2I("vz", "Vertez z;CPU;GPU", 30, -30., 30., 30, -30., 30.);
   hchi2_ = ibooker.book2I("chi2", "Vertex chi-squared;CPU;GPU", 40, 0., 20., 40, 0., 20.);
   hchi2oNdof_ = ibooker.book2I("chi2oNdof", "Vertex chi-squared/Ndof;CPU;GPU", 40, 0., 20., 40, 0., 20.);
-  hptv2_ = ibooker.book2I("ptsq", "Vertex p_T squared;CPU;GPU", 200, 0., 200., 200, 0., 200.);
+  hptv2_ = ibooker.book2I("ptsq", "Vertex #sum (p_{T})^{2};CPU;GPU", 200, 0., 200., 200, 0., 200.);
   hntrks_ = ibooker.book2I("ntrk", "#tracks associated;CPU;GPU", 100, -0.5, 99.5, 100, -0.5, 99.5);
   hntrks_ = ibooker.book2I("ntrk", "#tracks associated;CPU;GPU", 100, -0.5, 99.5, 100, -0.5, 99.5);
   hxdiff_ = ibooker.book1D("vxdiff", ";Vertex x difference (CPU - GPU);#entries", 100, -0.001, 0.001);
   hydiff_ = ibooker.book1D("vydiff", ";Vertex y difference (CPU - GPU);#entries", 100, -0.001, 0.001);
   hzdiff_ = ibooker.book1D("vzdiff", ";Vertex z difference (CPU - GPU);#entries", 100, -2.5, 2.5);
 }
 
-void SiPixelPhase1CompareVertexSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+void SiPixelCompareVertexSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
   // monitorpixelVertexSoA
   edm::ParameterSetDescription desc;
   desc.add<edm::InputTag>("pixelVertexSrcCPU", edm::InputTag("pixelVerticesSoA@cpu"));
@@ -182,4 +182,7 @@ void SiPixelPhase1CompareVertexSoA::fillDescriptions(edm::ConfigurationDescripti
   desc.add<double>("dzCut", 1.);
   descriptions.addWithDefaultLabel(desc);
 }
+DEFINE_FWK_MODULE(SiPixelCompareVertexSoA);
+
+using SiPixelPhase1CompareVertexSoA = SiPixelCompareVertexSoA;
 DEFINE_FWK_MODULE(SiPixelPhase1CompareVertexSoA);

DQM/SiPixelHeterogeneous/plugins/SiPixelMonitorTrackSoA.cc

@@ -80,7 +80,7 @@ void SiPixelPhase1MonitorTrackSoA::analyze(const edm::Event& iEvent, const edm::
     return;
   }
 
-  using helper = tracksUtilities<pixelTopology::Phase1>;
+  using helper = tracksUtilities<T>;
   auto& tsoa = *tsoaHandle.product();
   auto maxTracks = tsoa.view().metadata().size();
   auto const* quality = tsoa.view().quality();