From 566baeddcdc041b43fecf07c5aaffcebd260e4d8 Mon Sep 17 00:00:00 2001 From: Angela Czirkos Date: Tue, 6 Apr 2021 09:22:08 +0200 Subject: [PATCH] code-format --- .../plugins/Phase2TrackerRecHits.cc | 26 +- .../plugins/SiPixelClusterThresholds.h | 6 +- .../plugins/SiPixelDigisClustersFromSoA.cc | 93 +- .../plugins/SiPixelRawToClusterCUDA.cc | 136 +- .../plugins/SiPixelRawToClusterGPUKernel.cu | 1176 ++++++++--------- .../plugins/SiPixelRawToClusterGPUKernel.h | 367 +++-- .../plugins/gpuClusterChargeCut.h | 207 +-- .../SiPixelClusterizer/test/gpuClustering_t.h | 142 +- .../src/PixelCPEClusterRepair.cc | 320 +++-- .../interface/PixelClusterSelectorTopBottom.h | 13 +- .../interface/StripClusterSelectorTopBottom.h | 13 +- .../src/HITrackClusterRemover.cc | 320 ++--- .../src/HLTTrackClusterRemoverNew.cc | 254 ++-- .../src/JetCoreClusterSplitter.cc | 309 ++--- .../src/PixelClusterSelectorTopBottom.cc | 15 +- .../src/SeedClusterRemover.cc | 160 +-- .../src/SeedClusterRemoverPhase2.cc | 141 +- .../src/StripClusterSelectorTopBottom.cc | 20 +- 18 files changed, 1727 insertions(+), 1991 deletions(-) diff --git a/RecoLocalTracker/Phase2TrackerRecHits/plugins/Phase2TrackerRecHits.cc b/RecoLocalTracker/Phase2TrackerRecHits/plugins/Phase2TrackerRecHits.cc index d3d54e97c2f3f..c1ed80c795d0a 100644 --- a/RecoLocalTracker/Phase2TrackerRecHits/plugins/Phase2TrackerRecHits.cc +++ b/RecoLocalTracker/Phase2TrackerRecHits/plugins/Phase2TrackerRecHits.cc @@ -24,17 +24,14 @@ #include "RecoLocalTracker/Records/interface/TkPhase2OTCPERecord.h" class Phase2TrackerRecHits : public edm::global::EDProducer<> { - public: +public: explicit Phase2TrackerRecHits(const edm::ParameterSet& conf); ~Phase2TrackerRecHits() override{}; - void produce(edm::StreamID sid, edm::Event& event, - const edm::EventSetup& eventSetup) const final; + void produce(edm::StreamID sid, edm::Event& event, const edm::EventSetup& eventSetup) const final; - private: - edm::ESGetToken const - tTrackerGeom_; - edm::ESGetToken, - TkPhase2OTCPERecord> const tCPE_; +private: + edm::ESGetToken const tTrackerGeom_; + edm::ESGetToken, TkPhase2OTCPERecord> const tCPE_; edm::EDGetTokenT token_; }; @@ -42,13 +39,11 @@ class Phase2TrackerRecHits : public edm::global::EDProducer<> { Phase2TrackerRecHits::Phase2TrackerRecHits(edm::ParameterSet const& conf) : tTrackerGeom_(esConsumes()), tCPE_(esConsumes(conf.getParameter("Phase2StripCPE"))), - token_(consumes( - conf.getParameter("src"))) { + token_(consumes(conf.getParameter("src"))) { produces(); } -void Phase2TrackerRecHits::produce(edm::StreamID sid, edm::Event& event, - const edm::EventSetup& eventSetup) const { +void Phase2TrackerRecHits::produce(edm::StreamID sid, edm::Event& event, const edm::EventSetup& eventSetup) const { // Get the Clusters edm::Handle clusters; event.getByToken(token_, clusters); @@ -70,16 +65,15 @@ void Phase2TrackerRecHits::produce(edm::StreamID sid, edm::Event& event, const GeomDetUnit* geomDetUnit(tkGeom->idToDetUnit(detId)); // Container for the clusters that will be produced for this modules - Phase2TrackerRecHit1DCollectionNew::FastFiller rechits( - *outputRecHits, clusterDetSet.detId()); + Phase2TrackerRecHit1DCollectionNew::FastFiller rechits(*outputRecHits, clusterDetSet.detId()); for (const auto& clusterRef : clusterDetSet) { ClusterParameterEstimator::LocalValues lv = cpe->localParameters(clusterRef, *geomDetUnit); // Create a persistent edm::Ref to the cluster - edm::Ref - cluster = edmNew::makeRefTo(clusters, &clusterRef); + edm::Ref cluster = + edmNew::makeRefTo(clusters, &clusterRef); // Make a RecHit and add it to the DetSet Phase2TrackerRecHit1D hit(lv.first, lv.second, *geomDetUnit, cluster); diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterThresholds.h b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterThresholds.h index 93d4ccc4c0014..41dd430262f73 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterThresholds.h +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterThresholds.h @@ -2,15 +2,13 @@ #define RecoLocalTracker_SiPixelClusterizer_plugins_SiPixelClusterThresholds_h struct SiPixelClusterThresholds { - inline constexpr int32_t getThresholdForLayerOnCondition( - bool isLayer1) const noexcept { + inline constexpr int32_t getThresholdForLayerOnCondition(bool isLayer1) const noexcept { return isLayer1 ? layer1 : otherLayers; } const int32_t layer1; const int32_t otherLayers; }; -constexpr SiPixelClusterThresholds kSiPixelClusterThresholdsDefaultPhase1{ - .layer1 = 2000, .otherLayers = 4000}; +constexpr SiPixelClusterThresholds kSiPixelClusterThresholdsDefaultPhase1{.layer1 = 2000, .otherLayers = 4000}; #endif // RecoLocalTracker_SiPixelClusterizer_plugins_SiPixelClusterThresholds_h diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc index 439b310fa48f6..b6484dbea915e 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc @@ -21,15 +21,14 @@ #include "SiPixelClusterThresholds.h" class SiPixelDigisClustersFromSoA : public edm::global::EDProducer<> { - public: +public: explicit SiPixelDigisClustersFromSoA(const edm::ParameterSet& iConfig); ~SiPixelDigisClustersFromSoA() override = default; static void fillDescriptions(edm::ConfigurationDescriptions& descriptions); - private: - void produce(edm::StreamID, edm::Event& iEvent, - const edm::EventSetup& iSetup) const override; +private: + void produce(edm::StreamID, edm::Event& iEvent, const edm::EventSetup& iSetup) const override; const edm::ESGetToken topoToken_; @@ -38,34 +37,26 @@ class SiPixelDigisClustersFromSoA : public edm::global::EDProducer<> { edm::EDPutTokenT> digiPutToken_; edm::EDPutTokenT clusterPutToken_; - const SiPixelClusterThresholds - clusterThresholds_; // Cluster threshold in electrons + const SiPixelClusterThresholds clusterThresholds_; // Cluster threshold in electrons }; -SiPixelDigisClustersFromSoA::SiPixelDigisClustersFromSoA( - const edm::ParameterSet& iConfig) +SiPixelDigisClustersFromSoA::SiPixelDigisClustersFromSoA(const edm::ParameterSet& iConfig) : topoToken_(esConsumes()), - digiGetToken_(consumes( - iConfig.getParameter("src"))), + digiGetToken_(consumes(iConfig.getParameter("src"))), digiPutToken_(produces>()), clusterPutToken_(produces()), - clusterThresholds_{ - iConfig.getParameter("clusterThreshold_layer1"), - iConfig.getParameter("clusterThreshold_otherLayers")} {} + clusterThresholds_{iConfig.getParameter("clusterThreshold_layer1"), + iConfig.getParameter("clusterThreshold_otherLayers")} {} -void SiPixelDigisClustersFromSoA::fillDescriptions( - edm::ConfigurationDescriptions& descriptions) { +void SiPixelDigisClustersFromSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) { edm::ParameterSetDescription desc; desc.add("src", edm::InputTag("siPixelDigisSoA")); - desc.add("clusterThreshold_layer1", - kSiPixelClusterThresholdsDefaultPhase1.layer1); - desc.add("clusterThreshold_otherLayers", - kSiPixelClusterThresholdsDefaultPhase1.otherLayers); + desc.add("clusterThreshold_layer1", kSiPixelClusterThresholdsDefaultPhase1.layer1); + desc.add("clusterThreshold_otherLayers", kSiPixelClusterThresholdsDefaultPhase1.otherLayers); descriptions.addWithDefaultLabel(desc); } -void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, - const edm::EventSetup& iSetup) const { +void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, const edm::EventSetup& iSetup) const { const auto& digis = iEvent.get(digiGetToken_); const uint32_t nDigis = digis.size(); const auto& ttopo = iSetup.getData(topoToken_); @@ -76,7 +67,8 @@ void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, edm::DetSet* detDigis = nullptr; for (uint32_t i = 0; i < nDigis; i++) { - if (digis.pdigi(i) == 0) continue; + if (digis.pdigi(i) == 0) + continue; detDigis = &collection->find_or_insert(digis.rawIdArr(i)); if ((*detDigis).empty()) (*detDigis).data.reserve(64); // avoid the first relocations @@ -84,54 +76,50 @@ void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, } int32_t nclus = -1; - std::vector aclusters( - gpuClustering::maxNumClustersPerModules); + std::vector aclusters(gpuClustering::maxNumClustersPerModules); #ifdef EDM_ML_DEBUG auto totClustersFilled = 0; #endif auto fillClusters = [&](uint32_t detId) { - if (nclus < 0) return; // this in reality should never happen - edmNew::DetSetVector::FastFiller spc(*outputClusters, - detId); + if (nclus < 0) + return; // this in reality should never happen + edmNew::DetSetVector::FastFiller spc(*outputClusters, detId); auto layer = (DetId(detId).subdetId() == 1) ? ttopo.pxbLayer(detId) : 0; - auto clusterThreshold = - clusterThresholds_.getThresholdForLayerOnCondition(layer == 1); + auto clusterThreshold = clusterThresholds_.getThresholdForLayerOnCondition(layer == 1); for (int32_t ic = 0; ic < nclus + 1; ++ic) { auto const& acluster = aclusters[ic]; // in any case we cannot go out of sync with gpu... if (acluster.charge < clusterThreshold) - edm::LogWarning("SiPixelDigisClustersFromSoA") - << "cluster below charge Threshold " - << "Layer/DetId/clusId " << layer << '/' << detId << '/' << ic - << " size/charge " << acluster.isize << '/' << acluster.charge; - SiPixelCluster cluster(acluster.isize, acluster.adc, acluster.x, - acluster.y, acluster.xmin, acluster.ymin, ic); + edm::LogWarning("SiPixelDigisClustersFromSoA") << "cluster below charge Threshold " + << "Layer/DetId/clusId " << layer << '/' << detId << '/' << ic + << " size/charge " << acluster.isize << '/' << acluster.charge; + SiPixelCluster cluster(acluster.isize, acluster.adc, acluster.x, acluster.y, acluster.xmin, acluster.ymin, ic); #ifdef EDM_ML_DEBUG ++totClustersFilled; #endif LogDebug("SiPixelDigisClustersFromSoA") - << "putting in this cluster " << ic << " " << cluster.charge() << " " - << cluster.pixelADC().size(); + << "putting in this cluster " << ic << " " << cluster.charge() << " " << cluster.pixelADC().size(); // sort by row (x) spc.push_back(std::move(cluster)); - std::push_heap(spc.begin(), spc.end(), - [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { - return cl1.minPixelRow() < cl2.minPixelRow(); - }); + std::push_heap(spc.begin(), spc.end(), [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { + return cl1.minPixelRow() < cl2.minPixelRow(); + }); } - for (int32_t ic = 0; ic < nclus + 1; ++ic) aclusters[ic].clear(); + for (int32_t ic = 0; ic < nclus + 1; ++ic) + aclusters[ic].clear(); nclus = -1; // sort by row (x) - std::sort_heap(spc.begin(), spc.end(), - [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { - return cl1.minPixelRow() < cl2.minPixelRow(); - }); - if (spc.empty()) spc.abort(); + std::sort_heap(spc.begin(), spc.end(), [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { + return cl1.minPixelRow() < cl2.minPixelRow(); + }); + if (spc.empty()) + spc.abort(); }; for (uint32_t i = 0; i < nDigis; i++) { - if (digis.pdigi(i) == 0) continue; + if (digis.pdigi(i) == 0) + continue; if (digis.clus(i) > 9000) continue; // not in cluster; TODO add an assert for the size assert(digis.rawIdArr(i) > 109999); @@ -142,8 +130,7 @@ void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, if ((*detDigis).empty()) (*detDigis).data.reserve(64); // avoid the first relocations else { - edm::LogWarning("SiPixelDigisClustersFromSoA") - << "Problem det present twice in input! " << (*detDigis).detId(); + edm::LogWarning("SiPixelDigisClustersFromSoA") << "Problem det present twice in input! " << (*detDigis).detId(); } } (*detDigis).data.emplace_back(digis.pdigi(i)); @@ -159,10 +146,10 @@ void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, } // fill final clusters - if (detDigis) fillClusters((*detDigis).detId()); + if (detDigis) + fillClusters((*detDigis).detId()); #ifdef EDM_ML_DEBUG - LogDebug("SiPixelDigisClustersFromSoA") - << "filled " << totClustersFilled << " clusters"; + LogDebug("SiPixelDigisClustersFromSoA") << "filled " << totClustersFilled << " clusters"; #endif iEvent.put(digiPutToken_, std::move(collection)); diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterCUDA.cc b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterCUDA.cc index 1ee05eb42ccf3..690bae6886fe0 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterCUDA.cc +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterCUDA.cc @@ -40,36 +40,31 @@ #include "SiPixelClusterThresholds.h" #include "SiPixelRawToClusterGPUKernel.h" -class SiPixelRawToClusterCUDA - : public edm::stream::EDProducer { - public: +class SiPixelRawToClusterCUDA : public edm::stream::EDProducer { +public: explicit SiPixelRawToClusterCUDA(const edm::ParameterSet& iConfig); ~SiPixelRawToClusterCUDA() override = default; static void fillDescriptions(edm::ConfigurationDescriptions& descriptions); - private: - void acquire(const edm::Event& iEvent, const edm::EventSetup& iSetup, +private: + void acquire(const edm::Event& iEvent, + const edm::EventSetup& iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) override; void produce(edm::Event& iEvent, const edm::EventSetup& iSetup) override; edm::EDGetTokenT rawGetToken_; edm::EDPutTokenT> digiPutToken_; - edm::EDPutTokenT> - digiErrorPutToken_; + edm::EDPutTokenT> digiErrorPutToken_; edm::EDPutTokenT> clusterPutToken_; cms::cuda::ContextState ctxState_; edm::ESWatcher recordWatcher_; - edm::ESGetToken - gpuMapToken_; - edm::ESGetToken - gainsToken_; - edm::ESGetToken - cablingMapToken_; + edm::ESGetToken gpuMapToken_; + edm::ESGetToken gainsToken_; + edm::ESGetToken cablingMapToken_; std::unique_ptr cabling_; std::vector fedIds_; @@ -77,9 +72,7 @@ class SiPixelRawToClusterCUDA std::unique_ptr regions_; pixelgpudetails::SiPixelRawToClusterGPUKernel gpuAlgo_; - std::unique_ptr< - pixelgpudetails::SiPixelRawToClusterGPUKernel::WordFedAppender> - wordFedAppender_; + std::unique_ptr wordFedAppender_; PixelDataFormatter::Errors errors_; const bool isRun2_; @@ -88,55 +81,41 @@ class SiPixelRawToClusterCUDA const SiPixelClusterThresholds clusterThresholds_; }; -SiPixelRawToClusterCUDA::SiPixelRawToClusterCUDA( - const edm::ParameterSet& iConfig) - : rawGetToken_(consumes( - iConfig.getParameter("InputLabel"))), +SiPixelRawToClusterCUDA::SiPixelRawToClusterCUDA(const edm::ParameterSet& iConfig) + : rawGetToken_(consumes(iConfig.getParameter("InputLabel"))), digiPutToken_(produces>()), clusterPutToken_(produces>()), - gpuMapToken_(esConsumes()), - gainsToken_(esConsumes()), - cablingMapToken_( - esConsumes( - edm::ESInputTag( - "", iConfig.getParameter("CablingMapLabel")))), + gpuMapToken_(esConsumes()), + gainsToken_(esConsumes()), + cablingMapToken_(esConsumes( + edm::ESInputTag("", iConfig.getParameter("CablingMapLabel")))), isRun2_(iConfig.getParameter("isRun2")), includeErrors_(iConfig.getParameter("IncludeErrors")), useQuality_(iConfig.getParameter("UseQualityInfo")), - clusterThresholds_{ - iConfig.getParameter("clusterThreshold_layer1"), - iConfig.getParameter("clusterThreshold_otherLayers")} { + clusterThresholds_{iConfig.getParameter("clusterThreshold_layer1"), + iConfig.getParameter("clusterThreshold_otherLayers")} { if (includeErrors_) { digiErrorPutToken_ = produces>(); } // regions - if (!iConfig.getParameter("Regions") - .getParameterNames() - .empty()) { - regions_ = - std::make_unique(iConfig, consumesCollector()); + if (!iConfig.getParameter("Regions").getParameterNames().empty()) { + regions_ = std::make_unique(iConfig, consumesCollector()); } edm::Service cs; if (cs->enabled()) { - wordFedAppender_ = std::make_unique< - pixelgpudetails::SiPixelRawToClusterGPUKernel::WordFedAppender>(); + wordFedAppender_ = std::make_unique(); } } -void SiPixelRawToClusterCUDA::fillDescriptions( - edm::ConfigurationDescriptions& descriptions) { +void SiPixelRawToClusterCUDA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) { edm::ParameterSetDescription desc; desc.add("isRun2", true); desc.add("IncludeErrors", true); desc.add("UseQualityInfo", false); - desc.add("clusterThreshold_layer1", - kSiPixelClusterThresholdsDefaultPhase1.layer1); - desc.add("clusterThreshold_otherLayers", - kSiPixelClusterThresholdsDefaultPhase1.otherLayers); + desc.add("clusterThreshold_layer1", kSiPixelClusterThresholdsDefaultPhase1.layer1); + desc.add("clusterThreshold_otherLayers", kSiPixelClusterThresholdsDefaultPhase1.otherLayers); desc.add("InputLabel", edm::InputTag("rawDataCollector")); { edm::ParameterSetDescription psd0; @@ -147,23 +126,20 @@ void SiPixelRawToClusterCUDA::fillDescriptions( desc.add("Regions", psd0) ->setComment("## Empty Regions PSet means complete unpacking"); } - desc.add("CablingMapLabel", "") - ->setComment("CablingMap label"); // Tav + desc.add("CablingMapLabel", "")->setComment("CablingMap label"); // Tav descriptions.addWithDefaultLabel(desc); } -void SiPixelRawToClusterCUDA::acquire( - const edm::Event& iEvent, const edm::EventSetup& iSetup, - edm::WaitingTaskWithArenaHolder waitingTaskHolder) { - cms::cuda::ScopedContextAcquire ctx{iEvent.streamID(), - std::move(waitingTaskHolder), ctxState_}; +void SiPixelRawToClusterCUDA::acquire(const edm::Event& iEvent, + const edm::EventSetup& iSetup, + edm::WaitingTaskWithArenaHolder waitingTaskHolder) { + cms::cuda::ScopedContextAcquire ctx{iEvent.streamID(), std::move(waitingTaskHolder), ctxState_}; auto hgpuMap = iSetup.getHandle(gpuMapToken_); if (hgpuMap->hasQuality() != useQuality_) { - throw cms::Exception("LogicError") - << "UseQuality of the module (" << useQuality_ - << ") differs the one from SiPixelROCsStatusAndMappingWrapper. Please " - "fix your configuration."; + throw cms::Exception("LogicError") << "UseQuality of the module (" << useQuality_ + << ") differs the one from SiPixelROCsStatusAndMappingWrapper. Please " + "fix your configuration."; } // get the GPU product already here so that the async transfer can begin const auto* gpuMap = hgpuMap->getGPUProductAsync(ctx.stream()); @@ -177,14 +153,10 @@ void SiPixelRawToClusterCUDA::acquire( if (regions_) { regions_->run(iEvent, iSetup); - LogDebug("SiPixelRawToCluster") - << "region2unpack #feds: " << regions_->nFEDs(); - LogDebug("SiPixelRawToCluster") - << "region2unpack #modules (BPIX,EPIX,total): " - << regions_->nBarrelModules() << " " << regions_->nForwardModules() - << " " << regions_->nModules(); - modulesToUnpackRegional = hgpuMap->getModToUnpRegionalAsync( - *(regions_->modulesToUnpack()), ctx.stream()); + LogDebug("SiPixelRawToCluster") << "region2unpack #feds: " << regions_->nFEDs(); + LogDebug("SiPixelRawToCluster") << "region2unpack #modules (BPIX,EPIX,total): " << regions_->nBarrelModules() << " " + << regions_->nForwardModules() << " " << regions_->nModules(); + modulesToUnpackRegional = hgpuMap->getModToUnpRegionalAsync(*(regions_->modulesToUnpack()), ctx.stream()); gpuModulesToUnpack = modulesToUnpackRegional.get(); } else { gpuModulesToUnpack = hgpuMap->getModToUnpAllAsync(ctx.stream()); @@ -214,7 +186,8 @@ void SiPixelRawToClusterCUDA::acquire( // PixelDataFormatter::interpretRawData() ErrorChecker errorcheck; for (int fedId : fedIds_) { - if (regions_ && !regions_->mayUnpackFED(fedId)) continue; + if (regions_ && !regions_->mayUnpackFED(fedId)) + continue; // for GPU // first 150 index stores the fedId and next 150 will store the @@ -232,21 +205,18 @@ void SiPixelRawToClusterCUDA::acquire( } // check CRC bit - const cms_uint64_t* trailer = - reinterpret_cast(rawData.data()) + (nWords - 1); + const cms_uint64_t* trailer = reinterpret_cast(rawData.data()) + (nWords - 1); if (not errorcheck.checkCRC(errorsInEvent, fedId, trailer, errors_)) { continue; } // check headers - const cms_uint64_t* header = - reinterpret_cast(rawData.data()); + const cms_uint64_t* header = reinterpret_cast(rawData.data()); header--; bool moreHeaders = true; while (moreHeaders) { header++; - bool headerStatus = - errorcheck.checkHeader(errorsInEvent, fedId, header, errors_); + bool headerStatus = errorcheck.checkHeader(errorsInEvent, fedId, header, errors_); moreHeaders = headerStatus; } @@ -255,8 +225,7 @@ void SiPixelRawToClusterCUDA::acquire( trailer++; while (moreTrailers) { trailer--; - bool trailerStatus = errorcheck.checkTrailer(errorsInEvent, fedId, nWords, - trailer, errors_); + bool trailerStatus = errorcheck.checkTrailer(errorsInEvent, fedId, nWords, trailer, errors_); moreTrailers = trailerStatus; } @@ -269,15 +238,22 @@ void SiPixelRawToClusterCUDA::acquire( } // end of for loop - gpuAlgo_.makeClustersAsync( - isRun2_, clusterThresholds_, gpuMap, gpuModulesToUnpack, gpuGains, - *wordFedAppender_, std::move(errors_), wordCounterGPU, fedCounter, - useQuality_, includeErrors_, edm::MessageDrop::instance()->debugEnabled, - ctx.stream()); + gpuAlgo_.makeClustersAsync(isRun2_, + clusterThresholds_, + gpuMap, + gpuModulesToUnpack, + gpuGains, + *wordFedAppender_, + std::move(errors_), + wordCounterGPU, + fedCounter, + useQuality_, + includeErrors_, + edm::MessageDrop::instance()->debugEnabled, + ctx.stream()); } -void SiPixelRawToClusterCUDA::produce(edm::Event& iEvent, - const edm::EventSetup& iSetup) { +void SiPixelRawToClusterCUDA::produce(edm::Event& iEvent, const edm::EventSetup& iSetup) { cms::cuda::ScopedContextProduce ctx{ctxState_}; auto tmp = gpuAlgo_.getResults(); diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.cu b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.cu index a84cef8948c6f..95442d1f02d54 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.cu +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.cu @@ -34,647 +34,647 @@ namespace pixelgpudetails { -// number of words for all the FEDs -constexpr uint32_t MAX_FED_WORDS = - pixelgpudetails::MAX_FED * pixelgpudetails::MAX_WORD; - -SiPixelRawToClusterGPUKernel::WordFedAppender::WordFedAppender() { - word_ = cms::cuda::make_host_noncached_unique( - MAX_FED_WORDS, cudaHostAllocWriteCombined); - fedId_ = cms::cuda::make_host_noncached_unique( - MAX_FED_WORDS, cudaHostAllocWriteCombined); -} - -void SiPixelRawToClusterGPUKernel::WordFedAppender::initializeWordFed( - int fedId, unsigned int wordCounterGPU, const cms_uint32_t *src, - unsigned int length) { - std::memcpy(word_.get() + wordCounterGPU, src, sizeof(cms_uint32_t) * length); - std::memset(fedId_.get() + wordCounterGPU / 2, - fedId - FEDNumbering::MINSiPixeluTCAFEDID, length / 2); -} - -//////////////////// - -__device__ uint32_t getLink(uint32_t ww) { - return ((ww >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask); -} - -__device__ uint32_t getRoc(uint32_t ww) { - return ((ww >> pixelgpudetails::ROC_shift) & pixelgpudetails::ROC_mask); -} - -__device__ uint32_t getADC(uint32_t ww) { - return ((ww >> pixelgpudetails::ADC_shift) & pixelgpudetails::ADC_mask); -} - -__device__ bool isBarrel(uint32_t rawId) { - return (PixelSubdetector::PixelBarrel == - ((rawId >> DetId::kSubdetOffset) & DetId::kSubdetMask)); -} - -__device__ pixelgpudetails::DetIdGPU getRawId( - const SiPixelROCsStatusAndMapping *cablingMap, uint8_t fed, uint32_t link, - uint32_t roc) { - uint32_t index = fed * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + roc; - pixelgpudetails::DetIdGPU detId = {cablingMap->rawId[index], - cablingMap->rocInDet[index], - cablingMap->moduleId[index]}; - return detId; -} - -// reference -// http://cmsdoxygen.web.cern.ch/cmsdoxygen/CMSSW_9_2_0/doc/html/dd/d31/FrameConversion_8cc_source.html -// http://cmslxr.fnal.gov/source/CondFormats/SiPixelObjects/src/PixelROC.cc?v=CMSSW_9_2_0#0071 -// Convert local pixel to pixelgpudetails::global pixel -__device__ pixelgpudetails::Pixel frameConversion( - bool bpix, int side, uint32_t layer, uint32_t rocIdInDetUnit, - pixelgpudetails::Pixel local) { - int slopeRow = 0, slopeCol = 0; - int rowOffset = 0, colOffset = 0; - - if (bpix) { - if (side == -1 && layer != 1) { // -Z side: 4 non-flipped modules oriented - // like 'dddd', except Layer 1 - if (rocIdInDetUnit < 8) { - slopeRow = 1; - slopeCol = -1; - rowOffset = 0; - colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; - } else { - slopeRow = -1; - slopeCol = 1; - rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; - colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; - } // if roc - } else { // +Z side: 4 non-flipped modules oriented like 'pppp', but all 8 - // in layer1 - if (rocIdInDetUnit < 8) { - slopeRow = -1; - slopeCol = 1; - rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; - colOffset = rocIdInDetUnit * pixelgpudetails::numColsInRoc; - } else { - slopeRow = 1; - slopeCol = -1; - rowOffset = 0; - colOffset = (16 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; + // number of words for all the FEDs + constexpr uint32_t MAX_FED_WORDS = pixelgpudetails::MAX_FED * pixelgpudetails::MAX_WORD; + + SiPixelRawToClusterGPUKernel::WordFedAppender::WordFedAppender() { + word_ = cms::cuda::make_host_noncached_unique(MAX_FED_WORDS, cudaHostAllocWriteCombined); + fedId_ = cms::cuda::make_host_noncached_unique(MAX_FED_WORDS, cudaHostAllocWriteCombined); + } + + void SiPixelRawToClusterGPUKernel::WordFedAppender::initializeWordFed(int fedId, + unsigned int wordCounterGPU, + const cms_uint32_t *src, + unsigned int length) { + std::memcpy(word_.get() + wordCounterGPU, src, sizeof(cms_uint32_t) * length); + std::memset(fedId_.get() + wordCounterGPU / 2, fedId - FEDNumbering::MINSiPixeluTCAFEDID, length / 2); + } + + //////////////////// + + __device__ uint32_t getLink(uint32_t ww) { + return ((ww >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask); + } + + __device__ uint32_t getRoc(uint32_t ww) { return ((ww >> pixelgpudetails::ROC_shift) & pixelgpudetails::ROC_mask); } + + __device__ uint32_t getADC(uint32_t ww) { return ((ww >> pixelgpudetails::ADC_shift) & pixelgpudetails::ADC_mask); } + + __device__ bool isBarrel(uint32_t rawId) { + return (PixelSubdetector::PixelBarrel == ((rawId >> DetId::kSubdetOffset) & DetId::kSubdetMask)); + } + + __device__ pixelgpudetails::DetIdGPU getRawId(const SiPixelROCsStatusAndMapping *cablingMap, + uint8_t fed, + uint32_t link, + uint32_t roc) { + uint32_t index = fed * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + roc; + pixelgpudetails::DetIdGPU detId = { + cablingMap->rawId[index], cablingMap->rocInDet[index], cablingMap->moduleId[index]}; + return detId; + } + + // reference + // http://cmsdoxygen.web.cern.ch/cmsdoxygen/CMSSW_9_2_0/doc/html/dd/d31/FrameConversion_8cc_source.html + // http://cmslxr.fnal.gov/source/CondFormats/SiPixelObjects/src/PixelROC.cc?v=CMSSW_9_2_0#0071 + // Convert local pixel to pixelgpudetails::global pixel + __device__ pixelgpudetails::Pixel frameConversion( + bool bpix, int side, uint32_t layer, uint32_t rocIdInDetUnit, pixelgpudetails::Pixel local) { + int slopeRow = 0, slopeCol = 0; + int rowOffset = 0, colOffset = 0; + + if (bpix) { + if (side == -1 && layer != 1) { // -Z side: 4 non-flipped modules oriented + // like 'dddd', except Layer 1 + if (rocIdInDetUnit < 8) { + slopeRow = 1; + slopeCol = -1; + rowOffset = 0; + colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; + } else { + slopeRow = -1; + slopeCol = 1; + rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; + colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; + } // if roc + } else { // +Z side: 4 non-flipped modules oriented like 'pppp', but all 8 + // in layer1 + if (rocIdInDetUnit < 8) { + slopeRow = -1; + slopeCol = 1; + rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; + colOffset = rocIdInDetUnit * pixelgpudetails::numColsInRoc; + } else { + slopeRow = 1; + slopeCol = -1; + rowOffset = 0; + colOffset = (16 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; + } } + + } else { // fpix + if (side == -1) { // pannel 1 + if (rocIdInDetUnit < 8) { + slopeRow = 1; + slopeCol = -1; + rowOffset = 0; + colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; + } else { + slopeRow = -1; + slopeCol = 1; + rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; + colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; + } + } else { // pannel 2 + if (rocIdInDetUnit < 8) { + slopeRow = 1; + slopeCol = -1; + rowOffset = 0; + colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; + } else { + slopeRow = -1; + slopeCol = 1; + rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; + colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; + } + + } // side } - } else { // fpix - if (side == -1) { // pannel 1 - if (rocIdInDetUnit < 8) { - slopeRow = 1; - slopeCol = -1; - rowOffset = 0; - colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; - } else { - slopeRow = -1; - slopeCol = 1; - rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; - colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; + uint32_t gRow = rowOffset + slopeRow * local.row; + uint32_t gCol = colOffset + slopeCol * local.col; + // inside frameConversion row: gRow, column: gCol + pixelgpudetails::Pixel global = {gRow, gCol}; + return global; + } + + // error decoding and handling copied from + // EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc + __device__ uint8_t conversionError(uint8_t fedId, uint8_t status, bool debug = false) { + uint8_t errorType = 0; + + switch (status) { + case (1): { + if (debug) + printf("Error in Fed: %i, invalid channel Id (errorType = 35\n)", fedId); + errorType = 35; + break; } - } else { // pannel 2 - if (rocIdInDetUnit < 8) { - slopeRow = 1; - slopeCol = -1; - rowOffset = 0; - colOffset = (8 - rocIdInDetUnit) * pixelgpudetails::numColsInRoc - 1; - } else { - slopeRow = -1; - slopeCol = 1; - rowOffset = 2 * pixelgpudetails::numRowsInRoc - 1; - colOffset = (rocIdInDetUnit - 8) * pixelgpudetails::numColsInRoc; + case (2): { + if (debug) + printf("Error in Fed: %i, invalid ROC Id (errorType = 36)\n", fedId); + errorType = 36; + break; } + case (3): { + if (debug) + printf("Error in Fed: %i, invalid dcol/pixel value (errorType = 37)\n", fedId); + errorType = 37; + break; + } + case (4): { + if (debug) + printf("Error in Fed: %i, dcol/pixel read out of order (errorType = 38)\n", fedId); + errorType = 38; + break; + } + default: + if (debug) + printf("Cabling check returned unexpected result, status = %i\n", status); + }; - } // side + return errorType; } - uint32_t gRow = rowOffset + slopeRow * local.row; - uint32_t gCol = colOffset + slopeCol * local.col; - // inside frameConversion row: gRow, column: gCol - pixelgpudetails::Pixel global = {gRow, gCol}; - return global; -} - -// error decoding and handling copied from -// EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc -__device__ uint8_t conversionError(uint8_t fedId, uint8_t status, - bool debug = false) { - uint8_t errorType = 0; - - switch (status) { - case (1): { - if (debug) - printf("Error in Fed: %i, invalid channel Id (errorType = 35\n)", - fedId); - errorType = 35; - break; - } - case (2): { - if (debug) - printf("Error in Fed: %i, invalid ROC Id (errorType = 36)\n", fedId); - errorType = 36; - break; - } - case (3): { - if (debug) - printf("Error in Fed: %i, invalid dcol/pixel value (errorType = 37)\n", - fedId); - errorType = 37; - break; - } - case (4): { - if (debug) - printf( - "Error in Fed: %i, dcol/pixel read out of order (errorType = 38)\n", - fedId); - errorType = 38; - break; - } - default: - if (debug) - printf("Cabling check returned unexpected result, status = %i\n", - status); - }; - - return errorType; -} - -__device__ bool rocRowColIsValid(uint32_t rocRow, uint32_t rocCol) { - /// row and column in ROC representation - return ((rocRow < pixelgpudetails::numRowsInRoc) & - (rocCol < pixelgpudetails::numColsInRoc)); -} - -__device__ bool dcolIsValid(uint32_t dcol, uint32_t pxid) { - return ((dcol < 26) & (2 <= pxid) & (pxid < 162)); -} - -// error decoding and handling copied from -// EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc -__device__ uint8_t checkROC(uint32_t errorWord, uint8_t fedId, uint32_t link, - const SiPixelROCsStatusAndMapping *cablingMap, - bool debug = false) { - uint8_t errorType = - (errorWord >> pixelgpudetails::ROC_shift) & pixelgpudetails::ERROR_mask; - if (errorType < 25) return 0; - bool errorFound = false; - - switch (errorType) { - case (25): { - errorFound = true; - uint32_t index = fedId * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + 1; - if (index > 1 && index <= cablingMap->size) { - if (!(link == cablingMap->link[index] && 1 == cablingMap->roc[index])) - errorFound = false; + __device__ bool rocRowColIsValid(uint32_t rocRow, uint32_t rocCol) { + /// row and column in ROC representation + return ((rocRow < pixelgpudetails::numRowsInRoc) & (rocCol < pixelgpudetails::numColsInRoc)); + } + + __device__ bool dcolIsValid(uint32_t dcol, uint32_t pxid) { return ((dcol < 26) & (2 <= pxid) & (pxid < 162)); } + + // error decoding and handling copied from + // EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc + __device__ uint8_t checkROC(uint32_t errorWord, + uint8_t fedId, + uint32_t link, + const SiPixelROCsStatusAndMapping *cablingMap, + bool debug = false) { + uint8_t errorType = (errorWord >> pixelgpudetails::ROC_shift) & pixelgpudetails::ERROR_mask; + if (errorType < 25) + return 0; + bool errorFound = false; + + switch (errorType) { + case (25): { + errorFound = true; + uint32_t index = fedId * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + 1; + if (index > 1 && index <= cablingMap->size) { + if (!(link == cablingMap->link[index] && 1 == cablingMap->roc[index])) + errorFound = false; + } + if (debug and errorFound) + printf("Invalid ROC = 25 found (errorType = 25)\n"); + break; } - if (debug and errorFound) - printf("Invalid ROC = 25 found (errorType = 25)\n"); - break; - } - case (26): { - if (debug) printf("Gap word found (errorType = 26)\n"); - errorFound = true; - break; - } - case (27): { - if (debug) printf("Dummy word found (errorType = 27)\n"); - errorFound = true; - break; - } - case (28): { - if (debug) printf("Error fifo nearly full (errorType = 28)\n"); - errorFound = true; - break; - } - case (29): { - if (debug) printf("Timeout on a channel (errorType = 29)\n"); - if ((errorWord >> pixelgpudetails::OMIT_ERR_shift) & - pixelgpudetails::OMIT_ERR_mask) { + case (26): { if (debug) - printf("...first errorType=29 error, this gets masked out\n"); + printf("Gap word found (errorType = 26)\n"); + errorFound = true; + break; } - errorFound = true; - break; - } - case (30): { - if (debug) printf("TBM error trailer (errorType = 30)\n"); - int stateMatch_bits = 4; - int stateMatch_shift = 8; - uint32_t stateMatch_mask = ~(~uint32_t(0) << stateMatch_bits); - int stateMatch = (errorWord >> stateMatch_shift) & stateMatch_mask; - if (stateMatch != 1 && stateMatch != 8) { + case (27): { if (debug) - printf("FED error 30 with unexpected State Bits (errorType = 30)\n"); + printf("Dummy word found (errorType = 27)\n"); + errorFound = true; + break; } - if (stateMatch == 1) - errorType = 40; // 1=Overflow -> 40, 8=number of ROCs -> 30 - errorFound = true; - break; - } - case (31): { - if (debug) printf("Event number error (errorType = 31)\n"); - errorFound = true; - break; - } - default: - errorFound = false; - }; - - return errorFound ? errorType : 0; -} - -// error decoding and handling copied from -// EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc -__device__ uint32_t getErrRawID(uint8_t fedId, uint32_t errWord, - uint32_t errorType, - const SiPixelROCsStatusAndMapping *cablingMap, - bool debug = false) { - uint32_t rID = 0xffffffff; - - switch (errorType) { - case 25: - case 30: - case 31: - case 36: - case 40: { - uint32_t roc = 1; - uint32_t link = - (errWord >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask; - uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; - if (rID_temp != gpuClustering::invalidModuleId) rID = rID_temp; - break; - } - case 29: { - int chanNmbr = 0; - const int DB0_shift = 0; - const int DB1_shift = DB0_shift + 1; - const int DB2_shift = DB1_shift + 1; - const int DB3_shift = DB2_shift + 1; - const int DB4_shift = DB3_shift + 1; - const uint32_t DataBit_mask = ~(~uint32_t(0) << 1); - - int CH1 = (errWord >> DB0_shift) & DataBit_mask; - int CH2 = (errWord >> DB1_shift) & DataBit_mask; - int CH3 = (errWord >> DB2_shift) & DataBit_mask; - int CH4 = (errWord >> DB3_shift) & DataBit_mask; - int CH5 = (errWord >> DB4_shift) & DataBit_mask; - int BLOCK_bits = 3; - int BLOCK_shift = 8; - uint32_t BLOCK_mask = ~(~uint32_t(0) << BLOCK_bits); - int BLOCK = (errWord >> BLOCK_shift) & BLOCK_mask; - int localCH = 1 * CH1 + 2 * CH2 + 3 * CH3 + 4 * CH4 + 5 * CH5; - if (BLOCK % 2 == 0) - chanNmbr = (BLOCK / 2) * 9 + localCH; - else - chanNmbr = ((BLOCK - 1) / 2) * 9 + 4 + localCH; - if ((chanNmbr < 1) || (chanNmbr > 36)) - break; // signifies unexpected result - - uint32_t roc = 1; - uint32_t link = chanNmbr; - uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; - if (rID_temp != gpuClustering::invalidModuleId) rID = rID_temp; - break; - } - case 37: - case 38: { - uint32_t roc = - (errWord >> pixelgpudetails::ROC_shift) & pixelgpudetails::ROC_mask; - uint32_t link = - (errWord >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask; - uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; - if (rID_temp != gpuClustering::invalidModuleId) rID = rID_temp; - break; - } - default: - break; - }; - - return rID; -} - -// Kernel to perform Raw to Digi conversion -__global__ void RawToDigi_kernel( - const SiPixelROCsStatusAndMapping *cablingMap, - const unsigned char *modToUnp, const uint32_t wordCounter, - const uint32_t *word, const uint8_t *fedIds, uint16_t *xx, uint16_t *yy, - uint16_t *adc, uint32_t *pdigi, uint32_t *rawIdArr, uint16_t *moduleId, - cms::cuda::SimpleVector *err, bool useQualityInfo, - bool includeErrors, bool debug) { - // if (threadIdx.x==0) printf("Event: %u blockIdx.x: %u start: %u end: %u\n", - // eventno, blockIdx.x, begin, end); - - int32_t first = threadIdx.x + blockIdx.x * blockDim.x; - for (int32_t iloop = first, nend = wordCounter; iloop < nend; - iloop += blockDim.x * gridDim.x) { - auto gIndex = iloop; - xx[gIndex] = 0; - yy[gIndex] = 0; - adc[gIndex] = 0; - bool skipROC = false; - - uint8_t fedId = fedIds[gIndex / 2]; // +1200; - - // initialize (too many coninue below) - pdigi[gIndex] = 0; - rawIdArr[gIndex] = 0; - moduleId[gIndex] = gpuClustering::invalidModuleId; - - uint32_t ww = word[gIndex]; // Array containing 32 bit raw data - if (ww == 0) { - // 0 is an indicator of a noise/dead channel, skip these pixels during - // clusterization - continue; - } + case (28): { + if (debug) + printf("Error fifo nearly full (errorType = 28)\n"); + errorFound = true; + break; + } + case (29): { + if (debug) + printf("Timeout on a channel (errorType = 29)\n"); + if ((errorWord >> pixelgpudetails::OMIT_ERR_shift) & pixelgpudetails::OMIT_ERR_mask) { + if (debug) + printf("...first errorType=29 error, this gets masked out\n"); + } + errorFound = true; + break; + } + case (30): { + if (debug) + printf("TBM error trailer (errorType = 30)\n"); + int stateMatch_bits = 4; + int stateMatch_shift = 8; + uint32_t stateMatch_mask = ~(~uint32_t(0) << stateMatch_bits); + int stateMatch = (errorWord >> stateMatch_shift) & stateMatch_mask; + if (stateMatch != 1 && stateMatch != 8) { + if (debug) + printf("FED error 30 with unexpected State Bits (errorType = 30)\n"); + } + if (stateMatch == 1) + errorType = 40; // 1=Overflow -> 40, 8=number of ROCs -> 30 + errorFound = true; + break; + } + case (31): { + if (debug) + printf("Event number error (errorType = 31)\n"); + errorFound = true; + break; + } + default: + errorFound = false; + }; - uint32_t link = getLink(ww); // Extract link - uint32_t roc = getRoc(ww); // Extract Roc in link - pixelgpudetails::DetIdGPU detId = getRawId(cablingMap, fedId, link, roc); + return errorFound ? errorType : 0; + } - uint8_t errorType = checkROC(ww, fedId, link, cablingMap, debug); - skipROC = (roc < pixelgpudetails::maxROCIndex) ? false : (errorType != 0); - if (includeErrors and skipROC) { - uint32_t rID = getErrRawID(fedId, ww, errorType, cablingMap, debug); - err->push_back(SiPixelErrorCompact{rID, ww, errorType, fedId}); - continue; - } + // error decoding and handling copied from + // EventFilter/SiPixelRawToDigi/src/ErrorChecker.cc + __device__ uint32_t getErrRawID(uint8_t fedId, + uint32_t errWord, + uint32_t errorType, + const SiPixelROCsStatusAndMapping *cablingMap, + bool debug = false) { + uint32_t rID = 0xffffffff; + + switch (errorType) { + case 25: + case 30: + case 31: + case 36: + case 40: { + uint32_t roc = 1; + uint32_t link = (errWord >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask; + uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; + if (rID_temp != gpuClustering::invalidModuleId) + rID = rID_temp; + break; + } + case 29: { + int chanNmbr = 0; + const int DB0_shift = 0; + const int DB1_shift = DB0_shift + 1; + const int DB2_shift = DB1_shift + 1; + const int DB3_shift = DB2_shift + 1; + const int DB4_shift = DB3_shift + 1; + const uint32_t DataBit_mask = ~(~uint32_t(0) << 1); + + int CH1 = (errWord >> DB0_shift) & DataBit_mask; + int CH2 = (errWord >> DB1_shift) & DataBit_mask; + int CH3 = (errWord >> DB2_shift) & DataBit_mask; + int CH4 = (errWord >> DB3_shift) & DataBit_mask; + int CH5 = (errWord >> DB4_shift) & DataBit_mask; + int BLOCK_bits = 3; + int BLOCK_shift = 8; + uint32_t BLOCK_mask = ~(~uint32_t(0) << BLOCK_bits); + int BLOCK = (errWord >> BLOCK_shift) & BLOCK_mask; + int localCH = 1 * CH1 + 2 * CH2 + 3 * CH3 + 4 * CH4 + 5 * CH5; + if (BLOCK % 2 == 0) + chanNmbr = (BLOCK / 2) * 9 + localCH; + else + chanNmbr = ((BLOCK - 1) / 2) * 9 + 4 + localCH; + if ((chanNmbr < 1) || (chanNmbr > 36)) + break; // signifies unexpected result + + uint32_t roc = 1; + uint32_t link = chanNmbr; + uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; + if (rID_temp != gpuClustering::invalidModuleId) + rID = rID_temp; + break; + } + case 37: + case 38: { + uint32_t roc = (errWord >> pixelgpudetails::ROC_shift) & pixelgpudetails::ROC_mask; + uint32_t link = (errWord >> pixelgpudetails::LINK_shift) & pixelgpudetails::LINK_mask; + uint32_t rID_temp = getRawId(cablingMap, fedId, link, roc).rawId; + if (rID_temp != gpuClustering::invalidModuleId) + rID = rID_temp; + break; + } + default: + break; + }; - uint32_t rawId = detId.rawId; - uint32_t rocIdInDetUnit = detId.rocInDet; - bool barrel = isBarrel(rawId); + return rID; + } - uint32_t index = fedId * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + roc; - if (useQualityInfo) { - skipROC = cablingMap->badRocs[index]; - if (skipROC) continue; - } - skipROC = modToUnp[index]; - if (skipROC) continue; - - uint32_t layer = 0; - int side = 0, panel = 0, module = 0; - - if (barrel) { - layer = (rawId >> pixelgpudetails::layerStartBit) & - pixelgpudetails::layerMask; - module = (rawId >> pixelgpudetails::moduleStartBit) & - pixelgpudetails::moduleMask; - side = (module < 5) ? -1 : 1; - } else { - // endcap ids - layer = 0; - panel = (rawId >> pixelgpudetails::panelStartBit) & - pixelgpudetails::panelMask; - side = (panel == 1) ? -1 : 1; - } + // Kernel to perform Raw to Digi conversion + __global__ void RawToDigi_kernel(const SiPixelROCsStatusAndMapping *cablingMap, + const unsigned char *modToUnp, + const uint32_t wordCounter, + const uint32_t *word, + const uint8_t *fedIds, + uint16_t *xx, + uint16_t *yy, + uint16_t *adc, + uint32_t *pdigi, + uint32_t *rawIdArr, + uint16_t *moduleId, + cms::cuda::SimpleVector *err, + bool useQualityInfo, + bool includeErrors, + bool debug) { + // if (threadIdx.x==0) printf("Event: %u blockIdx.x: %u start: %u end: %u\n", + // eventno, blockIdx.x, begin, end); + + int32_t first = threadIdx.x + blockIdx.x * blockDim.x; + for (int32_t iloop = first, nend = wordCounter; iloop < nend; iloop += blockDim.x * gridDim.x) { + auto gIndex = iloop; + xx[gIndex] = 0; + yy[gIndex] = 0; + adc[gIndex] = 0; + bool skipROC = false; + + uint8_t fedId = fedIds[gIndex / 2]; // +1200; + + // initialize (too many coninue below) + pdigi[gIndex] = 0; + rawIdArr[gIndex] = 0; + moduleId[gIndex] = gpuClustering::invalidModuleId; + + uint32_t ww = word[gIndex]; // Array containing 32 bit raw data + if (ww == 0) { + // 0 is an indicator of a noise/dead channel, skip these pixels during + // clusterization + continue; + } - // ***special case of layer to 1 be handled here - pixelgpudetails::Pixel localPix; - if (layer == 1) { - uint32_t col = - (ww >> pixelgpudetails::COL_shift) & pixelgpudetails::COL_mask; - uint32_t row = - (ww >> pixelgpudetails::ROW_shift) & pixelgpudetails::ROW_mask; - localPix.row = row; - localPix.col = col; - if (includeErrors) { - if (not rocRowColIsValid(row, col)) { - uint8_t error = conversionError( - fedId, 3, debug); // use the device function and fill the arrays - err->push_back(SiPixelErrorCompact{rawId, ww, error, fedId}); - if (debug) printf("BPIX1 Error status: %i\n", error); + uint32_t link = getLink(ww); // Extract link + uint32_t roc = getRoc(ww); // Extract Roc in link + pixelgpudetails::DetIdGPU detId = getRawId(cablingMap, fedId, link, roc); + + uint8_t errorType = checkROC(ww, fedId, link, cablingMap, debug); + skipROC = (roc < pixelgpudetails::maxROCIndex) ? false : (errorType != 0); + if (includeErrors and skipROC) { + uint32_t rID = getErrRawID(fedId, ww, errorType, cablingMap, debug); + err->push_back(SiPixelErrorCompact{rID, ww, errorType, fedId}); + continue; + } + + uint32_t rawId = detId.rawId; + uint32_t rocIdInDetUnit = detId.rocInDet; + bool barrel = isBarrel(rawId); + + uint32_t index = fedId * MAX_LINK * MAX_ROC + (link - 1) * MAX_ROC + roc; + if (useQualityInfo) { + skipROC = cablingMap->badRocs[index]; + if (skipROC) continue; - } } - } else { - // ***conversion rules for dcol and pxid - uint32_t dcol = - (ww >> pixelgpudetails::DCOL_shift) & pixelgpudetails::DCOL_mask; - uint32_t pxid = - (ww >> pixelgpudetails::PXID_shift) & pixelgpudetails::PXID_mask; - uint32_t row = pixelgpudetails::numRowsInRoc - pxid / 2; - uint32_t col = dcol * 2 + pxid % 2; - localPix.row = row; - localPix.col = col; - if (includeErrors and not dcolIsValid(dcol, pxid)) { - uint8_t error = conversionError(fedId, 3, debug); - err->push_back(SiPixelErrorCompact{rawId, ww, error, fedId}); - if (debug) - printf("Error status: %i %d %d %d %d\n", error, dcol, pxid, fedId, - roc); + skipROC = modToUnp[index]; + if (skipROC) continue; + + uint32_t layer = 0; + int side = 0, panel = 0, module = 0; + + if (barrel) { + layer = (rawId >> pixelgpudetails::layerStartBit) & pixelgpudetails::layerMask; + module = (rawId >> pixelgpudetails::moduleStartBit) & pixelgpudetails::moduleMask; + side = (module < 5) ? -1 : 1; + } else { + // endcap ids + layer = 0; + panel = (rawId >> pixelgpudetails::panelStartBit) & pixelgpudetails::panelMask; + side = (panel == 1) ? -1 : 1; } - } - pixelgpudetails::Pixel globalPix = - frameConversion(barrel, side, layer, rocIdInDetUnit, localPix); - xx[gIndex] = globalPix.row; // origin shifting by 1 0-159 - yy[gIndex] = globalPix.col; // origin shifting by 1 0-415 - adc[gIndex] = getADC(ww); - pdigi[gIndex] = - pixelgpudetails::pack(globalPix.row, globalPix.col, adc[gIndex]); - moduleId[gIndex] = detId.moduleId; - rawIdArr[gIndex] = rawId; - } // end of loop (gIndex < end) - -} // end of Raw to Digi kernel - -__global__ void fillHitsModuleStart(uint32_t const *__restrict__ cluStart, - uint32_t *__restrict__ moduleStart) { - assert(gpuClustering::maxNumModules < - 2048); // easy to extend at least till 32*1024 - assert(1 == gridDim.x); - assert(0 == blockIdx.x); - - int first = threadIdx.x; - - // limit to maxHitsInModule() - for (int i = first, iend = gpuClustering::maxNumModules; i < iend; - i += blockDim.x) { - moduleStart[i + 1] = - std::min(gpuClustering::maxHitsInModule(), cluStart[i]); - } + // ***special case of layer to 1 be handled here + pixelgpudetails::Pixel localPix; + if (layer == 1) { + uint32_t col = (ww >> pixelgpudetails::COL_shift) & pixelgpudetails::COL_mask; + uint32_t row = (ww >> pixelgpudetails::ROW_shift) & pixelgpudetails::ROW_mask; + localPix.row = row; + localPix.col = col; + if (includeErrors) { + if (not rocRowColIsValid(row, col)) { + uint8_t error = conversionError(fedId, 3, debug); // use the device function and fill the arrays + err->push_back(SiPixelErrorCompact{rawId, ww, error, fedId}); + if (debug) + printf("BPIX1 Error status: %i\n", error); + continue; + } + } + } else { + // ***conversion rules for dcol and pxid + uint32_t dcol = (ww >> pixelgpudetails::DCOL_shift) & pixelgpudetails::DCOL_mask; + uint32_t pxid = (ww >> pixelgpudetails::PXID_shift) & pixelgpudetails::PXID_mask; + uint32_t row = pixelgpudetails::numRowsInRoc - pxid / 2; + uint32_t col = dcol * 2 + pxid % 2; + localPix.row = row; + localPix.col = col; + if (includeErrors and not dcolIsValid(dcol, pxid)) { + uint8_t error = conversionError(fedId, 3, debug); + err->push_back(SiPixelErrorCompact{rawId, ww, error, fedId}); + if (debug) + printf("Error status: %i %d %d %d %d\n", error, dcol, pxid, fedId, roc); + continue; + } + } - __shared__ uint32_t ws[32]; - cms::cuda::blockPrefixScan(moduleStart + 1, moduleStart + 1, 1024, ws); - cms::cuda::blockPrefixScan(moduleStart + 1025, moduleStart + 1025, - gpuClustering::maxNumModules - 1024, ws); + pixelgpudetails::Pixel globalPix = frameConversion(barrel, side, layer, rocIdInDetUnit, localPix); + xx[gIndex] = globalPix.row; // origin shifting by 1 0-159 + yy[gIndex] = globalPix.col; // origin shifting by 1 0-415 + adc[gIndex] = getADC(ww); + pdigi[gIndex] = pixelgpudetails::pack(globalPix.row, globalPix.col, adc[gIndex]); + moduleId[gIndex] = detId.moduleId; + rawIdArr[gIndex] = rawId; + } // end of loop (gIndex < end) - for (int i = first + 1025, iend = gpuClustering::maxNumModules + 1; i < iend; - i += blockDim.x) { - moduleStart[i] += moduleStart[1024]; - } - __syncthreads(); + } // end of Raw to Digi kernel + + __global__ void fillHitsModuleStart(uint32_t const *__restrict__ cluStart, uint32_t *__restrict__ moduleStart) { + assert(gpuClustering::maxNumModules < 2048); // easy to extend at least till 32*1024 + assert(1 == gridDim.x); + assert(0 == blockIdx.x); + + int first = threadIdx.x; + + // limit to maxHitsInModule() + for (int i = first, iend = gpuClustering::maxNumModules; i < iend; i += blockDim.x) { + moduleStart[i + 1] = std::min(gpuClustering::maxHitsInModule(), cluStart[i]); + } + + __shared__ uint32_t ws[32]; + cms::cuda::blockPrefixScan(moduleStart + 1, moduleStart + 1, 1024, ws); + cms::cuda::blockPrefixScan(moduleStart + 1025, moduleStart + 1025, gpuClustering::maxNumModules - 1024, ws); + + for (int i = first + 1025, iend = gpuClustering::maxNumModules + 1; i < iend; i += blockDim.x) { + moduleStart[i] += moduleStart[1024]; + } + __syncthreads(); #ifdef GPU_DEBUG - assert(0 == moduleStart[0]); - auto c0 = std::min(gpuClustering::maxHitsInModule(), cluStart[0]); - assert(c0 == moduleStart[1]); - assert(moduleStart[1024] >= moduleStart[1023]); - assert(moduleStart[1025] >= moduleStart[1024]); - assert(moduleStart[gpuClustering::maxNumModules] >= moduleStart[1025]); - - for (int i = first, iend = gpuClustering::maxNumModules + 1; i < iend; - i += blockDim.x) { - if (0 != i) assert(moduleStart[i] >= moduleStart[i - i]); - // [BPX1, BPX2, BPX3, BPX4, FP1, FP2, FP3, FN1, FN2, FN3, LAST_VALID] - // [ 0, 96, 320, 672, 1184, 1296, 1408, 1520, 1632, 1744, 1856] - if (i == 96 || i == 1184 || i == 1744 || i == gpuClustering::maxNumModules) - printf("moduleStart %d %d\n", i, moduleStart[i]); - } + assert(0 == moduleStart[0]); + auto c0 = std::min(gpuClustering::maxHitsInModule(), cluStart[0]); + assert(c0 == moduleStart[1]); + assert(moduleStart[1024] >= moduleStart[1023]); + assert(moduleStart[1025] >= moduleStart[1024]); + assert(moduleStart[gpuClustering::maxNumModules] >= moduleStart[1025]); + + for (int i = first, iend = gpuClustering::maxNumModules + 1; i < iend; i += blockDim.x) { + if (0 != i) + assert(moduleStart[i] >= moduleStart[i - i]); + // [BPX1, BPX2, BPX3, BPX4, FP1, FP2, FP3, FN1, FN2, FN3, LAST_VALID] + // [ 0, 96, 320, 672, 1184, 1296, 1408, 1520, 1632, 1744, 1856] + if (i == 96 || i == 1184 || i == 1744 || i == gpuClustering::maxNumModules) + printf("moduleStart %d %d\n", i, moduleStart[i]); + } #endif - // avoid overflow - auto constexpr maxNumClusters = gpuClustering::maxNumClusters; - for (int i = first, iend = gpuClustering::maxNumModules + 1; i < iend; - i += blockDim.x) { - moduleStart[i] = std::clamp(moduleStart[i], 0U, maxNumClusters); + // avoid overflow + auto constexpr maxNumClusters = gpuClustering::maxNumClusters; + for (int i = first, iend = gpuClustering::maxNumModules + 1; i < iend; i += blockDim.x) { + moduleStart[i] = std::clamp(moduleStart[i], 0U, maxNumClusters); + } } -} - -// Interface to outside -void SiPixelRawToClusterGPUKernel::makeClustersAsync( - bool isRun2, const SiPixelClusterThresholds clusterThresholds, - const SiPixelROCsStatusAndMapping *cablingMap, - const unsigned char *modToUnp, const SiPixelGainForHLTonGPU *gains, - const WordFedAppender &wordFed, SiPixelFormatterErrors &&errors, - const uint32_t wordCounter, const uint32_t fedCounter, bool useQualityInfo, - bool includeErrors, bool debug, cudaStream_t stream) { - nDigis = wordCounter; + + // Interface to outside + void SiPixelRawToClusterGPUKernel::makeClustersAsync(bool isRun2, + const SiPixelClusterThresholds clusterThresholds, + const SiPixelROCsStatusAndMapping *cablingMap, + const unsigned char *modToUnp, + const SiPixelGainForHLTonGPU *gains, + const WordFedAppender &wordFed, + SiPixelFormatterErrors &&errors, + const uint32_t wordCounter, + const uint32_t fedCounter, + bool useQualityInfo, + bool includeErrors, + bool debug, + cudaStream_t stream) { + nDigis = wordCounter; #ifdef GPU_DEBUG - std::cout << "decoding " << wordCounter << " digis. Max is " - << pixelgpudetails::MAX_FED_WORDS << std::endl; + std::cout << "decoding " << wordCounter << " digis. Max is " << pixelgpudetails::MAX_FED_WORDS << std::endl; #endif - digis_d = SiPixelDigisCUDA(pixelgpudetails::MAX_FED_WORDS, stream); - if (includeErrors) { - digiErrors_d = SiPixelDigiErrorsCUDA(pixelgpudetails::MAX_FED_WORDS, - std::move(errors), stream); - } - clusters_d = SiPixelClustersCUDA(gpuClustering::maxNumModules, stream); - - nModules_Clusters_h = cms::cuda::make_host_unique(2, stream); - - if (wordCounter) // protect in case of empty event.... - { - const int threadsPerBlock = 512; - const int blocks = - (wordCounter + threadsPerBlock - 1) / threadsPerBlock; // fill it all - - assert(0 == wordCounter % 2); - // wordCounter is the total no of words in each event to be trasfered on - // device - auto word_d = - cms::cuda::make_device_unique(wordCounter, stream); - auto fedId_d = - cms::cuda::make_device_unique(wordCounter, stream); - - cudaCheck(cudaMemcpyAsync(word_d.get(), wordFed.word(), - wordCounter * sizeof(uint32_t), cudaMemcpyDefault, - stream)); - cudaCheck(cudaMemcpyAsync(fedId_d.get(), wordFed.fedId(), - wordCounter * sizeof(uint8_t) / 2, - cudaMemcpyDefault, stream)); - - // Launch rawToDigi kernel - RawToDigi_kernel<<>>( - cablingMap, modToUnp, wordCounter, word_d.get(), fedId_d.get(), - digis_d.xx(), digis_d.yy(), digis_d.adc(), digis_d.pdigi(), - digis_d.rawIdArr(), digis_d.moduleInd(), - digiErrors_d.error(), // returns nullptr if default-constructed - useQualityInfo, includeErrors, debug); - cudaCheck(cudaGetLastError()); + digis_d = SiPixelDigisCUDA(pixelgpudetails::MAX_FED_WORDS, stream); + if (includeErrors) { + digiErrors_d = SiPixelDigiErrorsCUDA(pixelgpudetails::MAX_FED_WORDS, std::move(errors), stream); + } + clusters_d = SiPixelClustersCUDA(gpuClustering::maxNumModules, stream); + + nModules_Clusters_h = cms::cuda::make_host_unique(2, stream); + + if (wordCounter) // protect in case of empty event.... + { + const int threadsPerBlock = 512; + const int blocks = (wordCounter + threadsPerBlock - 1) / threadsPerBlock; // fill it all + + assert(0 == wordCounter % 2); + // wordCounter is the total no of words in each event to be trasfered on + // device + auto word_d = cms::cuda::make_device_unique(wordCounter, stream); + auto fedId_d = cms::cuda::make_device_unique(wordCounter, stream); + + cudaCheck( + cudaMemcpyAsync(word_d.get(), wordFed.word(), wordCounter * sizeof(uint32_t), cudaMemcpyDefault, stream)); + cudaCheck(cudaMemcpyAsync( + fedId_d.get(), wordFed.fedId(), wordCounter * sizeof(uint8_t) / 2, cudaMemcpyDefault, stream)); + + // Launch rawToDigi kernel + RawToDigi_kernel<<>>( + cablingMap, + modToUnp, + wordCounter, + word_d.get(), + fedId_d.get(), + digis_d.xx(), + digis_d.yy(), + digis_d.adc(), + digis_d.pdigi(), + digis_d.rawIdArr(), + digis_d.moduleInd(), + digiErrors_d.error(), // returns nullptr if default-constructed + useQualityInfo, + includeErrors, + debug); + cudaCheck(cudaGetLastError()); #ifdef GPU_DEBUG - cudaDeviceSynchronize(); - cudaCheck(cudaGetLastError()); + cudaDeviceSynchronize(); + cudaCheck(cudaGetLastError()); #endif - if (includeErrors) { - digiErrors_d.copyErrorToHostAsync(stream); + if (includeErrors) { + digiErrors_d.copyErrorToHostAsync(stream); + } } - } - // End of Raw2Digi and passing data for clustering - - { - // clusterizer ... - using namespace gpuClustering; - int threadsPerBlock = 256; - int blocks = - (std::max(int(wordCounter), int(gpuClustering::maxNumModules)) + - threadsPerBlock - 1) / - threadsPerBlock; - - gpuCalibPixel::calibDigis<<>>( - isRun2, digis_d.moduleInd(), digis_d.xx(), digis_d.yy(), digis_d.adc(), - gains, wordCounter, clusters_d.moduleStart(), clusters_d.clusInModule(), - clusters_d.clusModuleStart()); - cudaCheck(cudaGetLastError()); + // End of Raw2Digi and passing data for clustering + + { + // clusterizer ... + using namespace gpuClustering; + int threadsPerBlock = 256; + int blocks = + (std::max(int(wordCounter), int(gpuClustering::maxNumModules)) + threadsPerBlock - 1) / threadsPerBlock; + + gpuCalibPixel::calibDigis<<>>(isRun2, + digis_d.moduleInd(), + digis_d.xx(), + digis_d.yy(), + digis_d.adc(), + gains, + wordCounter, + clusters_d.moduleStart(), + clusters_d.clusInModule(), + clusters_d.clusModuleStart()); + cudaCheck(cudaGetLastError()); #ifdef GPU_DEBUG - cudaDeviceSynchronize(); - cudaCheck(cudaGetLastError()); + cudaDeviceSynchronize(); + cudaCheck(cudaGetLastError()); #endif #ifdef GPU_DEBUG - std::cout << "CUDA countModules kernel launch with " << blocks - << " blocks of " << threadsPerBlock << " threads\n"; + std::cout << "CUDA countModules kernel launch with " << blocks << " blocks of " << threadsPerBlock + << " threads\n"; #endif - countModules<<>>( - digis_d.moduleInd(), clusters_d.moduleStart(), digis_d.clus(), - wordCounter); - cudaCheck(cudaGetLastError()); + countModules<<>>( + digis_d.moduleInd(), clusters_d.moduleStart(), digis_d.clus(), wordCounter); + cudaCheck(cudaGetLastError()); - // read the number of modules into a data member, used by getProduct()) - cudaCheck(cudaMemcpyAsync(&(nModules_Clusters_h[0]), - clusters_d.moduleStart(), sizeof(uint32_t), - cudaMemcpyDefault, stream)); + // read the number of modules into a data member, used by getProduct()) + cudaCheck(cudaMemcpyAsync( + &(nModules_Clusters_h[0]), clusters_d.moduleStart(), sizeof(uint32_t), cudaMemcpyDefault, stream)); - threadsPerBlock = 256; - blocks = maxNumModules; + threadsPerBlock = 256; + blocks = maxNumModules; #ifdef GPU_DEBUG - std::cout << "CUDA findClus kernel launch with " << blocks << " blocks of " - << threadsPerBlock << " threads\n"; + std::cout << "CUDA findClus kernel launch with " << blocks << " blocks of " << threadsPerBlock << " threads\n"; #endif - findClus<<>>( - digis_d.moduleInd(), digis_d.xx(), digis_d.yy(), - clusters_d.moduleStart(), clusters_d.clusInModule(), - clusters_d.moduleId(), digis_d.clus(), wordCounter); - cudaCheck(cudaGetLastError()); + findClus<<>>(digis_d.moduleInd(), + digis_d.xx(), + digis_d.yy(), + clusters_d.moduleStart(), + clusters_d.clusInModule(), + clusters_d.moduleId(), + digis_d.clus(), + wordCounter); + cudaCheck(cudaGetLastError()); #ifdef GPU_DEBUG - cudaDeviceSynchronize(); - cudaCheck(cudaGetLastError()); + cudaDeviceSynchronize(); + cudaCheck(cudaGetLastError()); #endif - // apply charge cut - clusterChargeCut<<>>( - clusterThresholds, digis_d.moduleInd(), digis_d.adc(), - clusters_d.moduleStart(), clusters_d.clusInModule(), - clusters_d.moduleId(), digis_d.clus(), wordCounter); - cudaCheck(cudaGetLastError()); - - // count the module start indices already here (instead of - // rechits) so that the number of clusters/hits can be made - // available in the rechit producer without additional points of - // synchronization/ExternalWork - - // MUST be ONE block - fillHitsModuleStart<<<1, 1024, 0, stream>>>(clusters_d.clusInModule(), - clusters_d.clusModuleStart()); - - // last element holds the number of all clusters - cudaCheck(cudaMemcpyAsync( - &(nModules_Clusters_h[1]), - clusters_d.clusModuleStart() + gpuClustering::maxNumModules, - sizeof(uint32_t), cudaMemcpyDefault, stream)); + // apply charge cut + clusterChargeCut<<>>(clusterThresholds, + digis_d.moduleInd(), + digis_d.adc(), + clusters_d.moduleStart(), + clusters_d.clusInModule(), + clusters_d.moduleId(), + digis_d.clus(), + wordCounter); + cudaCheck(cudaGetLastError()); + + // count the module start indices already here (instead of + // rechits) so that the number of clusters/hits can be made + // available in the rechit producer without additional points of + // synchronization/ExternalWork + + // MUST be ONE block + fillHitsModuleStart<<<1, 1024, 0, stream>>>(clusters_d.clusInModule(), clusters_d.clusModuleStart()); + + // last element holds the number of all clusters + cudaCheck(cudaMemcpyAsync(&(nModules_Clusters_h[1]), + clusters_d.clusModuleStart() + gpuClustering::maxNumModules, + sizeof(uint32_t), + cudaMemcpyDefault, + stream)); #ifdef GPU_DEBUG - cudaDeviceSynchronize(); - cudaCheck(cudaGetLastError()); + cudaDeviceSynchronize(); + cudaCheck(cudaGetLastError()); #endif - } // end clusterizer scope -} + } // end clusterizer scope + } } // namespace pixelgpudetails diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.h b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.h index d2465ef94b428..f9897104dd421 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.h +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelRawToClusterGPUKernel.h @@ -23,195 +23,194 @@ class SiPixelGainForHLTonGPU; namespace pixelgpudetails { -// Phase 1 geometry constants -const uint32_t layerStartBit = 20; -const uint32_t ladderStartBit = 12; -const uint32_t moduleStartBit = 2; - -const uint32_t panelStartBit = 10; -const uint32_t diskStartBit = 18; -const uint32_t bladeStartBit = 12; - -const uint32_t layerMask = 0xF; -const uint32_t ladderMask = 0xFF; -const uint32_t moduleMask = 0x3FF; -const uint32_t panelMask = 0x3; -const uint32_t diskMask = 0xF; -const uint32_t bladeMask = 0x3F; - -const uint32_t LINK_bits = 6; -const uint32_t ROC_bits = 5; -const uint32_t DCOL_bits = 5; -const uint32_t PXID_bits = 8; -const uint32_t ADC_bits = 8; - -// special for layer 1 -const uint32_t LINK_bits_l1 = 6; -const uint32_t ROC_bits_l1 = 5; -const uint32_t COL_bits_l1 = 6; -const uint32_t ROW_bits_l1 = 7; -const uint32_t OMIT_ERR_bits = 1; - -const uint32_t maxROCIndex = 8; -const uint32_t numRowsInRoc = 80; -const uint32_t numColsInRoc = 52; - -const uint32_t MAX_WORD = 2000; - -const uint32_t ADC_shift = 0; -const uint32_t PXID_shift = ADC_shift + ADC_bits; -const uint32_t DCOL_shift = PXID_shift + PXID_bits; -const uint32_t ROC_shift = DCOL_shift + DCOL_bits; -const uint32_t LINK_shift = ROC_shift + ROC_bits_l1; -// special for layer 1 ROC -const uint32_t ROW_shift = ADC_shift + ADC_bits; -const uint32_t COL_shift = ROW_shift + ROW_bits_l1; -const uint32_t OMIT_ERR_shift = 20; - -const uint32_t LINK_mask = ~(~uint32_t(0) << LINK_bits_l1); -const uint32_t ROC_mask = ~(~uint32_t(0) << ROC_bits_l1); -const uint32_t COL_mask = ~(~uint32_t(0) << COL_bits_l1); -const uint32_t ROW_mask = ~(~uint32_t(0) << ROW_bits_l1); -const uint32_t DCOL_mask = ~(~uint32_t(0) << DCOL_bits); -const uint32_t PXID_mask = ~(~uint32_t(0) << PXID_bits); -const uint32_t ADC_mask = ~(~uint32_t(0) << ADC_bits); -const uint32_t ERROR_mask = ~(~uint32_t(0) << ROC_bits_l1); -const uint32_t OMIT_ERR_mask = ~(~uint32_t(0) << OMIT_ERR_bits); - -struct DetIdGPU { - uint32_t rawId; - uint32_t rocInDet; - uint32_t moduleId; -}; - -struct Pixel { - uint32_t row; - uint32_t col; -}; - -class Packing { - public: - using PackedDigiType = uint32_t; - - // Constructor: pre-computes masks and shifts from field widths - __host__ __device__ inline constexpr Packing(unsigned int row_w, - unsigned int column_w, - unsigned int time_w, - unsigned int adc_w) - : row_width(row_w), - column_width(column_w), - adc_width(adc_w), - row_shift(0), - column_shift(row_shift + row_w), - time_shift(column_shift + column_w), - adc_shift(time_shift + time_w), - row_mask(~(~0U << row_w)), - column_mask(~(~0U << column_w)), - time_mask(~(~0U << time_w)), - adc_mask(~(~0U << adc_w)), - rowcol_mask(~(~0U << (column_w + row_w))), - max_row(row_mask), - max_column(column_mask), - max_adc(adc_mask) {} - - uint32_t row_width; - uint32_t column_width; - uint32_t adc_width; - - uint32_t row_shift; - uint32_t column_shift; - uint32_t time_shift; - uint32_t adc_shift; - - PackedDigiType row_mask; - PackedDigiType column_mask; - PackedDigiType time_mask; - PackedDigiType adc_mask; - PackedDigiType rowcol_mask; - - uint32_t max_row; - uint32_t max_column; - uint32_t max_adc; -}; - -__host__ __device__ inline constexpr Packing packing() { - return Packing(11, 11, 0, 10); -} - -__host__ __device__ inline uint32_t pack(uint32_t row, uint32_t col, - uint32_t adc) { - constexpr Packing thePacking = packing(); - adc = std::min(adc, thePacking.max_adc); - - return (row << thePacking.row_shift) | (col << thePacking.column_shift) | - (adc << thePacking.adc_shift); -} - -constexpr uint32_t pixelToChannel(int row, int col) { - constexpr Packing thePacking = packing(); - return (row << thePacking.column_width) | col; -} - -class SiPixelRawToClusterGPUKernel { - public: - class WordFedAppender { - public: - WordFedAppender(); - ~WordFedAppender() = default; - - void initializeWordFed(int fedId, unsigned int wordCounterGPU, - const cms_uint32_t* src, unsigned int length); - - const unsigned int* word() const { return word_.get(); } - const unsigned char* fedId() const { return fedId_.get(); } - - private: - cms::cuda::host::noncached::unique_ptr word_; - cms::cuda::host::noncached::unique_ptr fedId_; + // Phase 1 geometry constants + const uint32_t layerStartBit = 20; + const uint32_t ladderStartBit = 12; + const uint32_t moduleStartBit = 2; + + const uint32_t panelStartBit = 10; + const uint32_t diskStartBit = 18; + const uint32_t bladeStartBit = 12; + + const uint32_t layerMask = 0xF; + const uint32_t ladderMask = 0xFF; + const uint32_t moduleMask = 0x3FF; + const uint32_t panelMask = 0x3; + const uint32_t diskMask = 0xF; + const uint32_t bladeMask = 0x3F; + + const uint32_t LINK_bits = 6; + const uint32_t ROC_bits = 5; + const uint32_t DCOL_bits = 5; + const uint32_t PXID_bits = 8; + const uint32_t ADC_bits = 8; + + // special for layer 1 + const uint32_t LINK_bits_l1 = 6; + const uint32_t ROC_bits_l1 = 5; + const uint32_t COL_bits_l1 = 6; + const uint32_t ROW_bits_l1 = 7; + const uint32_t OMIT_ERR_bits = 1; + + const uint32_t maxROCIndex = 8; + const uint32_t numRowsInRoc = 80; + const uint32_t numColsInRoc = 52; + + const uint32_t MAX_WORD = 2000; + + const uint32_t ADC_shift = 0; + const uint32_t PXID_shift = ADC_shift + ADC_bits; + const uint32_t DCOL_shift = PXID_shift + PXID_bits; + const uint32_t ROC_shift = DCOL_shift + DCOL_bits; + const uint32_t LINK_shift = ROC_shift + ROC_bits_l1; + // special for layer 1 ROC + const uint32_t ROW_shift = ADC_shift + ADC_bits; + const uint32_t COL_shift = ROW_shift + ROW_bits_l1; + const uint32_t OMIT_ERR_shift = 20; + + const uint32_t LINK_mask = ~(~uint32_t(0) << LINK_bits_l1); + const uint32_t ROC_mask = ~(~uint32_t(0) << ROC_bits_l1); + const uint32_t COL_mask = ~(~uint32_t(0) << COL_bits_l1); + const uint32_t ROW_mask = ~(~uint32_t(0) << ROW_bits_l1); + const uint32_t DCOL_mask = ~(~uint32_t(0) << DCOL_bits); + const uint32_t PXID_mask = ~(~uint32_t(0) << PXID_bits); + const uint32_t ADC_mask = ~(~uint32_t(0) << ADC_bits); + const uint32_t ERROR_mask = ~(~uint32_t(0) << ROC_bits_l1); + const uint32_t OMIT_ERR_mask = ~(~uint32_t(0) << OMIT_ERR_bits); + + struct DetIdGPU { + uint32_t rawId; + uint32_t rocInDet; + uint32_t moduleId; }; - SiPixelRawToClusterGPUKernel() = default; - ~SiPixelRawToClusterGPUKernel() = default; - - SiPixelRawToClusterGPUKernel(const SiPixelRawToClusterGPUKernel&) = delete; - SiPixelRawToClusterGPUKernel(SiPixelRawToClusterGPUKernel&&) = delete; - SiPixelRawToClusterGPUKernel& operator=(const SiPixelRawToClusterGPUKernel&) = - delete; - SiPixelRawToClusterGPUKernel& operator=(SiPixelRawToClusterGPUKernel&&) = - delete; - - void makeClustersAsync( - bool isRun2, const SiPixelClusterThresholds clusterThresholds, - const SiPixelROCsStatusAndMapping* cablingMap, - const unsigned char* modToUnp, const SiPixelGainForHLTonGPU* gains, - const WordFedAppender& wordFed, SiPixelFormatterErrors&& errors, - const uint32_t wordCounter, const uint32_t fedCounter, - bool useQualityInfo, bool includeErrors, bool debug, cudaStream_t stream); - - std::pair getResults() { - digis_d.setNModulesDigis(nModules_Clusters_h[0], nDigis); - clusters_d.setNClusters(nModules_Clusters_h[1]); - // need to explicitly deallocate while the associated CUDA - // stream is still alive - // - // technically the statement above is not true anymore now that - // the CUDA streams are cached within the cms::cuda::StreamCache, but it is - // still better to release as early as possible - nModules_Clusters_h.reset(); - return std::make_pair(std::move(digis_d), std::move(clusters_d)); - } + struct Pixel { + uint32_t row; + uint32_t col; + }; - SiPixelDigiErrorsCUDA&& getErrors() { return std::move(digiErrors_d); } + class Packing { + public: + using PackedDigiType = uint32_t; + + // Constructor: pre-computes masks and shifts from field widths + __host__ __device__ inline constexpr Packing(unsigned int row_w, + unsigned int column_w, + unsigned int time_w, + unsigned int adc_w) + : row_width(row_w), + column_width(column_w), + adc_width(adc_w), + row_shift(0), + column_shift(row_shift + row_w), + time_shift(column_shift + column_w), + adc_shift(time_shift + time_w), + row_mask(~(~0U << row_w)), + column_mask(~(~0U << column_w)), + time_mask(~(~0U << time_w)), + adc_mask(~(~0U << adc_w)), + rowcol_mask(~(~0U << (column_w + row_w))), + max_row(row_mask), + max_column(column_mask), + max_adc(adc_mask) {} + + uint32_t row_width; + uint32_t column_width; + uint32_t adc_width; + + uint32_t row_shift; + uint32_t column_shift; + uint32_t time_shift; + uint32_t adc_shift; + + PackedDigiType row_mask; + PackedDigiType column_mask; + PackedDigiType time_mask; + PackedDigiType adc_mask; + PackedDigiType rowcol_mask; + + uint32_t max_row; + uint32_t max_column; + uint32_t max_adc; + }; - private: - uint32_t nDigis = 0; + __host__ __device__ inline constexpr Packing packing() { return Packing(11, 11, 0, 10); } - // Data to be put in the event - cms::cuda::host::unique_ptr nModules_Clusters_h; - SiPixelDigisCUDA digis_d; - SiPixelClustersCUDA clusters_d; - SiPixelDigiErrorsCUDA digiErrors_d; -}; + __host__ __device__ inline uint32_t pack(uint32_t row, uint32_t col, uint32_t adc) { + constexpr Packing thePacking = packing(); + adc = std::min(adc, thePacking.max_adc); + + return (row << thePacking.row_shift) | (col << thePacking.column_shift) | (adc << thePacking.adc_shift); + } + + constexpr uint32_t pixelToChannel(int row, int col) { + constexpr Packing thePacking = packing(); + return (row << thePacking.column_width) | col; + } + + class SiPixelRawToClusterGPUKernel { + public: + class WordFedAppender { + public: + WordFedAppender(); + ~WordFedAppender() = default; + + void initializeWordFed(int fedId, unsigned int wordCounterGPU, const cms_uint32_t* src, unsigned int length); + + const unsigned int* word() const { return word_.get(); } + const unsigned char* fedId() const { return fedId_.get(); } + + private: + cms::cuda::host::noncached::unique_ptr word_; + cms::cuda::host::noncached::unique_ptr fedId_; + }; + + SiPixelRawToClusterGPUKernel() = default; + ~SiPixelRawToClusterGPUKernel() = default; + + SiPixelRawToClusterGPUKernel(const SiPixelRawToClusterGPUKernel&) = delete; + SiPixelRawToClusterGPUKernel(SiPixelRawToClusterGPUKernel&&) = delete; + SiPixelRawToClusterGPUKernel& operator=(const SiPixelRawToClusterGPUKernel&) = delete; + SiPixelRawToClusterGPUKernel& operator=(SiPixelRawToClusterGPUKernel&&) = delete; + + void makeClustersAsync(bool isRun2, + const SiPixelClusterThresholds clusterThresholds, + const SiPixelROCsStatusAndMapping* cablingMap, + const unsigned char* modToUnp, + const SiPixelGainForHLTonGPU* gains, + const WordFedAppender& wordFed, + SiPixelFormatterErrors&& errors, + const uint32_t wordCounter, + const uint32_t fedCounter, + bool useQualityInfo, + bool includeErrors, + bool debug, + cudaStream_t stream); + + std::pair getResults() { + digis_d.setNModulesDigis(nModules_Clusters_h[0], nDigis); + clusters_d.setNClusters(nModules_Clusters_h[1]); + // need to explicitly deallocate while the associated CUDA + // stream is still alive + // + // technically the statement above is not true anymore now that + // the CUDA streams are cached within the cms::cuda::StreamCache, but it is + // still better to release as early as possible + nModules_Clusters_h.reset(); + return std::make_pair(std::move(digis_d), std::move(clusters_d)); + } + + SiPixelDigiErrorsCUDA&& getErrors() { return std::move(digiErrors_d); } + + private: + uint32_t nDigis = 0; + + // Data to be put in the event + cms::cuda::host::unique_ptr nModules_Clusters_h; + SiPixelDigisCUDA digis_d; + SiPixelClustersCUDA clusters_d; + SiPixelDigiErrorsCUDA digiErrors_d; + }; } // namespace pixelgpudetails diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/gpuClusterChargeCut.h b/RecoLocalTracker/SiPixelClusterizer/plugins/gpuClusterChargeCut.h index b49edeb089068..13376cebc6f05 100644 --- a/RecoLocalTracker/SiPixelClusterizer/plugins/gpuClusterChargeCut.h +++ b/RecoLocalTracker/SiPixelClusterizer/plugins/gpuClusterChargeCut.h @@ -14,110 +14,121 @@ namespace gpuClustering { -__global__ void clusterChargeCut( - SiPixelClusterThresholds - clusterThresholds, // charge cut on cluster in electrons (for layer 1 - // and for other layers) - uint16_t* __restrict__ id, // module id of each pixel (modified if bad - // cluster) - uint16_t const* __restrict__ adc, // charge of each pixel - uint32_t const* __restrict__ moduleStart, // index of the first pixel of - // each module - uint32_t* __restrict__ nClustersInModule, // modified: number of clusters - // found in each module - uint32_t const* __restrict__ moduleId, // module id of each module - int32_t* __restrict__ clusterId, // modified: cluster id of each pixel - uint32_t numElements) { - __shared__ int32_t charge[maxNumClustersPerModules]; - __shared__ uint8_t ok[maxNumClustersPerModules]; - __shared__ uint16_t newclusId[maxNumClustersPerModules]; - - auto firstModule = blockIdx.x; - auto endModule = moduleStart[0]; - for (auto module = firstModule; module < endModule; module += gridDim.x) { - auto firstPixel = moduleStart[1 + module]; - auto thisModuleId = id[firstPixel]; - assert(thisModuleId < maxNumModules); - assert(thisModuleId == moduleId[module]); - - auto nclus = nClustersInModule[thisModuleId]; - if (nclus == 0) continue; - - if (threadIdx.x == 0 && nclus > maxNumClustersPerModules) - printf("Warning too many clusters in module %d in block %d: %d > %d\n", - thisModuleId, blockIdx.x, nclus, maxNumClustersPerModules); - - auto first = firstPixel + threadIdx.x; - - if (nclus > maxNumClustersPerModules) { - // remove excess FIXME find a way to cut charge first.... - for (auto i = first; i < numElements; i += blockDim.x) { - if (id[i] == invalidModuleId) continue; // not valid - if (id[i] != thisModuleId) break; // end of module - if (clusterId[i] >= maxNumClustersPerModules) { - id[i] = invalidModuleId; - clusterId[i] = invalidModuleId; + __global__ void clusterChargeCut( + SiPixelClusterThresholds clusterThresholds, // charge cut on cluster in electrons (for layer 1 + // and for other layers) + uint16_t* __restrict__ id, // module id of each pixel (modified if bad + // cluster) + uint16_t const* __restrict__ adc, // charge of each pixel + uint32_t const* __restrict__ moduleStart, // index of the first pixel of + // each module + uint32_t* __restrict__ nClustersInModule, // modified: number of clusters + // found in each module + uint32_t const* __restrict__ moduleId, // module id of each module + int32_t* __restrict__ clusterId, // modified: cluster id of each pixel + uint32_t numElements) { + __shared__ int32_t charge[maxNumClustersPerModules]; + __shared__ uint8_t ok[maxNumClustersPerModules]; + __shared__ uint16_t newclusId[maxNumClustersPerModules]; + + auto firstModule = blockIdx.x; + auto endModule = moduleStart[0]; + for (auto module = firstModule; module < endModule; module += gridDim.x) { + auto firstPixel = moduleStart[1 + module]; + auto thisModuleId = id[firstPixel]; + assert(thisModuleId < maxNumModules); + assert(thisModuleId == moduleId[module]); + + auto nclus = nClustersInModule[thisModuleId]; + if (nclus == 0) + continue; + + if (threadIdx.x == 0 && nclus > maxNumClustersPerModules) + printf("Warning too many clusters in module %d in block %d: %d > %d\n", + thisModuleId, + blockIdx.x, + nclus, + maxNumClustersPerModules); + + auto first = firstPixel + threadIdx.x; + + if (nclus > maxNumClustersPerModules) { + // remove excess FIXME find a way to cut charge first.... + for (auto i = first; i < numElements; i += blockDim.x) { + if (id[i] == invalidModuleId) + continue; // not valid + if (id[i] != thisModuleId) + break; // end of module + if (clusterId[i] >= maxNumClustersPerModules) { + id[i] = invalidModuleId; + clusterId[i] = invalidModuleId; + } } + nclus = maxNumClustersPerModules; } - nclus = maxNumClustersPerModules; - } #ifdef GPU_DEBUG - if (thisModuleId % 100 == 1) - if (threadIdx.x == 0) - printf("start cluster charge cut for module %d in block %d\n", - thisModuleId, blockIdx.x); + if (thisModuleId % 100 == 1) + if (threadIdx.x == 0) + printf("start cluster charge cut for module %d in block %d\n", thisModuleId, blockIdx.x); #endif - assert(nclus <= maxNumClustersPerModules); - for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { - charge[i] = 0; - } - __syncthreads(); - - for (auto i = first; i < numElements; i += blockDim.x) { - if (id[i] == invalidModuleId) continue; // not valid - if (id[i] != thisModuleId) break; // end of module - atomicAdd(&charge[clusterId[i]], adc[i]); - } - __syncthreads(); - - auto chargeCut = clusterThresholds.getThresholdForLayerOnCondition( - thisModuleId < phase1PixelTopology::layerStart[1]); - for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { - newclusId[i] = ok[i] = charge[i] > chargeCut ? 1 : 0; - } - - __syncthreads(); - - // renumber - __shared__ uint16_t ws[32]; - cms::cuda::blockPrefixScan(newclusId, nclus, ws); - - assert(nclus >= newclusId[nclus - 1]); - - if (nclus == newclusId[nclus - 1]) continue; - - nClustersInModule[thisModuleId] = newclusId[nclus - 1]; - __syncthreads(); - - // mark bad cluster again - for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { - if (0 == ok[i]) newclusId[i] = invalidModuleId + 1; - } - __syncthreads(); - - // reassign id - for (auto i = first; i < numElements; i += blockDim.x) { - if (id[i] == invalidModuleId) continue; // not valid - if (id[i] != thisModuleId) break; // end of module - clusterId[i] = newclusId[clusterId[i]] - 1; - if (clusterId[i] == invalidModuleId) id[i] = invalidModuleId; - } - - // done - } // loop on modules -} + assert(nclus <= maxNumClustersPerModules); + for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { + charge[i] = 0; + } + __syncthreads(); + + for (auto i = first; i < numElements; i += blockDim.x) { + if (id[i] == invalidModuleId) + continue; // not valid + if (id[i] != thisModuleId) + break; // end of module + atomicAdd(&charge[clusterId[i]], adc[i]); + } + __syncthreads(); + + auto chargeCut = + clusterThresholds.getThresholdForLayerOnCondition(thisModuleId < phase1PixelTopology::layerStart[1]); + for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { + newclusId[i] = ok[i] = charge[i] > chargeCut ? 1 : 0; + } + + __syncthreads(); + + // renumber + __shared__ uint16_t ws[32]; + cms::cuda::blockPrefixScan(newclusId, nclus, ws); + + assert(nclus >= newclusId[nclus - 1]); + + if (nclus == newclusId[nclus - 1]) + continue; + + nClustersInModule[thisModuleId] = newclusId[nclus - 1]; + __syncthreads(); + + // mark bad cluster again + for (auto i = threadIdx.x; i < nclus; i += blockDim.x) { + if (0 == ok[i]) + newclusId[i] = invalidModuleId + 1; + } + __syncthreads(); + + // reassign id + for (auto i = first; i < numElements; i += blockDim.x) { + if (id[i] == invalidModuleId) + continue; // not valid + if (id[i] != thisModuleId) + break; // end of module + clusterId[i] = newclusId[clusterId[i]] - 1; + if (clusterId[i] == invalidModuleId) + id[i] = invalidModuleId; + } + + // done + } // loop on modules + } } // namespace gpuClustering diff --git a/RecoLocalTracker/SiPixelClusterizer/test/gpuClustering_t.h b/RecoLocalTracker/SiPixelClusterizer/test/gpuClustering_t.h index af34cf34905db..4823a7c27906f 100644 --- a/RecoLocalTracker/SiPixelClusterizer/test/gpuClustering_t.h +++ b/RecoLocalTracker/SiPixelClusterizer/test/gpuClustering_t.h @@ -28,8 +28,7 @@ int main(void) { using namespace gpuClustering; constexpr int numElements = 256 * maxNumModules; - constexpr SiPixelClusterThresholds clusterThresholds( - kSiPixelClusterThresholdsDefaultPhase1); + constexpr SiPixelClusterThresholds clusterThresholds(kSiPixelClusterThresholdsDefaultPhase1); // these in reality are already on GPU auto h_id = std::make_unique(numElements); @@ -44,12 +43,9 @@ int main(void) { auto d_y = cms::cuda::make_device_unique(numElements, nullptr); auto d_adc = cms::cuda::make_device_unique(numElements, nullptr); auto d_clus = cms::cuda::make_device_unique(numElements, nullptr); - auto d_moduleStart = - cms::cuda::make_device_unique(maxNumModules + 1, nullptr); - auto d_clusInModule = - cms::cuda::make_device_unique(maxNumModules, nullptr); - auto d_moduleId = - cms::cuda::make_device_unique(maxNumModules, nullptr); + auto d_moduleStart = cms::cuda::make_device_unique(maxNumModules + 1, nullptr); + auto d_clusInModule = cms::cuda::make_device_unique(maxNumModules, nullptr); + auto d_moduleId = cms::cuda::make_device_unique(maxNumModules, nullptr); #else // __CUDACC__ auto h_moduleStart = std::make_unique(maxNumModules + 1); auto h_clusInModule = std::make_unique(maxNumModules); @@ -74,9 +70,11 @@ int main(void) { h_adc[n] = 1000; ++n; ++ncl; - if (MaxPixels <= ncl) break; + if (MaxPixels <= ncl) + break; } - if (MaxPixels <= ncl) break; + if (MaxPixels <= ncl) + break; } } @@ -190,7 +188,8 @@ int main(void) { } // all odd id for (int id = 11; id <= 1800; id += 2) { - if ((id / 20) % 2) h_id[n++] = invalidModuleId; // error + if ((id / 20) % 2) + h_id[n++] = invalidModuleId; // error for (int x = 0; x < 40; x += 4) { ++ncl; if ((id / 10) % 2) { @@ -213,7 +212,8 @@ int main(void) { h_y[n] = x + y[9 - k]; h_adc[n] = kn == 2 ? 10 : 1000; ++n; - if (y[k] == 3) continue; // hole + if (y[k] == 3) + continue; // hole if (id == 51) { h_id[n++] = invalidModuleId; h_id[n++] = invalidModuleId; @@ -233,8 +233,7 @@ int main(void) { ncl = 0; generateClusters(kkk); - std::cout << "created " << n << " digis in " << ncl << " clusters" - << std::endl; + std::cout << "created " << n << " digis in " << ncl << " clusters" << std::endl; assert(n <= numElements); uint32_t nModules = 0; @@ -243,106 +242,108 @@ int main(void) { size_t size16 = n * sizeof(unsigned short); // size_t size8 = n * sizeof(uint8_t); - cudaCheck(cudaMemcpy(d_moduleStart.get(), &nModules, sizeof(uint32_t), - cudaMemcpyHostToDevice)); - cudaCheck( - cudaMemcpy(d_id.get(), h_id.get(), size16, cudaMemcpyHostToDevice)); + cudaCheck(cudaMemcpy(d_moduleStart.get(), &nModules, sizeof(uint32_t), cudaMemcpyHostToDevice)); + cudaCheck(cudaMemcpy(d_id.get(), h_id.get(), size16, cudaMemcpyHostToDevice)); cudaCheck(cudaMemcpy(d_x.get(), h_x.get(), size16, cudaMemcpyHostToDevice)); cudaCheck(cudaMemcpy(d_y.get(), h_y.get(), size16, cudaMemcpyHostToDevice)); - cudaCheck( - cudaMemcpy(d_adc.get(), h_adc.get(), size16, cudaMemcpyHostToDevice)); + cudaCheck(cudaMemcpy(d_adc.get(), h_adc.get(), size16, cudaMemcpyHostToDevice)); // Launch CUDA Kernels int threadsPerBlock = (kkk == 5) ? 512 : ((kkk == 3) ? 128 : 256); int blocksPerGrid = (numElements + threadsPerBlock - 1) / threadsPerBlock; - std::cout << "CUDA countModules kernel launch with " << blocksPerGrid - << " blocks of " << threadsPerBlock << " threads\n"; + std::cout << "CUDA countModules kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock + << " threads\n"; - cms::cuda::launch(countModules, {blocksPerGrid, threadsPerBlock}, - d_id.get(), d_moduleStart.get(), d_clus.get(), n); + cms::cuda::launch(countModules, {blocksPerGrid, threadsPerBlock}, d_id.get(), d_moduleStart.get(), d_clus.get(), n); blocksPerGrid = maxNumModules; // nModules; - std::cout << "CUDA findModules kernel launch with " << blocksPerGrid - << " blocks of " << threadsPerBlock << " threads\n"; - cudaCheck( - cudaMemset(d_clusInModule.get(), 0, maxNumModules * sizeof(uint32_t))); - - cms::cuda::launch(findClus, {blocksPerGrid, threadsPerBlock}, d_id.get(), - d_x.get(), d_y.get(), d_moduleStart.get(), - d_clusInModule.get(), d_moduleId.get(), d_clus.get(), n); + std::cout << "CUDA findModules kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock + << " threads\n"; + cudaCheck(cudaMemset(d_clusInModule.get(), 0, maxNumModules * sizeof(uint32_t))); + + cms::cuda::launch(findClus, + {blocksPerGrid, threadsPerBlock}, + d_id.get(), + d_x.get(), + d_y.get(), + d_moduleStart.get(), + d_clusInModule.get(), + d_moduleId.get(), + d_clus.get(), + n); cudaDeviceSynchronize(); - cudaCheck(cudaMemcpy(&nModules, d_moduleStart.get(), sizeof(uint32_t), - cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(&nModules, d_moduleStart.get(), sizeof(uint32_t), cudaMemcpyDeviceToHost)); uint32_t nclus[maxNumModules], moduleId[nModules]; - cudaCheck(cudaMemcpy(&nclus, d_clusInModule.get(), - maxNumModules * sizeof(uint32_t), - cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(&nclus, d_clusInModule.get(), maxNumModules * sizeof(uint32_t), cudaMemcpyDeviceToHost)); - std::cout << "before charge cut found " - << std::accumulate(nclus, nclus + maxNumModules, 0) << " clusters" + std::cout << "before charge cut found " << std::accumulate(nclus, nclus + maxNumModules, 0) << " clusters" << std::endl; for (auto i = maxNumModules; i > 0; i--) if (nclus[i - 1] > 0) { - std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] - << std::endl; + std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] << std::endl; break; } if (ncl != std::accumulate(nclus, nclus + maxNumModules, 0)) std::cout << "ERROR!!!!! wrong number of cluster found" << std::endl; - cms::cuda::launch(clusterChargeCut, {blocksPerGrid, threadsPerBlock}, - clusterThresholds, d_id.get(), d_adc.get(), - d_moduleStart.get(), d_clusInModule.get(), - d_moduleId.get(), d_clus.get(), n); + cms::cuda::launch(clusterChargeCut, + {blocksPerGrid, threadsPerBlock}, + clusterThresholds, + d_id.get(), + d_adc.get(), + d_moduleStart.get(), + d_clusInModule.get(), + d_moduleId.get(), + d_clus.get(), + n); cudaDeviceSynchronize(); #else // __CUDACC__ h_moduleStart[0] = nModules; countModules(h_id.get(), h_moduleStart.get(), h_clus.get(), n); memset(h_clusInModule.get(), 0, maxNumModules * sizeof(uint32_t)); - findClus(h_id.get(), h_x.get(), h_y.get(), h_moduleStart.get(), - h_clusInModule.get(), h_moduleId.get(), h_clus.get(), n); + findClus( + h_id.get(), h_x.get(), h_y.get(), h_moduleStart.get(), h_clusInModule.get(), h_moduleId.get(), h_clus.get(), n); nModules = h_moduleStart[0]; auto nclus = h_clusInModule.get(); - std::cout << "before charge cut found " - << std::accumulate(nclus, nclus + maxNumModules, 0) << " clusters" + std::cout << "before charge cut found " << std::accumulate(nclus, nclus + maxNumModules, 0) << " clusters" << std::endl; for (auto i = maxNumModules; i > 0; i--) if (nclus[i - 1] > 0) { - std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] - << std::endl; + std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] << std::endl; break; } if (ncl != std::accumulate(nclus, nclus + maxNumModules, 0)) std::cout << "ERROR!!!!! wrong number of cluster found" << std::endl; - clusterChargeCut(clusterThresholds, h_id.get(), h_adc.get(), - h_moduleStart.get(), h_clusInModule.get(), - h_moduleId.get(), h_clus.get(), n); + clusterChargeCut(clusterThresholds, + h_id.get(), + h_adc.get(), + h_moduleStart.get(), + h_clusInModule.get(), + h_moduleId.get(), + h_clus.get(), + n); #endif // __CUDACC__ std::cout << "found " << nModules << " Modules active" << std::endl; #ifdef __CUDACC__ - cudaCheck( - cudaMemcpy(h_id.get(), d_id.get(), size16, cudaMemcpyDeviceToHost)); - cudaCheck( - cudaMemcpy(h_clus.get(), d_clus.get(), size32, cudaMemcpyDeviceToHost)); - cudaCheck(cudaMemcpy(&nclus, d_clusInModule.get(), - maxNumModules * sizeof(uint32_t), - cudaMemcpyDeviceToHost)); - cudaCheck(cudaMemcpy(&moduleId, d_moduleId.get(), - nModules * sizeof(uint32_t), cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(h_id.get(), d_id.get(), size16, cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(h_clus.get(), d_clus.get(), size32, cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(&nclus, d_clusInModule.get(), maxNumModules * sizeof(uint32_t), cudaMemcpyDeviceToHost)); + cudaCheck(cudaMemcpy(&moduleId, d_moduleId.get(), nModules * sizeof(uint32_t), cudaMemcpyDeviceToHost)); #endif // __CUDACC__ std::set clids; for (int i = 0; i < n; ++i) { assert(h_id[i] != 666); // only noise - if (h_id[i] == invalidModuleId) continue; + if (h_id[i] == invalidModuleId) + continue; assert(h_clus[i] >= 0); assert(h_clus[i] < int(nclus[h_id[i]])); clids.insert(h_id[i] * 1000 + h_clus[i]); @@ -355,10 +356,8 @@ int main(void) { assert(0 == (*p) % 1000); auto c = p; ++c; - std::cout << "first clusters " << *p << ' ' << *c << ' ' << nclus[cmid] - << ' ' << nclus[(*c) / 1000] << std::endl; - std::cout << "last cluster " << *clids.rbegin() << ' ' - << nclus[(*clids.rbegin()) / 1000] << std::endl; + std::cout << "first clusters " << *p << ' ' << *c << ' ' << nclus[cmid] << ' ' << nclus[(*c) / 1000] << std::endl; + std::cout << "last cluster " << *clids.rbegin() << ' ' << nclus[(*clids.rbegin()) / 1000] << std::endl; for (; c != clids.end(); ++c) { auto cc = *c; auto pp = *p; @@ -380,12 +379,11 @@ int main(void) { std::cout << "error " << mid << ": " << nc << ' ' << pnc << std::endl; } - std::cout << "found " << std::accumulate(nclus, nclus + maxNumModules, 0) - << ' ' << clids.size() << " clusters" << std::endl; + std::cout << "found " << std::accumulate(nclus, nclus + maxNumModules, 0) << ' ' << clids.size() << " clusters" + << std::endl; for (auto i = maxNumModules; i > 0; i--) if (nclus[i - 1] > 0) { - std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] - << std::endl; + std::cout << "last module is " << i - 1 << ' ' << nclus[i - 1] << std::endl; break; } // << " and " << seeds.size() << " seeds" << std::endl; diff --git a/RecoLocalTracker/SiPixelRecHits/src/PixelCPEClusterRepair.cc b/RecoLocalTracker/SiPixelRecHits/src/PixelCPEClusterRepair.cc index a3503c9d1e34b..906d323935465 100644 --- a/RecoLocalTracker/SiPixelRecHits/src/PixelCPEClusterRepair.cc +++ b/RecoLocalTracker/SiPixelRecHits/src/PixelCPEClusterRepair.cc @@ -27,23 +27,23 @@ using namespace SiPixelTemplateReco; using namespace std; namespace { -constexpr float micronsToCm = 1.0e-4; -constexpr int cluster_matrix_size_x = 13; -constexpr int cluster_matrix_size_y = 21; + constexpr float micronsToCm = 1.0e-4; + constexpr int cluster_matrix_size_x = 13; + constexpr int cluster_matrix_size_y = 21; } // namespace //----------------------------------------------------------------------------- // Constructor. // //----------------------------------------------------------------------------- -PixelCPEClusterRepair::PixelCPEClusterRepair( - edm::ParameterSet const& conf, const MagneticField* mag, - const TrackerGeometry& geom, const TrackerTopology& ttopo, - const SiPixelLorentzAngle* lorentzAngle, - const SiPixelTemplateDBObject* templateDBobject, - const SiPixel2DTemplateDBObject* templateDBobject2D) - : PixelCPEBase(conf, mag, geom, ttopo, lorentzAngle, nullptr, - templateDBobject, nullptr, 1) { +PixelCPEClusterRepair::PixelCPEClusterRepair(edm::ParameterSet const& conf, + const MagneticField* mag, + const TrackerGeometry& geom, + const TrackerTopology& ttopo, + const SiPixelLorentzAngle* lorentzAngle, + const SiPixelTemplateDBObject* templateDBobject, + const SiPixel2DTemplateDBObject* templateDBobject2D) + : PixelCPEBase(conf, mag, geom, ttopo, lorentzAngle, nullptr, templateDBobject, nullptr, 1) { LogDebug("PixelCPEClusterRepair::(constructor)") << endl; //--- Parameter to decide between DB or text file template access @@ -62,8 +62,7 @@ PixelCPEClusterRepair::PixelCPEClusterRepair( "Using SiPixelTemplateDBObject2D version " << (*templateDBobject2D).version() << "\n\n"; } else { - LogDebug("PixelCPEClusterRepair") - << "Loading templates for barrel and forward from ASCII files." << endl; + LogDebug("PixelCPEClusterRepair") << "Loading templates for barrel and forward from ASCII files." << endl; //--- (Archaic) Get configurable template IDs. This code executes only if // we loading pixels from ASCII // files, and then they are mandatory. @@ -71,32 +70,26 @@ PixelCPEClusterRepair::PixelCPEClusterRepair( forwardTemplateID_ = conf.getParameter("forwardTemplateID"); templateDir_ = conf.getParameter("directoryWithTemplates"); - if (!SiPixelTemplate::pushfile(barrelTemplateID_, thePixelTemp_, - templateDir_)) - throw cms::Exception("PixelCPEClusterRepair") - << "\nERROR: Template ID " << barrelTemplateID_ - << " not loaded correctly from text file. Reconstruction will " - "fail.\n\n"; + if (!SiPixelTemplate::pushfile(barrelTemplateID_, thePixelTemp_, templateDir_)) + throw cms::Exception("PixelCPEClusterRepair") << "\nERROR: Template ID " << barrelTemplateID_ + << " not loaded correctly from text file. Reconstruction will " + "fail.\n\n"; - if (!SiPixelTemplate::pushfile(forwardTemplateID_, thePixelTemp_, - templateDir_)) - throw cms::Exception("PixelCPEClusterRepair") - << "\nERROR: Template ID " << forwardTemplateID_ - << " not loaded correctly from text file. Reconstruction will " - "fail.\n\n"; + if (!SiPixelTemplate::pushfile(forwardTemplateID_, thePixelTemp_, templateDir_)) + throw cms::Exception("PixelCPEClusterRepair") << "\nERROR: Template ID " << forwardTemplateID_ + << " not loaded correctly from text file. Reconstruction will " + "fail.\n\n"; } speed_ = conf.getParameter("speed"); - LogDebug("PixelCPEClusterRepair::PixelCPEClusterRepair:") - << "Template speed = " << speed_ << "\n"; + LogDebug("PixelCPEClusterRepair::PixelCPEClusterRepair:") << "Template speed = " << speed_ << "\n"; // this returns the magnetic field value in kgauss (1T = 10 kgauss) int theMagField = mag->nominalValue(); if (theMagField >= 36 && theMagField < 39) { LogDebug("PixelCPEClusterRepair::PixelCPEClusterRepair:") - << "Magnetic field value is: " << theMagField - << " kgauss. Algorithm is being run \n"; + << "Magnetic field value is: " << theMagField << " kgauss. Algorithm is being run \n"; templateDBobject2D_ = templateDBobject2D; fill2DTemplIDs(); @@ -111,8 +104,7 @@ PixelCPEClusterRepair::PixelCPEClusterRepair( // can be: // XYX (XYZ = PXB, PXE) // XYZ n (XYZ as above, n = layer, wheel or disk = 1 .. 6 ;) - std::vector str_recommend2D = - conf.getParameter>("Recommend2D"); + std::vector str_recommend2D = conf.getParameter>("Recommend2D"); recommend2D_.reserve(str_recommend2D.size()); for (auto& str : str_recommend2D) { recommend2D_.push_back(str); @@ -135,18 +127,13 @@ void PixelCPEClusterRepair::fill2DTemplIDs() { unsigned m_detectors = dus.size(); for (unsigned int i = 1; i < 7; ++i) { LogDebug("PixelCPEClusterRepair:: LookingForFirstStrip") - << "Subdetector " << i << " GeomDetEnumerator " - << GeomDetEnumerators::tkDetEnum[i] << " offset " + << "Subdetector " << i << " GeomDetEnumerator " << GeomDetEnumerators::tkDetEnum[i] << " offset " << geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i]) << " is it strip? " << (geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i]) != dus.size() - ? dus[geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i])] - ->type() - .isOuterTracker() + ? dus[geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i])]->type().isOuterTracker() : false); if (geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i]) != dus.size() && - dus[geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i])] - ->type() - .isOuterTracker()) { + dus[geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i])]->type().isOuterTracker()) { if (geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i]) < m_detectors) m_detectors = geom_.offsetDU(GeomDetEnumerators::tkDetEnum[i]); } @@ -154,19 +141,16 @@ void PixelCPEClusterRepair::fill2DTemplIDs() { LogDebug("LookingForFirstStrip") << " Chosen offset: " << m_detectors; m_DetParams.resize(m_detectors); - LogDebug("PixelCPEClusterRepair::fillDetParams():") - << "caching " << m_detectors << " pixel detectors" << endl; + LogDebug("PixelCPEClusterRepair::fillDetParams():") << "caching " << m_detectors << " pixel detectors" << endl; // Loop through detectors and store 2D template ID bool printed_info = false; for (unsigned i = 0; i != m_detectors; ++i) { auto& p = m_DetParams[i]; - p.detTemplateId2D = - templateDBobject2D_->getTemplateID(p.theDet->geographicalId()); + p.detTemplateId2D = templateDBobject2D_->getTemplateID(p.theDet->geographicalId()); if (p.detTemplateId != p.detTemplateId2D && !printed_info) { edm::LogWarning("PixelCPEClusterRepair") - << "different template ID between 1D and 2D " << p.detTemplateId - << " " << p.detTemplateId2D << endl; + << "different template ID between 1D and 2D " << p.detTemplateId << " " << p.detTemplateId2D << endl; printed_info = true; } } @@ -177,8 +161,7 @@ void PixelCPEClusterRepair::fill2DTemplIDs() { //----------------------------------------------------------------------------- PixelCPEClusterRepair::~PixelCPEClusterRepair() {} -std::unique_ptr -PixelCPEClusterRepair::createClusterParam(const SiPixelCluster& cl) const { +std::unique_ptr PixelCPEClusterRepair::createClusterParam(const SiPixelCluster& cl) const { return std::make_unique(cl); } @@ -189,15 +172,12 @@ PixelCPEClusterRepair::createClusterParam(const SiPixelCluster& cl) const { //------------------------------------------------------------------ // The main call to the template code. //------------------------------------------------------------------ -LocalPoint PixelCPEClusterRepair::localPosition( - DetParam const& theDetParam, ClusterParam& theClusterParamBase) const { - ClusterParamTemplate& theClusterParam = - static_cast(theClusterParamBase); +LocalPoint PixelCPEClusterRepair::localPosition(DetParam const& theDetParam, ClusterParam& theClusterParamBase) const { + ClusterParamTemplate& theClusterParam = static_cast(theClusterParamBase); bool filled_from_2d = false; if (!GeomDetEnumerators::isTrackerPixel(theDetParam.thePart)) - throw cms::Exception("PixelCPEClusterRepair::localPosition :") - << "A non-pixel detector type in here?"; + throw cms::Exception("PixelCPEClusterRepair::localPosition :") << "A non-pixel detector type in here?"; int ID1 = -9999; int ID2 = -9999; @@ -235,14 +215,12 @@ LocalPoint PixelCPEClusterRepair::localPosition( LocalPoint lp; if (theClusterParam.with_track_angle) //--- Update call with trk angles needed for bow/kink corrections // Gavril - lp = theDetParam.theTopol->localPosition(MeasurementPoint(tmp_x, tmp_y), - theClusterParam.loc_trk_pred); + lp = theDetParam.theTopol->localPosition(MeasurementPoint(tmp_x, tmp_y), theClusterParam.loc_trk_pred); else { - edm::LogError("PixelCPEClusterRepair") - << "@SUB = PixelCPEClusterRepair::localPosition" - << "Should never be here. PixelCPEClusterRepair should always be " - "called " - "with track angles. This is a bad error !!! "; + edm::LogError("PixelCPEClusterRepair") << "@SUB = PixelCPEClusterRepair::localPosition" + << "Should never be here. PixelCPEClusterRepair should always be " + "called " + "with track angles. This is a bad error !!! "; lp = theDetParam.theTopol->localPosition(MeasurementPoint(tmp_x, tmp_y)); } @@ -281,10 +259,8 @@ LocalPoint PixelCPEClusterRepair::localPosition( //--- Prepare struct that passes pointers to TemplateReco. It doesn't own // anything. - SiPixelTemplateReco::ClusMatrix clusterPayload{&clustMatrix[0][0], xdouble, - ydouble, mrow, mcol}; - SiPixelTemplateReco2D::ClusMatrix clusterPayload2d{ - &clustMatrix2[0][0], xdouble, ydouble, mrow, mcol}; + SiPixelTemplateReco::ClusMatrix clusterPayload{&clustMatrix[0][0], xdouble, ydouble, mrow, mcol}; + SiPixelTemplateReco2D::ClusMatrix clusterPayload2d{&clustMatrix2[0][0], xdouble, ydouble, mrow, mcol}; //--- Copy clust's pixels (calibrated in electrons) into clustMatrix; memset(clustMatrix, 0, sizeof(float) * mrow * mcol); // Wipe it clean. @@ -293,7 +269,8 @@ LocalPoint PixelCPEClusterRepair::localPosition( int irow = int(pix.x) - row_offset; int icol = int(pix.y) - col_offset; // &&& Do we ever get pixels that are out of bounds ??? Need to check. - if ((irow < mrow) & (icol < mcol)) clustMatrix[irow][icol] = float(pix.adc); + if ((irow < mrow) & (icol < mcol)) + clustMatrix[irow][icol] = float(pix.adc); } // &&& Save for later: fillClustMatrix( float * clustMatrix ); @@ -340,19 +317,19 @@ LocalPoint PixelCPEClusterRepair::localPosition( //------------------------------------------------------------------ // Helper function to aggregate call & handling of Template Reco //------------------------------------------------------------------ -void PixelCPEClusterRepair::callTempReco1D( - DetParam const& theDetParam, ClusterParamTemplate& theClusterParam, - SiPixelTemplateReco::ClusMatrix& clusterPayload, int ID, - LocalPoint& lp) const { +void PixelCPEClusterRepair::callTempReco1D(DetParam const& theDetParam, + ClusterParamTemplate& theClusterParam, + SiPixelTemplateReco::ClusMatrix& clusterPayload, + int ID, + LocalPoint& lp) const { SiPixelTemplate templ(thePixelTemp_); // Output: float nonsense = -99999.9f; // nonsense init value - theClusterParam.templXrec_ = theClusterParam.templYrec_ = - theClusterParam.templSigmaX_ = theClusterParam.templSigmaY_ = nonsense; + theClusterParam.templXrec_ = theClusterParam.templYrec_ = theClusterParam.templSigmaX_ = + theClusterParam.templSigmaY_ = nonsense; // If the template recontruction fails, we want to return 1.0 for now - theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = - theClusterParam.probabilityQ_ = 1.f; + theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = theClusterParam.probabilityQ_ = 1.f; theClusterParam.qBin_ = 0; // We have a boolean denoting whether the reco failed or not theClusterParam.hasFilledProb_ = false; @@ -367,13 +344,26 @@ void PixelCPEClusterRepair::callTempReco1D( std::vector> zeropix; int nypix = 0, nxpix = 0; // - theClusterParam.ierr = PixelTempReco1D( - ID, theClusterParam.cotalpha, theClusterParam.cotbeta, locBz, locBx, - clusterPayload, templ, theClusterParam.templYrec_, - theClusterParam.templSigmaY_, theClusterParam.probabilityY_, - theClusterParam.templXrec_, theClusterParam.templSigmaX_, - theClusterParam.probabilityX_, theClusterParam.qBin_, speed_, deadpix, - zeropix, theClusterParam.probabilityQ_, nypix, nxpix); + theClusterParam.ierr = PixelTempReco1D(ID, + theClusterParam.cotalpha, + theClusterParam.cotbeta, + locBz, + locBx, + clusterPayload, + templ, + theClusterParam.templYrec_, + theClusterParam.templSigmaY_, + theClusterParam.probabilityY_, + theClusterParam.templXrec_, + theClusterParam.templSigmaX_, + theClusterParam.probabilityX_, + theClusterParam.qBin_, + speed_, + deadpix, + zeropix, + theClusterParam.probabilityQ_, + nypix, + nxpix); // ****************************************************************** //--- Check exit status @@ -381,8 +371,7 @@ void PixelCPEClusterRepair::callTempReco1D( LogDebug("PixelCPEClusterRepair::localPosition") << "reconstruction failed with error " << theClusterParam.ierr << "\n"; - theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = - theClusterParam.probabilityQ_ = 0.f; + theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = theClusterParam.probabilityQ_ = 0.f; theClusterParam.qBin_ = 0; // Gavril: what do we do in this case ? For now, just return the cluster @@ -396,24 +385,20 @@ void PixelCPEClusterRepair::callTempReco1D( lorentz_drift = 10.0f; // in microns // GG: trk angles needed to correct for bows/kinks if (theClusterParam.with_track_angle) { - theClusterParam.templXrec_ = - theDetParam.theTopol->localX(theClusterParam.theCluster->x(), - theClusterParam.loc_trk_pred) - - lorentz_drift * micronsToCm; // rough Lorentz drift correction - theClusterParam.templYrec_ = theDetParam.theTopol->localY( - theClusterParam.theCluster->y(), theClusterParam.loc_trk_pred); - } else { - edm::LogError("PixelCPEClusterRepair") - << "@SUB = PixelCPEClusterRepair::localPosition" - << "Should never be here. PixelCPEClusterRepair should always be " - "called " - "with track angles. This is a bad error !!! "; - - theClusterParam.templXrec_ = - theDetParam.theTopol->localX(theClusterParam.theCluster->x()) - - lorentz_drift * micronsToCm; // rough Lorentz drift correction + theClusterParam.templXrec_ = theDetParam.theTopol->localX(theClusterParam.theCluster->x(), + theClusterParam.loc_trk_pred) - + lorentz_drift * micronsToCm; // rough Lorentz drift correction theClusterParam.templYrec_ = - theDetParam.theTopol->localY(theClusterParam.theCluster->y()); + theDetParam.theTopol->localY(theClusterParam.theCluster->y(), theClusterParam.loc_trk_pred); + } else { + edm::LogError("PixelCPEClusterRepair") << "@SUB = PixelCPEClusterRepair::localPosition" + << "Should never be here. PixelCPEClusterRepair should always be " + "called " + "with track angles. This is a bad error !!! "; + + theClusterParam.templXrec_ = theDetParam.theTopol->localX(theClusterParam.theCluster->x()) - + lorentz_drift * micronsToCm; // rough Lorentz drift correction + theClusterParam.templYrec_ = theDetParam.theTopol->localY(theClusterParam.theCluster->y()); } } else { //--- Template Reco succeeded. The probabilities are filled. @@ -433,19 +418,19 @@ void PixelCPEClusterRepair::callTempReco1D( //------------------------------------------------------------------ // Helper function to aggregate call & handling of Template 2D fit //------------------------------------------------------------------ -void PixelCPEClusterRepair::callTempReco2D( - DetParam const& theDetParam, ClusterParamTemplate& theClusterParam, - SiPixelTemplateReco2D::ClusMatrix& clusterPayload, int ID, - LocalPoint& lp) const { +void PixelCPEClusterRepair::callTempReco2D(DetParam const& theDetParam, + ClusterParamTemplate& theClusterParam, + SiPixelTemplateReco2D::ClusMatrix& clusterPayload, + int ID, + LocalPoint& lp) const { SiPixelTemplate2D templ2d(thePixelTemp2D_); // Output: float nonsense = -99999.9f; // nonsense init value - theClusterParam.templXrec_ = theClusterParam.templYrec_ = - theClusterParam.templSigmaX_ = theClusterParam.templSigmaY_ = nonsense; + theClusterParam.templXrec_ = theClusterParam.templYrec_ = theClusterParam.templSigmaX_ = + theClusterParam.templSigmaY_ = nonsense; // If the template recontruction fails, we want to return 1.0 for now - theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = - theClusterParam.probabilityQ_ = 1.f; + theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = theClusterParam.probabilityQ_ = 1.f; theClusterParam.qBin_ = 0; // We have a boolean denoting whether the reco failed or not theClusterParam.hasFilledProb_ = false; @@ -483,54 +468,58 @@ void PixelCPEClusterRepair::callTempReco2D( theClusterParam.ierr2 = 8; } else { - theClusterParam.ierr2 = PixelTempReco2D( - ID, theClusterParam.cotalpha, theClusterParam.cotbeta, locBz, locBx, - theClusterParam.edgeTypeY_, theClusterParam.edgeTypeX_, clusterPayload, - templ2d, theClusterParam.templYrec_, theClusterParam.templSigmaY_, - theClusterParam.templXrec_, theClusterParam.templSigmaX_, - theClusterParam.templProbXY_, theClusterParam.probabilityQ_, - theClusterParam.qBin_, deltay, npixels); + theClusterParam.ierr2 = PixelTempReco2D(ID, + theClusterParam.cotalpha, + theClusterParam.cotbeta, + locBz, + locBx, + theClusterParam.edgeTypeY_, + theClusterParam.edgeTypeX_, + clusterPayload, + templ2d, + theClusterParam.templYrec_, + theClusterParam.templSigmaY_, + theClusterParam.templXrec_, + theClusterParam.templSigmaX_, + theClusterParam.templProbXY_, + theClusterParam.probabilityQ_, + theClusterParam.qBin_, + deltay, + npixels); } // ****************************************************************** //--- Check exit status if UNLIKELY (theClusterParam.ierr2 != 0) { LogDebug("PixelCPEClusterRepair::localPosition") - << "2D reconstruction failed with error " << theClusterParam.ierr2 - << "\n"; + << "2D reconstruction failed with error " << theClusterParam.ierr2 << "\n"; - theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = - theClusterParam.probabilityQ_ = 0.f; + theClusterParam.probabilityX_ = theClusterParam.probabilityY_ = theClusterParam.probabilityQ_ = 0.f; theClusterParam.qBin_ = 0; // GG: what do we do in this case? For now, just return the cluster center // of gravity in microns In the x case, apply a rough Lorentz drift average // correction float lorentz_drift = -999.9; if (!GeomDetEnumerators::isEndcap(theDetParam.thePart)) - lorentz_drift = - 60.0f; // in microns // &&& replace with a constant (globally) + lorentz_drift = 60.0f; // in microns // &&& replace with a constant (globally) else lorentz_drift = 10.0f; // in microns // GG: trk angles needed to correct for bows/kinks if (theClusterParam.with_track_angle) { - theClusterParam.templXrec_ = - theDetParam.theTopol->localX(theClusterParam.theCluster->x(), - theClusterParam.loc_trk_pred) - - lorentz_drift * micronsToCm; // rough Lorentz drift correction - theClusterParam.templYrec_ = theDetParam.theTopol->localY( - theClusterParam.theCluster->y(), theClusterParam.loc_trk_pred); - } else { - edm::LogError("PixelCPEClusterRepair") - << "@SUB = PixelCPEClusterRepair::localPosition" - << "Should never be here. PixelCPEClusterRepair should always be " - "called " - "with track angles. This is a bad error !!! "; - - theClusterParam.templXrec_ = - theDetParam.theTopol->localX(theClusterParam.theCluster->x()) - - lorentz_drift * micronsToCm; // rough Lorentz drift correction + theClusterParam.templXrec_ = theDetParam.theTopol->localX(theClusterParam.theCluster->x(), + theClusterParam.loc_trk_pred) - + lorentz_drift * micronsToCm; // rough Lorentz drift correction theClusterParam.templYrec_ = - theDetParam.theTopol->localY(theClusterParam.theCluster->y()); + theDetParam.theTopol->localY(theClusterParam.theCluster->y(), theClusterParam.loc_trk_pred); + } else { + edm::LogError("PixelCPEClusterRepair") << "@SUB = PixelCPEClusterRepair::localPosition" + << "Should never be here. PixelCPEClusterRepair should always be " + "called " + "with track angles. This is a bad error !!! "; + + theClusterParam.templXrec_ = theDetParam.theTopol->localX(theClusterParam.theCluster->x()) - + lorentz_drift * micronsToCm; // rough Lorentz drift correction + theClusterParam.templYrec_ = theDetParam.theTopol->localY(theClusterParam.theCluster->y()); } } else { //--- Template Reco succeeded. @@ -550,9 +539,10 @@ void PixelCPEClusterRepair::callTempReco2D( // Helper function to see if we should recommend 2D reco to be run on this // cluster // --------------------------------------------------------------------------------- -void PixelCPEClusterRepair::checkRecommend2D( - DetParam const& theDetParam, ClusterParamTemplate& theClusterParam, - SiPixelTemplateReco::ClusMatrix& clusterPayload, int ID) const +void PixelCPEClusterRepair::checkRecommend2D(DetParam const& theDetParam, + ClusterParamTemplate& theClusterParam, + SiPixelTemplateReco::ClusMatrix& clusterPayload, + int ID) const { // recommend2D is false by default @@ -563,7 +553,8 @@ void PixelCPEClusterRepair::checkRecommend2D( bool recommend = false; for (auto& rec : recommend2D_) { recommend = rec.recommend(id, ttopo_); - if (recommend) break; + if (recommend) + break; } // only run on those layers recommended by configuration @@ -580,8 +571,7 @@ void PixelCPEClusterRepair::checkRecommend2D( } // The 1d pixel template SiPixelTemplate templ(thePixelTemp_); - if (!templ.interpolate(ID, theClusterParam.cotalpha, theClusterParam.cotbeta, - theDetParam.bz, theDetParam.bx)) { + if (!templ.interpolate(ID, theClusterParam.cotalpha, theClusterParam.cotbeta, theDetParam.bz, theDetParam.bx)) { // error setting up template, return false theClusterParam.recommended2D_ = false; return; @@ -590,7 +580,8 @@ void PixelCPEClusterRepair::checkRecommend2D( // length of the cluster taking into account double sized pixels float nypix = clusterPayload.mcol; for (int i = 0; i < clusterPayload.mcol; i++) { - if (clusterPayload.ydouble[i]) nypix += 1.; + if (clusterPayload.ydouble[i]) + nypix += 1.; } // templ.clsleny() is the expected length of the cluster along y axis. @@ -641,10 +632,8 @@ void PixelCPEClusterRepair::checkRecommend2D( //------------------------------------------------------------------ // localError() relies on localPosition() being called FIRST!!! //------------------------------------------------------------------ -LocalError PixelCPEClusterRepair::localError( - DetParam const& theDetParam, ClusterParam& theClusterParamBase) const { - ClusterParamTemplate& theClusterParam = - static_cast(theClusterParamBase); +LocalError PixelCPEClusterRepair::localError(DetParam const& theDetParam, ClusterParam& theClusterParamBase) const { + ClusterParamTemplate& theClusterParam = static_cast(theClusterParamBase); //--- Default is the maximum error used for edge clusters. //--- (never used, in fact: let comment it out, shut up the complains of the @@ -653,11 +642,9 @@ LocalError PixelCPEClusterRepair::localError( // Check if the errors were already set at the clusters splitting level if (theClusterParam.theCluster->getSplitClusterErrorX() > 0.0f && - theClusterParam.theCluster->getSplitClusterErrorX() < - clusterSplitMaxError_ && + theClusterParam.theCluster->getSplitClusterErrorX() < clusterSplitMaxError_ && theClusterParam.theCluster->getSplitClusterErrorY() > 0.0f && - theClusterParam.theCluster->getSplitClusterErrorY() < - clusterSplitMaxError_) { + theClusterParam.theCluster->getSplitClusterErrorY() < clusterSplitMaxError_) { xerr = theClusterParam.theCluster->getSplitClusterErrorX() * micronsToCm; yerr = theClusterParam.theCluster->getSplitClusterErrorY() * micronsToCm; @@ -675,8 +662,7 @@ LocalError PixelCPEClusterRepair::localError( // errors. // if UNLIKELY (!GeomDetEnumerators::isTrackerPixel(theDetParam.thePart)) - throw cms::Exception("PixelCPEClusterRepair::localPosition :") - << "A non-pixel detector type in here?"; + throw cms::Exception("PixelCPEClusterRepair::localPosition :") << "A non-pixel detector type in here?"; // Assign better errors based on the residuals for failed template cases if (GeomDetEnumerators::isBarrel(theDetParam.thePart)) { @@ -690,8 +676,7 @@ LocalError PixelCPEClusterRepair::localError( // Use special edge hit errors (derived from observed RMS's) for edge hits // that we did not run the 2D reco on // - else if (!theClusterParamBase.filled_from_2d && - (theClusterParam.edgeTypeX_ || theClusterParam.edgeTypeY_)) { + else if (!theClusterParamBase.filled_from_2d && (theClusterParam.edgeTypeX_ || theClusterParam.edgeTypeY_)) { // for edge pixels assign errors according to observed residual RMS if (theClusterParam.edgeTypeX_ && !theClusterParam.edgeTypeY_) { xerr = xEdgeXError_ * micronsToCm; @@ -712,12 +697,11 @@ LocalError PixelCPEClusterRepair::localError( } if (theVerboseLevel > 9) { - LogDebug("PixelCPEClusterRepair") - << " Sizex = " << theClusterParam.theCluster->sizeX() - << " Sizey = " << theClusterParam.theCluster->sizeY() - << " Edgex = " << theClusterParam.edgeTypeX_ - << " Edgey = " << theClusterParam.edgeTypeY_ << " ErrX = " << xerr - << " ErrY = " << yerr; + LogDebug("PixelCPEClusterRepair") << " Sizex = " << theClusterParam.theCluster->sizeX() + << " Sizey = " << theClusterParam.theCluster->sizeY() + << " Edgex = " << theClusterParam.edgeTypeX_ + << " Edgey = " << theClusterParam.edgeTypeY_ << " ErrX = " << xerr + << " ErrY = " << yerr; } if (!(xerr > 0.0f)) @@ -736,8 +720,7 @@ PixelCPEClusterRepair::Rule::Rule(const std::string& str) { boost::cmatch match; // match and check it works if (!regex_match(str.c_str(), match, rule)) - throw cms::Exception("Configuration") - << "Rule '" << str << "' not understood.\n"; + throw cms::Exception("Configuration") << "Rule '" << str << "' not understood.\n"; // subdet subdet_ = -1; @@ -747,8 +730,7 @@ PixelCPEClusterRepair::Rule::Rule(const std::string& str) { subdet_ = PixelSubdetector::PixelEndcap; if (subdet_ == -1) { throw cms::Exception("PixelCPEClusterRepair::Configuration") - << "Detector '" << match[1].first - << "' not understood. Should be PXB, PXE.\n"; + << "Detector '" << match[1].first << "' not understood. Should be PXB, PXE.\n"; } // layer (if present) if (match[3].first != match[3].second) { @@ -758,8 +740,7 @@ PixelCPEClusterRepair::Rule::Rule(const std::string& str) { } } // end Rule::Rule -void PixelCPEClusterRepair::fillPSetDescription( - edm::ParameterSetDescription& desc) { +void PixelCPEClusterRepair::fillPSetDescription(edm::ParameterSetDescription& desc) { desc.add("barrelTemplateID", 0); desc.add("forwardTemplateID", 0); desc.add("directoryWithTemplates", 0); @@ -767,7 +748,6 @@ void PixelCPEClusterRepair::fillPSetDescription( desc.add("UseClusterSplitter", false); desc.add("MaxSizeMismatchInY", 0.3); desc.add("MinChargeRatio", 0.8); - desc.add>("Recommend2D", - {"PXB 2", "PXB 3", "PXB 4"}); + desc.add>("Recommend2D", {"PXB 2", "PXB 3", "PXB 4"}); desc.add("RunDamagedClusters", false); } diff --git a/RecoLocalTracker/SubCollectionProducers/interface/PixelClusterSelectorTopBottom.h b/RecoLocalTracker/SubCollectionProducers/interface/PixelClusterSelectorTopBottom.h index f874c32e9608f..be441a3ec5397 100644 --- a/RecoLocalTracker/SubCollectionProducers/interface/PixelClusterSelectorTopBottom.h +++ b/RecoLocalTracker/SubCollectionProducers/interface/PixelClusterSelectorTopBottom.h @@ -21,21 +21,18 @@ #include "Geometry/TrackerGeometryBuilder/interface/TrackerGeometry.h" class PixelClusterSelectorTopBottom : public edm::global::EDProducer<> { - public: +public: explicit PixelClusterSelectorTopBottom(const edm::ParameterSet& cfg) : tTrackerGeom_(esConsumes()), - token_(consumes( - cfg.getParameter("label"))), + token_(consumes(cfg.getParameter("label"))), y_(cfg.getParameter("y")) { produces(); } - void produce(edm::StreamID, edm::Event& event, - const edm::EventSetup& setup) const override; + void produce(edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; edm::EDGetTokenT token_; double y_; }; diff --git a/RecoLocalTracker/SubCollectionProducers/interface/StripClusterSelectorTopBottom.h b/RecoLocalTracker/SubCollectionProducers/interface/StripClusterSelectorTopBottom.h index 4fee963500a38..3e7abcb3e30f7 100644 --- a/RecoLocalTracker/SubCollectionProducers/interface/StripClusterSelectorTopBottom.h +++ b/RecoLocalTracker/SubCollectionProducers/interface/StripClusterSelectorTopBottom.h @@ -22,21 +22,18 @@ #include "Geometry/TrackerGeometryBuilder/interface/TrackerGeometry.h" class StripClusterSelectorTopBottom : public edm::global::EDProducer<> { - public: +public: explicit StripClusterSelectorTopBottom(const edm::ParameterSet& cfg) : tTrackerGeom_(esConsumes()), - token_(consumes>( - cfg.getParameter("label"))), + token_(consumes>(cfg.getParameter("label"))), y_(cfg.getParameter("y")) { produces>(); } - void produce(edm::StreamID, edm::Event& event, - const edm::EventSetup& setup) const override; + void produce(edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; edm::EDGetTokenT> token_; double y_; }; diff --git a/RecoLocalTracker/SubCollectionProducers/src/HITrackClusterRemover.cc b/RecoLocalTracker/SubCollectionProducers/src/HITrackClusterRemover.cc index 8200ef98f6494..1ad5425f55176 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/HITrackClusterRemover.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/HITrackClusterRemover.cc @@ -34,14 +34,13 @@ // class HITrackClusterRemover : public edm::stream::EDProducer<> { - public: +public: HITrackClusterRemover(const edm::ParameterSet &iConfig); ~HITrackClusterRemover() override; void produce(edm::Event &iEvent, const edm::EventSetup &iSetup) override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; struct ParamBlock { ParamBlock() : isSet_(false), usesCharge_(false) {} ParamBlock(const edm::ParameterSet &iConfig) @@ -51,14 +50,9 @@ class HITrackClusterRemover : public edm::stream::EDProducer<> { cutOnPixelCharge_(iConfig.exists("minGoodPixelCharge")), cutOnStripCharge_(iConfig.exists("minGoodStripCharge")), maxChi2_(iConfig.getParameter("maxChi2")), - maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") - : 0), - minGoodPixelCharge_(cutOnPixelCharge_ ? iConfig.getParameter( - "minGoodPixelCharge") - : 0), - minGoodStripCharge_(cutOnStripCharge_ ? iConfig.getParameter( - "minGoodStripCharge") - : 0), + maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") : 0), + minGoodPixelCharge_(cutOnPixelCharge_ ? iConfig.getParameter("minGoodPixelCharge") : 0), + minGoodStripCharge_(cutOnStripCharge_ ? iConfig.getParameter("minGoodStripCharge") : 0), maxSize_(usesSize_ ? iConfig.getParameter("maxSize") : 0) {} bool isSet_, usesCharge_, usesSize_, cutOnPixelCharge_, cutOnStripCharge_; float maxChi2_, maxCharge_, minGoodPixelCharge_, minGoodStripCharge_; @@ -69,10 +63,8 @@ class HITrackClusterRemover : public edm::stream::EDProducer<> { bool doTracks_; bool doStrip_, doPixel_; bool mergeOld_; - typedef edm::ContainerMask > - PixelMaskContainer; - typedef edm::ContainerMask > - StripMaskContainer; + typedef edm::ContainerMask > PixelMaskContainer; + typedef edm::ContainerMask > StripMaskContainer; edm::EDGetTokenT > pixelClusters_; edm::EDGetTokenT > stripClusters_; edm::EDGetTokenT tracks_; @@ -80,34 +72,34 @@ class HITrackClusterRemover : public edm::stream::EDProducer<> { edm::EDGetTokenT oldPxlMaskToken_; edm::EDGetTokenT oldStrMaskToken_; std::vector > > overrideTrkQuals_; - edm::EDGetTokenT rphiRecHitToken_, - stereoRecHitToken_; + edm::EDGetTokenT rphiRecHitToken_, stereoRecHitToken_; // edm::EDGetTokenT pixelRecHitsToken_; ParamBlock pblocks_[NumberOfParamBlocks]; - void readPSet(const edm::ParameterSet &iConfig, const std::string &name, - int id1 = -1, int id2 = -1, int id3 = -1, int id4 = -1, - int id5 = -1, int id6 = -1); + void readPSet(const edm::ParameterSet &iConfig, + const std::string &name, + int id1 = -1, + int id2 = -1, + int id3 = -1, + int id4 = -1, + int id5 = -1, + int id6 = -1); std::vector pixels, strips; // avoid unneed alloc/dealloc of this edm::ProductID pixelSourceProdID, stripSourceProdID; // ProdIDs refs must point to (for consistency tests) - inline void process(const TrackingRecHit *hit, unsigned char chi2, - const TrackerGeometry *tg); - inline void process(const OmniClusterRef &cluRef, SiStripDetId &detid, - bool fromTrack); + inline void process(const TrackingRecHit *hit, unsigned char chi2, const TrackerGeometry *tg); + inline void process(const OmniClusterRef &cluRef, SiStripDetId &detid, bool fromTrack); template - std::unique_ptr > cleanup( - const edmNew::DetSetVector &oldClusters, - const std::vector &isGood, - reco::ClusterRemovalInfo::Indices &refs, - const reco::ClusterRemovalInfo::Indices *oldRefs); + std::unique_ptr > cleanup(const edmNew::DetSetVector &oldClusters, + const std::vector &isGood, + reco::ClusterRemovalInfo::Indices &refs, + const reco::ClusterRemovalInfo::Indices *oldRefs); // Carries in full removal info about a given det from oldRefs - void mergeOld(reco::ClusterRemovalInfo::Indices &refs, - const reco::ClusterRemovalInfo::Indices &oldRefs); + void mergeOld(reco::ClusterRemovalInfo::Indices &refs, const reco::ClusterRemovalInfo::Indices &oldRefs); bool clusterWasteSolution_, doStripChargeCheck_, doPixelChargeCheck_; std::string stripRecHits_, pixelRecHits_; @@ -124,20 +116,24 @@ using namespace std; using namespace edm; using namespace reco; -void HITrackClusterRemover::readPSet(const edm::ParameterSet &iConfig, - const std::string &name, int id1, int id2, - int id3, int id4, int id5, int id6) { +void HITrackClusterRemover::readPSet( + const edm::ParameterSet &iConfig, const std::string &name, int id1, int id2, int id3, int id4, int id5, int id6) { if (iConfig.exists(name)) { ParamBlock pblock(iConfig.getParameter(name)); if (id1 == -1) { fill(pblocks_, pblocks_ + NumberOfParamBlocks, pblock); } else { pblocks_[id1] = pblock; - if (id2 != -1) pblocks_[id2] = pblock; - if (id3 != -1) pblocks_[id3] = pblock; - if (id4 != -1) pblocks_[id4] = pblock; - if (id5 != -1) pblocks_[id5] = pblock; - if (id6 != -1) pblocks_[id6] = pblock; + if (id2 != -1) + pblocks_[id2] = pblock; + if (id3 != -1) + pblocks_[id3] = pblock; + if (id4 != -1) + pblocks_[id4] = pblock; + if (id5 != -1) + pblocks_[id5] = pblock; + if (id6 != -1) + pblocks_[id6] = pblock; } } } @@ -145,48 +141,36 @@ void HITrackClusterRemover::readPSet(const edm::ParameterSet &iConfig, HITrackClusterRemover::HITrackClusterRemover(const ParameterSet &iConfig) : tTrackerGeom_(esConsumes()), doTracks_(iConfig.exists("trajectories")), - doStrip_(iConfig.existsAs("doStrip") - ? iConfig.getParameter("doStrip") - : true), - doPixel_(iConfig.existsAs("doPixel") - ? iConfig.getParameter("doPixel") - : true), + doStrip_(iConfig.existsAs("doStrip") ? iConfig.getParameter("doStrip") : true), + doPixel_(iConfig.existsAs("doPixel") ? iConfig.getParameter("doPixel") : true), mergeOld_(iConfig.exists("oldClusterRemovalInfo")), clusterWasteSolution_(true), - doStripChargeCheck_(iConfig.existsAs("doStripChargeCheck") - ? iConfig.getParameter("doStripChargeCheck") - : false), - doPixelChargeCheck_(iConfig.existsAs("doPixelChargeCheck") - ? iConfig.getParameter("doPixelChargeCheck") - : false), - stripRecHits_(doStripChargeCheck_ - ? iConfig.getParameter("stripRecHits") - : std::string("siStripMatchedRecHits")), - pixelRecHits_(doPixelChargeCheck_ - ? iConfig.getParameter("pixelRecHits") - : std::string("siPixelRecHits")) { - mergeOld_ = - mergeOld_ && - !iConfig.getParameter("oldClusterRemovalInfo").label().empty(); + doStripChargeCheck_( + iConfig.existsAs("doStripChargeCheck") ? iConfig.getParameter("doStripChargeCheck") : false), + doPixelChargeCheck_( + iConfig.existsAs("doPixelChargeCheck") ? iConfig.getParameter("doPixelChargeCheck") : false), + stripRecHits_(doStripChargeCheck_ ? iConfig.getParameter("stripRecHits") + : std::string("siStripMatchedRecHits")), + pixelRecHits_(doPixelChargeCheck_ ? iConfig.getParameter("pixelRecHits") + : std::string("siPixelRecHits")) { + mergeOld_ = mergeOld_ && !iConfig.getParameter("oldClusterRemovalInfo").label().empty(); if (iConfig.exists("overrideTrkQuals")) - overrideTrkQuals_.push_back(consumes >( - iConfig.getParameter("overrideTrkQuals"))); + overrideTrkQuals_.push_back(consumes >(iConfig.getParameter("overrideTrkQuals"))); if (iConfig.exists("clusterLessSolution")) clusterWasteSolution_ = !iConfig.getParameter("clusterLessSolution"); if ((doPixelChargeCheck_ && !doPixel_) || (doStripChargeCheck_ && !doStrip_)) - throw cms::Exception("Configuration Error") - << "HITrackClusterRemover: Charge check asked without cluster " - "collection "; + throw cms::Exception("Configuration Error") << "HITrackClusterRemover: Charge check asked without cluster " + "collection "; if (doPixelChargeCheck_) - throw cms::Exception("Configuration Error") - << "HITrackClusterRemover: Pixel cluster charge check not yet " - "implemented"; + throw cms::Exception("Configuration Error") << "HITrackClusterRemover: Pixel cluster charge check not yet " + "implemented"; if (doPixel_ && clusterWasteSolution_) produces >(); if (doStrip_ && clusterWasteSolution_) produces >(); - if (clusterWasteSolution_) produces(); + if (clusterWasteSolution_) + produces(); assert(!clusterWasteSolution_); @@ -210,14 +194,12 @@ HITrackClusterRemover::HITrackClusterRemover(const ParameterSet &iConfig) bool usingCharge = false; for (size_t i = 0; i < NumberOfParamBlocks; ++i) { if (!pblocks_[i].isSet_) - throw cms::Exception("Configuration Error") - << "HITrackClusterRemover: Missing configuration for detector with " - "subDetID = " - << (i + 1); + throw cms::Exception("Configuration Error") << "HITrackClusterRemover: Missing configuration for detector with " + "subDetID = " + << (i + 1); if (pblocks_[i].usesCharge_ && !usingCharge) { - throw cms::Exception("Configuration Error") - << "HITrackClusterRemover: Configuration for subDetID = " << (i + 1) - << " uses cluster charge, which is not enabled."; + throw cms::Exception("Configuration Error") << "HITrackClusterRemover: Configuration for subDetID = " << (i + 1) + << " uses cluster charge, which is not enabled."; } } @@ -229,38 +211,28 @@ HITrackClusterRemover::HITrackClusterRemover(const ParameterSet &iConfig) filterTracks_ = false; if (iConfig.exists("TrackQuality")) { filterTracks_ = true; - trackQuality_ = reco::TrackBase::qualityByName( - iConfig.getParameter("TrackQuality")); + trackQuality_ = reco::TrackBase::qualityByName(iConfig.getParameter("TrackQuality")); minNumberOfLayersWithMeasBeforeFiltering_ = iConfig.existsAs("minNumberOfLayersWithMeasBeforeFiltering") - ? iConfig.getParameter( - "minNumberOfLayersWithMeasBeforeFiltering") + ? iConfig.getParameter("minNumberOfLayersWithMeasBeforeFiltering") : 0; } if (doTracks_) - tracks_ = consumes( - iConfig.getParameter("trajectories")); + tracks_ = consumes(iConfig.getParameter("trajectories")); if (doPixel_) - pixelClusters_ = consumes >( - iConfig.getParameter("pixelClusters")); + pixelClusters_ = consumes >(iConfig.getParameter("pixelClusters")); if (doStrip_) - stripClusters_ = consumes >( - iConfig.getParameter("stripClusters")); + stripClusters_ = consumes >(iConfig.getParameter("stripClusters")); if (mergeOld_) { - oldRemovalInfo_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldPxlMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldStrMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); + oldRemovalInfo_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldPxlMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldStrMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); } if (doStripChargeCheck_) { - rphiRecHitToken_ = consumes( - InputTag(stripRecHits_, "rphiRecHit")); - stereoRecHitToken_ = consumes( - InputTag(stripRecHits_, "stereoRecHit")); + rphiRecHitToken_ = consumes(InputTag(stripRecHits_, "rphiRecHit")); + stereoRecHitToken_ = consumes(InputTag(stripRecHits_, "stereoRecHit")); } // if(doPixelChargeCheck_) pixelRecHitsToken_ = // consumes(InputTag(pixelRecHits_)); @@ -268,9 +240,7 @@ HITrackClusterRemover::HITrackClusterRemover(const ParameterSet &iConfig) HITrackClusterRemover::~HITrackClusterRemover() {} -void HITrackClusterRemover::mergeOld( - ClusterRemovalInfo::Indices &refs, - const ClusterRemovalInfo::Indices &oldRefs) { +void HITrackClusterRemover::mergeOld(ClusterRemovalInfo::Indices &refs, const ClusterRemovalInfo::Indices &oldRefs) { for (size_t i = 0, n = refs.size(); i < n; ++i) { refs[i] = oldRefs[refs[i]]; } @@ -279,7 +249,8 @@ void HITrackClusterRemover::mergeOld( template std::unique_ptr > HITrackClusterRemover::cleanup( const edmNew::DetSetVector &oldClusters, - const std::vector &isGood, reco::ClusterRemovalInfo::Indices &refs, + const std::vector &isGood, + reco::ClusterRemovalInfo::Indices &refs, const reco::ClusterRemovalInfo::Indices *oldRefs) { typedef typename edmNew::DetSetVector DSV; typedef typename edmNew::DetSetVector::FastFiller DSF; @@ -292,18 +263,16 @@ std::unique_ptr > HITrackClusterRemover::cleanup( // cluster removal loop const T *firstOffset = &oldClusters.data().front(); - for (typename DSV::const_iterator itdet = oldClusters.begin(), - enddet = oldClusters.end(); - itdet != enddet; ++itdet) { + for (typename DSV::const_iterator itdet = oldClusters.begin(), enddet = oldClusters.end(); itdet != enddet; ++itdet) { DS oldDS = *itdet; - if (oldDS.empty()) continue; // skip empty detsets + if (oldDS.empty()) + continue; // skip empty detsets uint32_t id = oldDS.detId(); DSF outds(*output, id); - for (typename DS::const_iterator it = oldDS.begin(), ed = oldDS.end(); - it != ed; ++it) { + for (typename DS::const_iterator it = oldDS.begin(), ed = oldDS.end(); it != ed; ++it) { uint32_t index = ((&*it) - firstOffset); countOld++; if (isGood[index]) { @@ -314,18 +283,20 @@ std::unique_ptr > HITrackClusterRemover::cleanup( // " reference " << index << " to " << (refs.size() - 1) << std::endl; } } - if (outds.empty()) outds.abort(); // not write in an empty DSV + if (outds.empty()) + outds.abort(); // not write in an empty DSV } - if (oldRefs != nullptr) mergeOld(refs, *oldRefs); + if (oldRefs != nullptr) + mergeOld(refs, *oldRefs); return output; } float HITrackClusterRemover::sensorThickness(const SiStripDetId &detid) const { if (detid.subdetId() >= SiStripDetId::TIB) { - if (detid.subdetId() == SiStripDetId::TOB) return 0.047; - if (detid.moduleGeometry() == SiStripModuleGeometry::W5 || - detid.moduleGeometry() == SiStripModuleGeometry::W6 || + if (detid.subdetId() == SiStripDetId::TOB) + return 0.047; + if (detid.moduleGeometry() == SiStripModuleGeometry::W5 || detid.moduleGeometry() == SiStripModuleGeometry::W6 || detid.moduleGeometry() == SiStripModuleGeometry::W7) return 0.047; return 0.029; // so it is TEC ring 1-4 or TIB or TOB; @@ -335,50 +306,45 @@ float HITrackClusterRemover::sensorThickness(const SiStripDetId &detid) const { return 0.027; } -void HITrackClusterRemover::process(OmniClusterRef const &ocluster, - SiStripDetId &detid, bool fromTrack) { +void HITrackClusterRemover::process(OmniClusterRef const &ocluster, SiStripDetId &detid, bool fromTrack) { SiStripRecHit2D::ClusterRef cluster = ocluster.cluster_strip(); if (cluster.id() != stripSourceProdID) throw cms::Exception("Inconsistent Data") - << "HITrackClusterRemover: strip cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << stripSourceProdID << ")\n."; + << "HITrackClusterRemover: strip cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << stripSourceProdID << ")\n."; uint32_t subdet = detid.subdetId(); assert(cluster.id() == stripSourceProdID); - if (pblocks_[subdet - 1].usesSize_ && - (cluster->amplitudes().size() > pblocks_[subdet - 1].maxSize_)) + if (pblocks_[subdet - 1].usesSize_ && (cluster->amplitudes().size() > pblocks_[subdet - 1].maxSize_)) return; if (!fromTrack) { if (pblocks_[subdet - 1].cutOnStripCharge_ && - (cluster->charge() > - (pblocks_[subdet - 1].minGoodStripCharge_ * sensorThickness(detid)))) + (cluster->charge() > (pblocks_[subdet - 1].minGoodStripCharge_ * sensorThickness(detid)))) return; } if (collectedStrips_.size() <= cluster.key()) - edm::LogError("BadCollectionSize") - << collectedStrips_.size() << " is smaller than " << cluster.key(); + edm::LogError("BadCollectionSize") << collectedStrips_.size() << " is smaller than " << cluster.key(); assert(collectedStrips_.size() > cluster.key()); strips[cluster.key()] = false; - if (!clusterWasteSolution_) collectedStrips_[cluster.key()] = true; + if (!clusterWasteSolution_) + collectedStrips_[cluster.key()] = true; } -void HITrackClusterRemover::process(const TrackingRecHit *hit, - unsigned char chi2, - const TrackerGeometry *tg) { +void HITrackClusterRemover::process(const TrackingRecHit *hit, unsigned char chi2, const TrackerGeometry *tg) { SiStripDetId detid = hit->geographicalId(); uint32_t subdet = detid.subdetId(); assert((subdet > 0) && (subdet <= NumberOfParamBlocks)); // chi2 cut - if (chi2 > Traj2TrackHits::toChi2x5(pblocks_[subdet - 1].maxChi2_)) return; + if (chi2 > Traj2TrackHits::toChi2x5(pblocks_[subdet - 1].maxChi2_)) + return; if (GeomDetEnumerators::isTrackerPixel(tg->geomDetSubDetector(subdet))) { - if (!doPixel_) return; + if (!doPixel_) + return; // this is a pixel, and i *know* it is const SiPixelRecHit *pixelHit = static_cast(hit); @@ -386,10 +352,8 @@ void HITrackClusterRemover::process(const TrackingRecHit *hit, if (cluster.id() != pixelSourceProdID) throw cms::Exception("Inconsistent Data") - << "HITrackClusterRemover: pixel cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << pixelSourceProdID << ")\n."; + << "HITrackClusterRemover: pixel cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << pixelSourceProdID << ")\n."; assert(cluster.id() == pixelSourceProdID); // DBG// cout << "HIT NEW PIXEL DETID = " << detid.rawId() << ", Cluster [ " @@ -397,8 +361,7 @@ void HITrackClusterRemover::process(const TrackingRecHit *hit, // endl; // if requested, cut on cluster size - if (pblocks_[subdet - 1].usesSize_ && - (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) + if (pblocks_[subdet - 1].usesSize_ && (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) return; // mark as used @@ -408,34 +371,32 @@ void HITrackClusterRemover::process(const TrackingRecHit *hit, assert(collectedPixels_.size() > cluster.key()); // assert(detid.rawId() == cluster->geographicalId()); //This condition // fails - if (!clusterWasteSolution_) collectedPixels_[cluster.key()] = true; + if (!clusterWasteSolution_) + collectedPixels_[cluster.key()] = true; } else { // aka Strip - if (!doStrip_) return; + if (!doStrip_) + return; const type_info &hitType = typeid(*hit); if (hitType == typeid(SiStripRecHit2D)) { - const SiStripRecHit2D *stripHit = - static_cast(hit); + const SiStripRecHit2D *stripHit = static_cast(hit); // DBG// cout << "Plain RecHit 2D: " << endl; process(stripHit->omniClusterRef(), detid, true); } else if (hitType == typeid(SiStripRecHit1D)) { const SiStripRecHit1D *hit1D = static_cast(hit); process(hit1D->omniClusterRef(), detid, true); } else if (hitType == typeid(SiStripMatchedRecHit2D)) { - const SiStripMatchedRecHit2D *matchHit = - static_cast(hit); + const SiStripMatchedRecHit2D *matchHit = static_cast(hit); // DBG// cout << "Matched RecHit 2D: " << endl; process(matchHit->monoClusterRef(), detid, true); process(matchHit->stereoClusterRef(), detid, true); } else if (hitType == typeid(ProjectedSiStripRecHit2D)) { - const ProjectedSiStripRecHit2D *projHit = - static_cast(hit); + const ProjectedSiStripRecHit2D *projHit = static_cast(hit); // DBG// cout << "Projected RecHit 2D: " << endl; process(projHit->originalHit().omniClusterRef(), detid, true); } else throw cms::Exception("NOT IMPLEMENTED") - << "Don't know how to handle " << hitType.name() << " on detid " - << detid.rawId() << "\n"; + << "Don't know how to handle " << hitType.name() << " on detid " << detid.rawId() << "\n"; } } @@ -497,10 +458,9 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { labelsForToken(oldRemovalInfo_, labels); throw cms::Exception("Inconsistent Data") << "HITrackClusterRemover: " - << "Input collection product IDs are [pixel: " << pixelClusters.id() - << ", strip: " << stripClusters.id() << "] \n" - << "\t but the *old* ClusterRemovalInfo " << labels.productInstance - << " refers as 'new product ids' to " + << "Input collection product IDs are [pixel: " << pixelClusters.id() << ", strip: " << stripClusters.id() + << "] \n" + << "\t but the *old* ClusterRemovalInfo " << labels.productInstance << " refers as 'new product ids' to " << "[pixel: " << oldRemovalInfo->pixelNewRefProd().id() << ", strip: " << oldRemovalInfo->stripNewRefProd().id() << "]\n" << "NOTA BENE: when running HITrackClusterRemover with an old " @@ -525,9 +485,8 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { edm::Handle oldStrMask; iEvent.getByToken(oldPxlMaskToken_, oldPxlMask); iEvent.getByToken(oldStrMaskToken_, oldStrMask); - LogDebug("HITrackClusterRemover") - << "to merge in, " << oldStrMask->size() << " strp and " - << oldPxlMask->size() << " pxl"; + LogDebug("HITrackClusterRemover") << "to merge in, " << oldStrMask->size() << " strp and " << oldPxlMask->size() + << " pxl"; oldStrMask->copyMaskTo(collectedStrips_); oldPxlMask->copyMaskTo(collectedPixels_); assert(stripClusters->dataSize() >= collectedStrips_.size()); @@ -561,9 +520,9 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { goodTk = (qual & (1 << trackQuality_)) >> trackQuality_; } else goodTk = (track.quality(trackQuality_)); - if (!goodTk) continue; - if (track.hitPattern().trackerLayersWithMeasurement() < - minNumberOfLayersWithMeasBeforeFiltering_) + if (!goodTk) + continue; + if (track.hitPattern().trackerLayersWithMeasurement() < minNumberOfLayersWithMeasBeforeFiltering_) continue; } auto const &chi2sX5 = track.extra()->chi2sX5(); @@ -571,7 +530,8 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { auto hb = track.recHitsBegin(); for (unsigned int h = 0; h < track.recHitsSize(); h++) { auto hit = *(hb + h); - if (!hit->isValid()) continue; + if (!hit->isValid()) + continue; process(hit, chi2sX5[h], tgh); } } @@ -580,21 +540,19 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { if (doStripChargeCheck_) { edm::Handle rechitsrphi; iEvent.getByToken(rphiRecHitToken_, rechitsrphi); - const SiStripRecHit2DCollection::DataContainer *rphiRecHits = - &(rechitsrphi).product()->data(); - for (SiStripRecHit2DCollection::DataContainer::const_iterator recHit = - rphiRecHits->begin(); - recHit != rphiRecHits->end(); recHit++) { + const SiStripRecHit2DCollection::DataContainer *rphiRecHits = &(rechitsrphi).product()->data(); + for (SiStripRecHit2DCollection::DataContainer::const_iterator recHit = rphiRecHits->begin(); + recHit != rphiRecHits->end(); + recHit++) { SiStripDetId detid = recHit->geographicalId(); process(recHit->omniClusterRef(), detid, false); } edm::Handle rechitsstereo; iEvent.getByToken(stereoRecHitToken_, rechitsstereo); - const SiStripRecHit2DCollection::DataContainer *stereoRecHits = - &(rechitsstereo).product()->data(); - for (SiStripRecHit2DCollection::DataContainer::const_iterator recHit = - stereoRecHits->begin(); - recHit != stereoRecHits->end(); recHit++) { + const SiStripRecHit2DCollection::DataContainer *stereoRecHits = &(rechitsstereo).product()->data(); + for (SiStripRecHit2DCollection::DataContainer::const_iterator recHit = stereoRecHits->begin(); + recHit != stereoRecHits->end(); + recHit++) { SiStripDetId detid = recHit->geographicalId(); process(recHit->omniClusterRef(), detid, false); } @@ -606,16 +564,14 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { if (doPixel_ && clusterWasteSolution_) { OrphanHandle > newPixels = iEvent.put( - cleanup(*pixelClusters, pixels, cri->pixelIndices(), - mergeOld_ ? &oldRemovalInfo->pixelIndices() : nullptr)); + cleanup(*pixelClusters, pixels, cri->pixelIndices(), mergeOld_ ? &oldRemovalInfo->pixelIndices() : nullptr)); // DBG// std::cout << "HITrackClusterRemover: Wrote pixel " << // newPixels.id() << " from " << pixelSourceProdID << std::endl; cri->setNewPixelClusters(newPixels); } if (doStrip_ && clusterWasteSolution_) { OrphanHandle > newStrips = iEvent.put( - cleanup(*stripClusters, strips, cri->stripIndices(), - mergeOld_ ? &oldRemovalInfo->stripIndices() : nullptr)); + cleanup(*stripClusters, strips, cri->stripIndices(), mergeOld_ ? &oldRemovalInfo->stripIndices() : nullptr)); // DBG// std::cout << "HITrackClusterRemover: Wrote strip " << // newStrips.id() << " from " << stripSourceProdID << std::endl; cri->setNewStripClusters(newStrips); @@ -640,22 +596,18 @@ void HITrackClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { // removedStripClsuterRefs(new // edmNew::DetSetVector()); - LogDebug("HITrackClusterRemover") - << "total strip to skip: " - << std::count(collectedStrips_.begin(), collectedStrips_.end(), true); + LogDebug("HITrackClusterRemover") << "total strip to skip: " + << std::count(collectedStrips_.begin(), collectedStrips_.end(), true); // std::cout << "HITrackClusterRemover " <<"total strip to skip: // "<( - edm::RefProd >(stripClusters), - collectedStrips_)); - - LogDebug("HITrackClusterRemover") - << "total pxl to skip: " - << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); - iEvent.put(std::make_unique( - edm::RefProd >(pixelClusters), - collectedPixels_)); + iEvent.put(std::make_unique(edm::RefProd >(stripClusters), + collectedStrips_)); + + LogDebug("HITrackClusterRemover") << "total pxl to skip: " + << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); + iEvent.put(std::make_unique(edm::RefProd >(pixelClusters), + collectedPixels_)); } collectedStrips_.clear(); collectedPixels_.clear(); diff --git a/RecoLocalTracker/SubCollectionProducers/src/HLTTrackClusterRemoverNew.cc b/RecoLocalTracker/SubCollectionProducers/src/HLTTrackClusterRemoverNew.cc index b53e30fa6fe02..26b833660cb0a 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/HLTTrackClusterRemoverNew.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/HLTTrackClusterRemoverNew.cc @@ -32,14 +32,13 @@ // class HLTTrackClusterRemoverNew final : public edm::stream::EDProducer<> { - public: +public: HLTTrackClusterRemoverNew(const edm::ParameterSet &iConfig); ~HLTTrackClusterRemoverNew() override; void produce(edm::Event &iEvent, const edm::EventSetup &iSetup) override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; struct ParamBlock { ParamBlock() : isSet_(false), usesCharge_(false) {} ParamBlock(const edm::ParameterSet &iConfig) @@ -49,14 +48,9 @@ class HLTTrackClusterRemoverNew final : public edm::stream::EDProducer<> { cutOnPixelCharge_(iConfig.exists("minGoodPixelCharge")), cutOnStripCharge_(iConfig.exists("minGoodStripCharge")), maxChi2_(iConfig.getParameter("maxChi2")), - maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") - : 0), - minGoodPixelCharge_(cutOnPixelCharge_ ? iConfig.getParameter( - "minGoodPixelCharge") - : 0), - minGoodStripCharge_(cutOnStripCharge_ ? iConfig.getParameter( - "minGoodStripCharge") - : 0), + maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") : 0), + minGoodPixelCharge_(cutOnPixelCharge_ ? iConfig.getParameter("minGoodPixelCharge") : 0), + minGoodStripCharge_(cutOnStripCharge_ ? iConfig.getParameter("minGoodStripCharge") : 0), maxSize_(usesSize_ ? iConfig.getParameter("maxSize") : 0) {} bool isSet_, usesCharge_, usesSize_, cutOnPixelCharge_, cutOnStripCharge_; float maxChi2_, maxCharge_, minGoodPixelCharge_, minGoodStripCharge_; @@ -68,10 +62,8 @@ class HLTTrackClusterRemoverNew final : public edm::stream::EDProducer<> { bool doStrip_, doPixel_; bool mergeOld_; - typedef edm::ContainerMask > - PixelMaskContainer; - typedef edm::ContainerMask > - StripMaskContainer; + typedef edm::ContainerMask > PixelMaskContainer; + typedef edm::ContainerMask > StripMaskContainer; edm::EDGetTokenT > pixelClusters_; edm::EDGetTokenT > stripClusters_; edm::EDGetTokenT oldPxlMaskToken_; @@ -79,28 +71,30 @@ class HLTTrackClusterRemoverNew final : public edm::stream::EDProducer<> { edm::EDGetTokenT > trajectories_; ParamBlock pblocks_[NumberOfParamBlocks]; - void readPSet(const edm::ParameterSet &iConfig, const std::string &name, - int id1 = -1, int id2 = -1, int id3 = -1, int id4 = -1, - int id5 = -1, int id6 = -1); + void readPSet(const edm::ParameterSet &iConfig, + const std::string &name, + int id1 = -1, + int id2 = -1, + int id3 = -1, + int id4 = -1, + int id5 = -1, + int id6 = -1); std::vector pixels, strips; // avoid unneed alloc/dealloc of this edm::ProductID pixelSourceProdID, stripSourceProdID; // ProdIDs refs must point to (for consistency tests) - inline void process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg); + inline void process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg); inline void process(const OmniClusterRef &cluRef, uint32_t subdet); template - std::unique_ptr > cleanup( - const edmNew::DetSetVector &oldClusters, - const std::vector &isGood, - reco::ClusterRemovalInfo::Indices &refs, - const reco::ClusterRemovalInfo::Indices *oldRefs); + std::unique_ptr > cleanup(const edmNew::DetSetVector &oldClusters, + const std::vector &isGood, + reco::ClusterRemovalInfo::Indices &refs, + const reco::ClusterRemovalInfo::Indices *oldRefs); // Carries in full removal info about a given det from oldRefs - void mergeOld(reco::ClusterRemovalInfo::Indices &refs, - const reco::ClusterRemovalInfo::Indices &oldRefs); + void mergeOld(reco::ClusterRemovalInfo::Indices &refs, const reco::ClusterRemovalInfo::Indices &oldRefs); bool makeProducts_; bool doStripChargeCheck_, doPixelChargeCheck_; @@ -112,63 +106,54 @@ using namespace std; using namespace edm; using namespace reco; -void HLTTrackClusterRemoverNew::readPSet(const edm::ParameterSet &iConfig, - const std::string &name, int id1, - int id2, int id3, int id4, int id5, - int id6) { +void HLTTrackClusterRemoverNew::readPSet( + const edm::ParameterSet &iConfig, const std::string &name, int id1, int id2, int id3, int id4, int id5, int id6) { if (iConfig.exists(name)) { ParamBlock pblock(iConfig.getParameter(name)); if (id1 == -1) { fill(pblocks_, pblocks_ + NumberOfParamBlocks, pblock); } else { pblocks_[id1] = pblock; - if (id2 != -1) pblocks_[id2] = pblock; - if (id3 != -1) pblocks_[id3] = pblock; - if (id4 != -1) pblocks_[id4] = pblock; - if (id5 != -1) pblocks_[id5] = pblock; - if (id6 != -1) pblocks_[id6] = pblock; + if (id2 != -1) + pblocks_[id2] = pblock; + if (id3 != -1) + pblocks_[id3] = pblock; + if (id4 != -1) + pblocks_[id4] = pblock; + if (id5 != -1) + pblocks_[id5] = pblock; + if (id6 != -1) + pblocks_[id6] = pblock; } } } -HLTTrackClusterRemoverNew::HLTTrackClusterRemoverNew( - const ParameterSet &iConfig) +HLTTrackClusterRemoverNew::HLTTrackClusterRemoverNew(const ParameterSet &iConfig) : tTrackerGeom_(esConsumes()), doTracks_(iConfig.exists("trajectories")), - doStrip_(iConfig.existsAs("doStrip") - ? iConfig.getParameter("doStrip") - : true), - doPixel_(iConfig.existsAs("doPixel") - ? iConfig.getParameter("doPixel") - : true), + doStrip_(iConfig.existsAs("doStrip") ? iConfig.getParameter("doStrip") : true), + doPixel_(iConfig.existsAs("doPixel") ? iConfig.getParameter("doPixel") : true), mergeOld_(false), makeProducts_(true), - doStripChargeCheck_(iConfig.existsAs("doStripChargeCheck") - ? iConfig.getParameter("doStripChargeCheck") - : false), - doPixelChargeCheck_(iConfig.existsAs("doPixelChargeCheck") - ? iConfig.getParameter("doPixelChargeCheck") - : false) + doStripChargeCheck_( + iConfig.existsAs("doStripChargeCheck") ? iConfig.getParameter("doStripChargeCheck") : false), + doPixelChargeCheck_( + iConfig.existsAs("doPixelChargeCheck") ? iConfig.getParameter("doPixelChargeCheck") : false) { if (iConfig.exists("oldClusterRemovalInfo")) { - oldPxlMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldStrMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - if (not(iConfig.getParameter("oldClusterRemovalInfo") == - edm::InputTag())) + oldPxlMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldStrMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + if (not(iConfig.getParameter("oldClusterRemovalInfo") == edm::InputTag())) mergeOld_ = true; } if ((doPixelChargeCheck_ && !doPixel_) || (doStripChargeCheck_ && !doStrip_)) - throw cms::Exception("Configuration Error") - << "HLTTrackClusterRemoverNew: Charge check asked without cluster " - "collection "; + throw cms::Exception("Configuration Error") << "HLTTrackClusterRemoverNew: Charge check asked without cluster " + "collection "; if (doPixelChargeCheck_) - throw cms::Exception("Configuration Error") - << "HLTTrackClusterRemoverNew: Pixel cluster charge check not yet " - "implemented"; + throw cms::Exception("Configuration Error") << "HLTTrackClusterRemoverNew: Pixel cluster charge check not yet " + "implemented"; fill(pblocks_, pblocks_ + NumberOfParamBlocks, ParamBlock()); readPSet(iConfig, "Common", -1); @@ -190,33 +175,27 @@ HLTTrackClusterRemoverNew::HLTTrackClusterRemoverNew( bool usingCharge = false; for (size_t i = 0; i < NumberOfParamBlocks; ++i) { if (!pblocks_[i].isSet_) - throw cms::Exception("Configuration Error") - << "HLTTrackClusterRemoverNew: Missing configuration for detector " - "with subDetID = " - << (i + 1); + throw cms::Exception("Configuration Error") << "HLTTrackClusterRemoverNew: Missing configuration for detector " + "with subDetID = " + << (i + 1); if (pblocks_[i].usesCharge_ && !usingCharge) { throw cms::Exception("Configuration Error") - << "HLTTrackClusterRemoverNew: Configuration for subDetID = " - << (i + 1) << " uses cluster charge, which is not enabled."; + << "HLTTrackClusterRemoverNew: Configuration for subDetID = " << (i + 1) + << " uses cluster charge, which is not enabled."; } } // trajectories_ = consumes // >(iConfig.getParameter("trajectories")); if (doTracks_) - trajectories_ = consumes >( - iConfig.getParameter("trajectories")); + trajectories_ = consumes >(iConfig.getParameter("trajectories")); if (doPixel_) - pixelClusters_ = consumes >( - iConfig.getParameter("pixelClusters")); + pixelClusters_ = consumes >(iConfig.getParameter("pixelClusters")); if (doStrip_) - stripClusters_ = consumes >( - iConfig.getParameter("stripClusters")); + stripClusters_ = consumes >(iConfig.getParameter("stripClusters")); if (mergeOld_) { - oldPxlMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldStrMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); + oldPxlMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldStrMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); } // produces >(); @@ -228,9 +207,8 @@ HLTTrackClusterRemoverNew::HLTTrackClusterRemoverNew( HLTTrackClusterRemoverNew::~HLTTrackClusterRemoverNew() {} -void HLTTrackClusterRemoverNew::mergeOld( - ClusterRemovalInfo::Indices &refs, - const ClusterRemovalInfo::Indices &oldRefs) { +void HLTTrackClusterRemoverNew::mergeOld(ClusterRemovalInfo::Indices &refs, + const ClusterRemovalInfo::Indices &oldRefs) { for (size_t i = 0, n = refs.size(); i < n; ++i) { refs[i] = oldRefs[refs[i]]; } @@ -239,7 +217,8 @@ void HLTTrackClusterRemoverNew::mergeOld( template std::unique_ptr > HLTTrackClusterRemoverNew::cleanup( const edmNew::DetSetVector &oldClusters, - const std::vector &isGood, reco::ClusterRemovalInfo::Indices &refs, + const std::vector &isGood, + reco::ClusterRemovalInfo::Indices &refs, const reco::ClusterRemovalInfo::Indices *oldRefs) { typedef typename edmNew::DetSetVector DSV; typedef typename edmNew::DetSetVector::FastFiller DSF; @@ -252,18 +231,16 @@ std::unique_ptr > HLTTrackClusterRemoverNew::cleanup( // cluster removal loop const T *firstOffset = &oldClusters.data().front(); - for (typename DSV::const_iterator itdet = oldClusters.begin(), - enddet = oldClusters.end(); - itdet != enddet; ++itdet) { + for (typename DSV::const_iterator itdet = oldClusters.begin(), enddet = oldClusters.end(); itdet != enddet; ++itdet) { DS oldDS = *itdet; - if (oldDS.empty()) continue; // skip empty detsets + if (oldDS.empty()) + continue; // skip empty detsets uint32_t id = oldDS.detId(); DSF outds(*output, id); - for (typename DS::const_iterator it = oldDS.begin(), ed = oldDS.end(); - it != ed; ++it) { + for (typename DS::const_iterator it = oldDS.begin(), ed = oldDS.end(); it != ed; ++it) { uint32_t index = ((&*it) - firstOffset); countOld++; if (isGood[index]) { @@ -275,56 +252,52 @@ std::unique_ptr > HLTTrackClusterRemoverNew::cleanup( // - 1) << std::endl; } } - if (outds.empty()) outds.abort(); // not write in an empty DSV + if (outds.empty()) + outds.abort(); // not write in an empty DSV } // double fraction = countNew / (double) countOld; // std::cout<<"fraction: "< clusterReg.key()); } collectedRegStrips_[clusterReg.key()] = true; } -void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg) { +void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg) { DetId detid = hit->geographicalId(); uint32_t subdet = detid.subdetId(); assert((subdet > 0) && (subdet <= NumberOfParamBlocks)); // chi2 cut - if (chi2 > pblocks_[subdet - 1].maxChi2_) return; + if (chi2 > pblocks_[subdet - 1].maxChi2_) + return; if (GeomDetEnumerators::isTrackerPixel(tg->geomDetSubDetector(subdet))) { // std::cout<<"process pxl hit"<(hit); SiPixelRecHit::ClusterRef cluster = pixelHit->cluster(); if (cluster.id() != pixelSourceProdID) throw cms::Exception("Inconsistent Data") - << "HLTTrackClusterRemoverNew: pixel cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << pixelSourceProdID << ")\n."; + << "HLTTrackClusterRemoverNew: pixel cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << pixelSourceProdID << ")\n."; assert(cluster.id() == pixelSourceProdID); // DBG// cout << "HIT NEW PIXEL DETID = " << detid.rawId() << ", Cluster [ " @@ -332,8 +305,7 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, // endl; // if requested, cut on cluster size - if (pblocks_[subdet - 1].usesSize_ && - (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) + if (pblocks_[subdet - 1].usesSize_ && (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) return; // mark as used @@ -341,11 +313,11 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, assert(collectedPixels_.size() > cluster.key()); collectedPixels_[cluster.key()] = true; } else { // aka Strip - if (!doStrip_) return; + if (!doStrip_) + return; const type_info &hitType = typeid(*hit); if (hitType == typeid(SiStripRecHit2D)) { - const SiStripRecHit2D *stripHit = - static_cast(hit); + const SiStripRecHit2D *stripHit = static_cast(hit); // DBG// cout << "Plain RecHit 2D: " << endl; process(stripHit->omniClusterRef(), subdet); // int clusCharge=0; @@ -365,8 +337,7 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, //(SiStripRecHit1D) chi2: " << chi2 << " [" << subdet << " --> charge: " //<< clusCharge << "]" << std::endl; } else if (hitType == typeid(SiStripMatchedRecHit2D)) { - const SiStripMatchedRecHit2D *matchHit = - static_cast(hit); + const SiStripMatchedRecHit2D *matchHit = static_cast(hit); // DBG// cout << "Matched RecHit 2D: " << endl; process(matchHit->monoClusterRef(), subdet); // int clusCharge=0; @@ -388,8 +359,7 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, //--> charge: " << clusCharge << "]" << std::endl; } else if (hitType == typeid(ProjectedSiStripRecHit2D)) { - const ProjectedSiStripRecHit2D *projHit = - static_cast(hit); + const ProjectedSiStripRecHit2D *projHit = static_cast(hit); // DBG// cout << "Projected RecHit 2D: " << endl; process(projHit->originalHit().omniClusterRef(), subdet); // int clusCharge=0; @@ -401,8 +371,7 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, // std::endl; } else throw cms::Exception("NOT IMPLEMENTED") - << "Don't know how to handle " << hitType.name() << " on detid " - << detid.rawId() << "\n"; + << "Don't know how to handle " << hitType.name() << " on detid " << detid.rawId() << "\n"; } } @@ -417,8 +386,7 @@ void HLTTrackClusterRemoverNew::process(const TrackingRecHit *hit, float chi2, * \-- Old ClusterRemovalInfo "old" ID */ -void HLTTrackClusterRemoverNew::produce(Event &iEvent, - const EventSetup &iSetup) { +void HLTTrackClusterRemoverNew::produce(Event &iEvent, const EventSetup &iSetup) { ProductID pixelOldProdID, stripOldProdID; const auto &tgh = &iSetup.getData(tTrackerGeom_); @@ -444,9 +412,8 @@ void HLTTrackClusterRemoverNew::produce(Event &iEvent, edm::Handle oldStrMask; iEvent.getByToken(oldPxlMaskToken_, oldPxlMask); iEvent.getByToken(oldStrMaskToken_, oldStrMask); - LogDebug("TrackClusterRemover") - << "to merge in, " << oldStrMask->size() << " strp and " - << oldPxlMask->size() << " pxl"; + LogDebug("TrackClusterRemover") << "to merge in, " << oldStrMask->size() << " strp and " << oldPxlMask->size() + << " pxl"; oldStrMask->copyMaskTo(collectedRegStrips_); oldPxlMask->copyMaskTo(collectedPixels_); collectedRegStrips_.resize(stripClusters->dataSize(), false); @@ -459,16 +426,15 @@ void HLTTrackClusterRemoverNew::produce(Event &iEvent, // trajectories->begin(), ed = trajectories->end(); it != ed; ++it) { // const Trajectory &tj = * it->key; - for (std::vector::const_iterator it = trajectories->begin(), - ed = trajectories->end(); - it != ed; ++it) { + for (std::vector::const_iterator it = trajectories->begin(), ed = trajectories->end(); it != ed; ++it) { const Trajectory &tj = *it; const std::vector &tms = tj.measurements(); std::vector::const_iterator itm, endtm; for (itm = tms.begin(), endtm = tms.end(); itm != endtm; ++itm) { const TrackingRecHit *hit = itm->recHit()->hit(); - if (!hit->isValid()) continue; + if (!hit->isValid()) + continue; // std::cout<<"process hit"<estimate(), tgh); } @@ -487,18 +453,21 @@ void HLTTrackClusterRemoverNew::produce(Event &iEvent, auto const &clusters = stripClusters->data(); for (auto const &item : stripClusters->ids()) { - if (!item.isValid()) continue; // not umpacked + if (!item.isValid()) + continue; // not umpacked DetId detid = item.id; uint32_t subdet = detid.subdetId(); - if (!pblocks_[subdet - 1].cutOnStripCharge_) continue; + if (!pblocks_[subdet - 1].cutOnStripCharge_) + continue; // std::cout << " i: " << i << " --> detid: " << detid << " --> // subdet: " << subdet << std::endl; for (int i = item.offset; i < item.offset + int(item.size); ++i) { int clusCharge = 0; - for (auto cAmp : clusters[i].amplitudes()) clusCharge += cAmp; + for (auto cAmp : clusters[i].amplitudes()) + clusCharge += cAmp; // if (clusCharge < pblocks_[subdet-1].minGoodStripCharge_) // std::cout << " clusCharge: " << clusCharge << std::endl; @@ -512,22 +481,17 @@ void HLTTrackClusterRemoverNew::produce(Event &iEvent, // << std::endl; std::cout << " total strip to skip: // "<( - edm::RefProd >(stripClusters), - collectedRegStrips_)); - - LogDebug("TrackClusterRemover") - << "total pxl to skip: " - << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); - iEvent.put(std::make_unique( - edm::RefProd >(pixelClusters), - collectedPixels_)); + iEvent.put(std::make_unique(edm::RefProd >(stripClusters), + collectedRegStrips_)); + + LogDebug("TrackClusterRemover") << "total pxl to skip: " + << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); + iEvent.put(std::make_unique(edm::RefProd >(pixelClusters), + collectedPixels_)); collectedRegStrips_.clear(); collectedPixels_.clear(); diff --git a/RecoLocalTracker/SubCollectionProducers/src/JetCoreClusterSplitter.cc b/RecoLocalTracker/SubCollectionProducers/src/JetCoreClusterSplitter.cc index dd7fd38d6e38b..7b4cf18e00dcc 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/JetCoreClusterSplitter.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/JetCoreClusterSplitter.cc @@ -20,30 +20,30 @@ #include "RecoLocalTracker/Records/interface/TkPixelCPERecord.h" class JetCoreClusterSplitter : public edm::stream::EDProducer<> { - public: +public: JetCoreClusterSplitter(const edm::ParameterSet& iConfig); ~JetCoreClusterSplitter() override; void produce(edm::Event& iEvent, const edm::EventSetup& iSetup) override; - private: +private: bool split(const SiPixelCluster& aCluster, edmNew::DetSetVector::FastFiller& filler, - float expectedADC, int sizeY, int sizeX, float jetZOverRho); + float expectedADC, + int sizeY, + int sizeX, + float jetZOverRho); std::vector fittingSplit(const SiPixelCluster& aCluster, - float expectedADC, int sizeY, - int sizeX, float jetZOverRho, + float expectedADC, + int sizeY, + int sizeX, + float jetZOverRho, unsigned int nSplitted); - std::pair closestClusters( - const std::vector& distanceMap); - std::multimap secondDistDiffScore( - const std::vector>& distanceMap); - std::multimap secondDistScore( - const std::vector>& distanceMap); - std::multimap distScore( - const std::vector>& distanceMap); - - edm::ESGetToken const - tTrackingGeom_; + std::pair closestClusters(const std::vector& distanceMap); + std::multimap secondDistDiffScore(const std::vector>& distanceMap); + std::multimap secondDistScore(const std::vector>& distanceMap); + std::multimap distScore(const std::vector>& distanceMap); + + edm::ESGetToken const tTrackingGeom_; edm::ESGetToken const tCPE_; bool verbose; @@ -62,18 +62,15 @@ class JetCoreClusterSplitter : public edm::stream::EDProducer<> { JetCoreClusterSplitter::JetCoreClusterSplitter(const edm::ParameterSet& iConfig) : tTrackingGeom_(esConsumes()), - tCPE_(esConsumes( - edm::ESInputTag("", iConfig.getParameter("pixelCPE")))), + tCPE_(esConsumes(edm::ESInputTag("", iConfig.getParameter("pixelCPE")))), verbose(iConfig.getParameter("verbose")), ptMin_(iConfig.getParameter("ptMin")), deltaR_(iConfig.getParameter("deltaRmax")), chargeFracMin_(iConfig.getParameter("chargeFractionMin")), - pixelClusters_(consumes>( - iConfig.getParameter("pixelClusters"))), - vertices_(consumes( - iConfig.getParameter("vertices"))), - cores_(consumes>( - iConfig.getParameter("cores"))), + pixelClusters_( + consumes>(iConfig.getParameter("pixelClusters"))), + vertices_(consumes(iConfig.getParameter("vertices"))), + cores_(consumes>(iConfig.getParameter("cores"))), forceXError_(iConfig.getParameter("forceXError")), forceYError_(iConfig.getParameter("forceYError")), fractionalWidth_(iConfig.getParameter("fractionalWidth")), @@ -86,13 +83,9 @@ JetCoreClusterSplitter::JetCoreClusterSplitter(const edm::ParameterSet& iConfig) JetCoreClusterSplitter::~JetCoreClusterSplitter() {} -bool SortPixels(const SiPixelCluster::Pixel& i, - const SiPixelCluster::Pixel& j) { - return (i.adc > j.adc); -} +bool SortPixels(const SiPixelCluster::Pixel& i, const SiPixelCluster::Pixel& j) { return (i.adc > j.adc); } -void JetCoreClusterSplitter::produce(edm::Event& iEvent, - const edm::EventSetup& iSetup) { +void JetCoreClusterSplitter::produce(edm::Event& iEvent, const edm::EventSetup& iSetup) { using namespace edm; const auto& geometry = &iSetup.getData(tTrackingGeom_); @@ -109,21 +102,17 @@ void JetCoreClusterSplitter::produce(edm::Event& iEvent, const PixelClusterParameterEstimator* pp = &iSetup.getData(tCPE_); auto output = std::make_unique>(); - edmNew::DetSetVector::const_iterator detIt = - inputPixelClusters->begin(); + edmNew::DetSetVector::const_iterator detIt = inputPixelClusters->begin(); for (; detIt != inputPixelClusters->end(); detIt++) { - edmNew::DetSetVector::FastFiller filler(*output, - detIt->id()); + edmNew::DetSetVector::FastFiller filler(*output, detIt->id()); const edmNew::DetSet& detset = *detIt; const GeomDet* det = geometry->idToDet(detset.id()); for (auto cluster = detset.begin(); cluster != detset.end(); cluster++) { const SiPixelCluster& aCluster = *cluster; bool hasBeenSplit = false; bool shouldBeSplit = false; - GlobalPoint cPos = det->surface().toGlobal( - pp->localParametersV(aCluster, - (*geometry->idToDetUnit(detIt->id())))[0] - .first); + GlobalPoint cPos = + det->surface().toGlobal(pp->localParametersV(aCluster, (*geometry->idToDetUnit(detIt->id())))[0].first); GlobalPoint ppv(pv.position().x(), pv.position().y(), pv.position().z()); GlobalVector clusterDir = cPos - ppv; for (unsigned int ji = 0; ji < cores->size(); ji++) { @@ -133,38 +122,32 @@ void JetCoreClusterSplitter::produce(edm::Event& iEvent, if (Geom::deltaR(jetDir, clusterDir) < deltaR_) { // check if the cluster has to be splitted - bool isEndCap = - (std::abs(cPos.z()) > 30.f); // FIXME: check detID instead! + bool isEndCap = (std::abs(cPos.z()) > 30.f); // FIXME: check detID instead! float jetZOverRho = jet.momentum().Z() / jet.momentum().Rho(); if (isEndCap) jetZOverRho = jet.momentum().Rho() / jet.momentum().Z(); - float expSizeY = std::sqrt( - (1.3f * 1.3f) + (1.9f * 1.9f) * jetZOverRho * jetZOverRho); - if (expSizeY < 1.f) expSizeY = 1.f; + float expSizeY = std::sqrt((1.3f * 1.3f) + (1.9f * 1.9f) * jetZOverRho * jetZOverRho); + if (expSizeY < 1.f) + expSizeY = 1.f; float expSizeX = 1.5f; if (isEndCap) { expSizeX = expSizeY; expSizeY = 1.5f; } // in endcap col/rows are switched - float expCharge = std::sqrt(1.08f + jetZOverRho * jetZOverRho) * - centralMIPCharge_; + float expCharge = std::sqrt(1.08f + jetZOverRho * jetZOverRho) * centralMIPCharge_; if (aCluster.charge() > expCharge * chargeFracMin_ && - (aCluster.sizeX() > expSizeX + 1 || - aCluster.sizeY() > expSizeY + 1)) { + (aCluster.sizeX() > expSizeX + 1 || aCluster.sizeY() > expSizeY + 1)) { shouldBeSplit = true; if (verbose) - std::cout << "Trying to split: charge and deltaR " - << aCluster.charge() << " " - << Geom::deltaR(jetDir, clusterDir) << " size x y " - << aCluster.sizeX() << " " << aCluster.sizeY() - << " exp. size (x,y) " << expSizeX << " " << expSizeY - << " detid " << detIt->id() << std::endl; + std::cout << "Trying to split: charge and deltaR " << aCluster.charge() << " " + << Geom::deltaR(jetDir, clusterDir) << " size x y " << aCluster.sizeX() << " " + << aCluster.sizeY() << " exp. size (x,y) " << expSizeX << " " << expSizeY << " detid " + << detIt->id() << std::endl; if (verbose) std::cout << "jetZOverRho=" << jetZOverRho << std::endl; - if (split(aCluster, filler, expCharge, expSizeY, expSizeX, - jetZOverRho)) { + if (split(aCluster, filler, expCharge, expSizeY, expSizeX, jetZOverRho)) { hasBeenSplit = true; } } @@ -176,54 +159,48 @@ void JetCoreClusterSplitter::produce(edm::Event& iEvent, if (shouldBeSplit) { // blowup the error if we failed to split a splittable cluster (does // it ever happen) - c.setSplitClusterErrorX( - c.sizeX() * - (100.f / 3.f)); // this is not really blowing up .. TODO: tune + c.setSplitClusterErrorX(c.sizeX() * (100.f / 3.f)); // this is not really blowing up .. TODO: tune c.setSplitClusterErrorY(c.sizeY() * (150.f / 3.f)); } filler.push_back(c); - std::push_heap( - filler.begin(), filler.end(), - [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { - return cl1.minPixelRow() < cl2.minPixelRow(); - }); + std::push_heap(filler.begin(), filler.end(), [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { + return cl1.minPixelRow() < cl2.minPixelRow(); + }); } } // loop over clusters - std::sort_heap(filler.begin(), filler.end(), - [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { - return cl1.minPixelRow() < cl2.minPixelRow(); - }); + std::sort_heap(filler.begin(), filler.end(), [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { + return cl1.minPixelRow() < cl2.minPixelRow(); + }); } // loop over det iEvent.put(std::move(output)); } -bool JetCoreClusterSplitter::split( - const SiPixelCluster& aCluster, - edmNew::DetSetVector::FastFiller& filler, float expectedADC, - int sizeY, int sizeX, float jetZOverRho) { +bool JetCoreClusterSplitter::split(const SiPixelCluster& aCluster, + edmNew::DetSetVector::FastFiller& filler, + float expectedADC, + int sizeY, + int sizeX, + float jetZOverRho) { // This function should test several configuration of splitting, and then // return the one with best chi2 - std::vector sp = - fittingSplit(aCluster, expectedADC, sizeY, sizeX, jetZOverRho, - std::floor(aCluster.charge() / expectedADC + 0.5f)); + std::vector sp = fittingSplit( + aCluster, expectedADC, sizeY, sizeX, jetZOverRho, std::floor(aCluster.charge() / expectedADC + 0.5f)); // for the config with best chi2 for (unsigned int i = 0; i < sp.size(); i++) { filler.push_back(sp[i]); - std::push_heap(filler.begin(), filler.end(), - [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { - return cl1.minPixelRow() < cl2.minPixelRow(); - }); + std::push_heap(filler.begin(), filler.end(), [](SiPixelCluster const& cl1, SiPixelCluster const& cl2) { + return cl1.minPixelRow() < cl2.minPixelRow(); + }); } return (!sp.empty()); } // order with fast algo and pick first and second instead? -std::pair JetCoreClusterSplitter::closestClusters( - const std::vector& distanceMap) { +std::pair JetCoreClusterSplitter::closestClusters(const std::vector& distanceMap) { float minDist = std::numeric_limits::max(); float secondMinDist = std::numeric_limits::max(); for (unsigned int i = 0; i < distanceMap.size(); i++) { @@ -248,8 +225,7 @@ std::multimap JetCoreClusterSplitter::secondDistDiffScore( return scores; } -std::multimap JetCoreClusterSplitter::secondDistScore( - const std::vector>& distanceMap) { +std::multimap JetCoreClusterSplitter::secondDistScore(const std::vector>& distanceMap) { std::multimap scores; for (unsigned int j = 0; j < distanceMap.size(); j++) { std::pair d = closestClusters(distanceMap[j]); @@ -258,8 +234,7 @@ std::multimap JetCoreClusterSplitter::secondDistScore( return scores; } -std::multimap JetCoreClusterSplitter::distScore( - const std::vector>& distanceMap) { +std::multimap JetCoreClusterSplitter::distScore(const std::vector>& distanceMap) { std::multimap scores; for (unsigned int j = 0; j < distanceMap.size(); j++) { std::pair d = closestClusters(distanceMap[j]); @@ -268,9 +243,12 @@ std::multimap JetCoreClusterSplitter::distScore( return scores; } -std::vector JetCoreClusterSplitter::fittingSplit( - const SiPixelCluster& aCluster, float expectedADC, int sizeY, int sizeX, - float jetZOverRho, unsigned int nSplitted) { +std::vector JetCoreClusterSplitter::fittingSplit(const SiPixelCluster& aCluster, + float expectedADC, + int sizeY, + int sizeX, + float jetZOverRho, + unsigned int nSplitted) { std::vector output; unsigned int meanExp = nSplitted; @@ -287,21 +265,18 @@ std::vector JetCoreClusterSplitter::fittingSplit( std::vector originalpixels = aCluster.pixels(); std::vector> pixels; for (unsigned int j = 0; j < originalpixels.size(); j++) { - int sub = originalpixels[j].adc / chargePerUnit_ * expectedADC / - centralMIPCharge_; - if (sub < 1) sub = 1; + int sub = originalpixels[j].adc / chargePerUnit_ * expectedADC / centralMIPCharge_; + if (sub < 1) + sub = 1; int perDiv = originalpixels[j].adc / sub; if (verbose) - std::cout << "Splitting " << j << " in [ " << pixels.size() << " , " - << pixels.size() + sub - << " ], expected numb of clusters: " << meanExp - << " original pixel (x,y) " << originalpixels[j].x << " " - << originalpixels[j].y << " sub " << sub << std::endl; + std::cout << "Splitting " << j << " in [ " << pixels.size() << " , " << pixels.size() + sub + << " ], expected numb of clusters: " << meanExp << " original pixel (x,y) " << originalpixels[j].x + << " " << originalpixels[j].y << " sub " << sub << std::endl; for (int k = 0; k < sub; k++) { - if (k == sub - 1) perDiv = originalpixels[j].adc - perDiv * k; - pixels.push_back(std::make_pair( - j, SiPixelCluster::Pixel(originalpixels[j].x, originalpixels[j].y, - perDiv))); + if (k == sub - 1) + perDiv = originalpixels[j].adc - perDiv * k; + pixels.push_back(std::make_pair(j, SiPixelCluster::Pixel(originalpixels[j].x, originalpixels[j].y, perDiv))); } } std::vector clusterForPixel(pixels.size()); @@ -318,42 +293,34 @@ std::vector JetCoreClusterSplitter::fittingSplit( while (!stop && remainingSteps > 0) { remainingSteps--; // Compute all distances - std::vector> distanceMapX(originalpixels.size(), - std::vector(meanExp)); - std::vector> distanceMapY(originalpixels.size(), - std::vector(meanExp)); - std::vector> distanceMap(originalpixels.size(), - std::vector(meanExp)); + std::vector> distanceMapX(originalpixels.size(), std::vector(meanExp)); + std::vector> distanceMapY(originalpixels.size(), std::vector(meanExp)); + std::vector> distanceMap(originalpixels.size(), std::vector(meanExp)); for (unsigned int j = 0; j < originalpixels.size(); j++) { if (verbose) - std::cout << "Original Pixel pos " << j << " " << pixels[j].second.x - << " " << pixels[j].second.y << std::endl; + std::cout << "Original Pixel pos " << j << " " << pixels[j].second.x << " " << pixels[j].second.y << std::endl; for (unsigned int i = 0; i < meanExp; i++) { distanceMapX[j][i] = 1.f * originalpixels[j].x - clx[i]; distanceMapY[j][i] = 1.f * originalpixels[j].y - cly[i]; float dist = 0; // float sizeX=2; if (std::abs(distanceMapX[j][i]) > sizeX / 2.f) { - dist += (std::abs(distanceMapX[j][i]) - sizeX / 2.f + 1.f) * - (std::abs(distanceMapX[j][i]) - sizeX / 2.f + 1.f); + dist += + (std::abs(distanceMapX[j][i]) - sizeX / 2.f + 1.f) * (std::abs(distanceMapX[j][i]) - sizeX / 2.f + 1.f); } else { - dist += (2.f * distanceMapX[j][i] / sizeX) * - (2.f * distanceMapX[j][i] / sizeX); + dist += (2.f * distanceMapX[j][i] / sizeX) * (2.f * distanceMapX[j][i] / sizeX); } if (std::abs(distanceMapY[j][i]) > sizeY / 2.f) { dist += 1.f * (std::abs(distanceMapY[j][i]) - sizeY / 2.f + 1.f) * (std::abs(distanceMapY[j][i]) - sizeY / 2.f + 1.f); } else { - dist += 1.f * (2.f * distanceMapY[j][i] / sizeY) * - (2.f * distanceMapY[j][i] / sizeY); + dist += 1.f * (2.f * distanceMapY[j][i] / sizeY) * (2.f * distanceMapY[j][i] / sizeY); } distanceMap[j][i] = sqrt(dist); if (verbose) - std::cout << "Cluster " << i << " Original Pixel " << j - << " distances: " << distanceMapX[j][i] << " " - << distanceMapY[j][i] << " " << distanceMap[j][i] - << std::endl; + std::cout << "Cluster " << i << " Original Pixel " << j << " distances: " << distanceMapX[j][i] << " " + << distanceMapY[j][i] << " " << distanceMap[j][i] << std::endl; } } // Compute scores for sequential addition. The first index is the @@ -367,16 +334,13 @@ std::vector JetCoreClusterSplitter::fittingSplit( // Iterate starting from the ones with furthest second best clusters, i.e. // easy choices std::vector weightOfPixel(pixels.size()); - for (std::multimap::iterator it = scores.begin(); - it != scores.end(); it++) { + for (std::multimap::iterator it = scores.begin(); it != scores.end(); it++) { int pixel_index = it->second; if (verbose) - std::cout << "Original Pixel " << pixel_index << " with score " - << it->first << std::endl; + std::cout << "Original Pixel " << pixel_index << " with score " << it->first << std::endl; // find cluster that is both close and has some charge still to assign int subpixel_counter = 0; - for (auto subpixel = pixels.begin(); subpixel != pixels.end(); - ++subpixel, ++subpixel_counter) { + for (auto subpixel = pixels.begin(); subpixel != pixels.end(); ++subpixel, ++subpixel_counter) { if (subpixel->first > pixel_index) { break; } else if (subpixel->first != pixel_index) { @@ -384,20 +348,14 @@ std::vector JetCoreClusterSplitter::fittingSplit( } else { float maxEst = 0; int cl = -1; - for (unsigned int subcluster_index = 0; subcluster_index < meanExp; - subcluster_index++) { - float nsig = - (cls[subcluster_index] - expectedADC) / - (expectedADC * - fractionalWidth_); // 20% uncertainty? realistic from Landau? - float clQest = - 1.f / (1.f + std::exp(nsig)) + 1e-6f; // 1./(1.+exp(x*x-3*3)) - float clDest = - 1.f / (distanceMap[pixel_index][subcluster_index] + 0.05f); + for (unsigned int subcluster_index = 0; subcluster_index < meanExp; subcluster_index++) { + float nsig = (cls[subcluster_index] - expectedADC) / + (expectedADC * fractionalWidth_); // 20% uncertainty? realistic from Landau? + float clQest = 1.f / (1.f + std::exp(nsig)) + 1e-6f; // 1./(1.+exp(x*x-3*3)) + float clDest = 1.f / (distanceMap[pixel_index][subcluster_index] + 0.05f); if (verbose) - std::cout << " Q: " << clQest << " D: " << clDest << " " - << distanceMap[pixel_index][subcluster_index] + std::cout << " Q: " << clQest << " D: " << clDest << " " << distanceMap[pixel_index][subcluster_index] << std::endl; float est = clQest * clDest; if (est > maxEst) { @@ -409,15 +367,14 @@ std::vector JetCoreClusterSplitter::fittingSplit( clusterForPixel[subpixel_counter] = cl; weightOfPixel[subpixel_counter] = maxEst; if (verbose) - std::cout << "Pixel weight j cl " << weightOfPixel[subpixel_counter] - << " " << subpixel_counter << " " << cl << std::endl; + std::cout << "Pixel weight j cl " << weightOfPixel[subpixel_counter] << " " << subpixel_counter << " " << cl + << std::endl; } } } // Recompute cluster centers stop = true; - for (unsigned int subcluster_index = 0; subcluster_index < meanExp; - subcluster_index++) { + for (unsigned int subcluster_index = 0; subcluster_index < meanExp; subcluster_index++) { if (std::abs(clx[subcluster_index] - oldclx[subcluster_index]) > 0.01f) stop = false; // still moving if (std::abs(cly[subcluster_index] - oldcly[subcluster_index]) > 0.01f) @@ -428,66 +385,53 @@ std::vector JetCoreClusterSplitter::fittingSplit( cly[subcluster_index] = 0; cls[subcluster_index] = 1e-99; } - for (unsigned int pixel_index = 0; pixel_index < pixels.size(); - pixel_index++) { - if (clusterForPixel[pixel_index] < 0) continue; + for (unsigned int pixel_index = 0; pixel_index < pixels.size(); pixel_index++) { + if (clusterForPixel[pixel_index] < 0) + continue; if (verbose) - std::cout << "j " << pixel_index << " x " - << pixels[pixel_index].second.x << " * y " - << pixels[pixel_index].second.y << " * ADC " - << pixels[pixel_index].second.adc << " * W " + std::cout << "j " << pixel_index << " x " << pixels[pixel_index].second.x << " * y " + << pixels[pixel_index].second.y << " * ADC " << pixels[pixel_index].second.adc << " * W " << weightOfPixel[pixel_index] << std::endl; - clx[clusterForPixel[pixel_index]] += - pixels[pixel_index].second.x * pixels[pixel_index].second.adc; - cly[clusterForPixel[pixel_index]] += - pixels[pixel_index].second.y * pixels[pixel_index].second.adc; + clx[clusterForPixel[pixel_index]] += pixels[pixel_index].second.x * pixels[pixel_index].second.adc; + cly[clusterForPixel[pixel_index]] += pixels[pixel_index].second.y * pixels[pixel_index].second.adc; cls[clusterForPixel[pixel_index]] += pixels[pixel_index].second.adc; } - for (unsigned int subcluster_index = 0; subcluster_index < meanExp; - subcluster_index++) { + for (unsigned int subcluster_index = 0; subcluster_index < meanExp; subcluster_index++) { if (cls[subcluster_index] != 0) { clx[subcluster_index] /= cls[subcluster_index]; cly[subcluster_index] /= cls[subcluster_index]; } if (verbose) - std::cout << "Center for cluster " << subcluster_index << " x,y " - << clx[subcluster_index] << " " << cly[subcluster_index] - << std::endl; + std::cout << "Center for cluster " << subcluster_index << " x,y " << clx[subcluster_index] << " " + << cly[subcluster_index] << std::endl; cls[subcluster_index] = 0; } } - if (verbose) std::cout << "maxstep " << remainingSteps << std::endl; + if (verbose) + std::cout << "maxstep " << remainingSteps << std::endl; // accumulate pixel with same cl std::vector> pixelsForCl(meanExp); for (int cl = 0; cl < (int)meanExp; cl++) { for (unsigned int j = 0; j < pixels.size(); j++) { - if (clusterForPixel[j] == cl and - pixels[j].second.adc != 0) { // for each pixel of cluster - // cl find the other pixels - // with same x,y and - // accumulate+reset their adc + if (clusterForPixel[j] == cl and pixels[j].second.adc != 0) { // for each pixel of cluster + // cl find the other pixels + // with same x,y and + // accumulate+reset their adc for (unsigned int k = j + 1; k < pixels.size(); k++) { - if (pixels[k].second.adc != 0 and - pixels[k].second.x == pixels[j].second.x and - pixels[k].second.y == pixels[j].second.y and - clusterForPixel[k] == cl) { + if (pixels[k].second.adc != 0 and pixels[k].second.x == pixels[j].second.x and + pixels[k].second.y == pixels[j].second.y and clusterForPixel[k] == cl) { if (verbose) - std::cout << "Resetting all sub-pixel for location " - << pixels[k].second.x << ", " << pixels[k].second.y - << " at index " << k << " associated to cl " - << clusterForPixel[k] << std::endl; + std::cout << "Resetting all sub-pixel for location " << pixels[k].second.x << ", " << pixels[k].second.y + << " at index " << k << " associated to cl " << clusterForPixel[k] << std::endl; pixels[j].second.adc += pixels[k].second.adc; pixels[k].second.adc = 0; } } for (unsigned int p = 0; p < pixels.size(); ++p) if (verbose) - std::cout << "index, x, y, ADC: " << p << ", " << pixels[p].second.x - << ", " << pixels[p].second.y << ", " - << pixels[p].second.adc << " associated to cl " - << clusterForPixel[p] << std::endl - << "Adding pixel " << pixels[j].second.x << ", " - << pixels[j].second.y << " to cluster " << cl + std::cout << "index, x, y, ADC: " << p << ", " << pixels[p].second.x << ", " << pixels[p].second.y << ", " + << pixels[p].second.adc << " associated to cl " << clusterForPixel[p] << std::endl + << "Adding pixel " << pixels[j].second.x << ", " << pixels[j].second.y << " to cluster " << cl << std::endl; pixelsForCl[cl].push_back(pixels[j].second); } @@ -499,10 +443,10 @@ std::vector JetCoreClusterSplitter::fittingSplit( //>(512,std::vector(512,0))); for (int cl = 0; cl < (int)meanExp; cl++) { - if (verbose) std::cout << "Pixels of cl " << cl << " "; + if (verbose) + std::cout << "Pixels of cl " << cl << " "; for (unsigned int j = 0; j < pixelsForCl[cl].size(); j++) { - SiPixelCluster::PixelPos newpix(pixelsForCl[cl][j].x, - pixelsForCl[cl][j].y); + SiPixelCluster::PixelPos newpix(pixelsForCl[cl][j].x, pixelsForCl[cl][j].y); if (verbose) std::cout << pixelsForCl[cl][j].x << "," << pixelsForCl[cl][j].y << "|"; if (j == 0) { @@ -511,10 +455,13 @@ std::vector JetCoreClusterSplitter::fittingSplit( output.back().add(newpix, pixelsForCl[cl][j].adc); } } - if (verbose) std::cout << std::endl; + if (verbose) + std::cout << std::endl; if (!pixelsForCl[cl].empty()) { - if (forceXError_ > 0) output.back().setSplitClusterErrorX(forceXError_); - if (forceYError_ > 0) output.back().setSplitClusterErrorY(forceYError_); + if (forceXError_ > 0) + output.back().setSplitClusterErrorX(forceXError_); + if (forceYError_ > 0) + output.back().setSplitClusterErrorY(forceYError_); } } // if(verbose) std::cout << "Weights" << std::endl; diff --git a/RecoLocalTracker/SubCollectionProducers/src/PixelClusterSelectorTopBottom.cc b/RecoLocalTracker/SubCollectionProducers/src/PixelClusterSelectorTopBottom.cc index 7ba94042a75e8..0d7ce8b61efd1 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/PixelClusterSelectorTopBottom.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/PixelClusterSelectorTopBottom.cc @@ -2,8 +2,7 @@ #include "FWCore/Framework/interface/MakerMacros.h" -void PixelClusterSelectorTopBottom::produce( - edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const { +void PixelClusterSelectorTopBottom::produce(edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const { edm::Handle input; event.getByToken(token_, input); @@ -11,22 +10,18 @@ void PixelClusterSelectorTopBottom::produce( auto output = std::make_unique(); - for (SiPixelClusterCollectionNew::const_iterator clustSet = input->begin(); - clustSet != input->end(); ++clustSet) { + for (SiPixelClusterCollectionNew::const_iterator clustSet = input->begin(); clustSet != input->end(); ++clustSet) { edmNew::DetSet::const_iterator clustIt = clustSet->begin(); edmNew::DetSet::const_iterator end = clustSet->end(); DetId detIdObject(clustSet->detId()); edmNew::DetSetVector::FastFiller spc(*output, detIdObject); - const PixelGeomDetUnit* theGeomDet = - dynamic_cast(theTracker.idToDet(detIdObject)); + const PixelGeomDetUnit* theGeomDet = dynamic_cast(theTracker.idToDet(detIdObject)); const PixelTopology* topol = (&(theGeomDet->specificTopology())); for (; clustIt != end; ++clustIt) { - LocalPoint lpclust = topol->localPosition( - MeasurementPoint((*clustIt).x(), (*clustIt).y())); - GlobalPoint GPclust = theGeomDet->surface().toGlobal( - Local3DPoint(lpclust.x(), lpclust.y(), lpclust.z())); + LocalPoint lpclust = topol->localPosition(MeasurementPoint((*clustIt).x(), (*clustIt).y())); + GlobalPoint GPclust = theGeomDet->surface().toGlobal(Local3DPoint(lpclust.x(), lpclust.y(), lpclust.z())); double clustY = GPclust.y(); if ((clustY * y_) > 0) { spc.push_back(*clustIt); diff --git a/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemover.cc b/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemover.cc index ad565a089db16..ea40d236525c2 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemover.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemover.cc @@ -30,13 +30,12 @@ // class SeedClusterRemover : public edm::stream::EDProducer<> { - public: +public: SeedClusterRemover(const edm::ParameterSet &iConfig); void produce(edm::Event &iEvent, const edm::EventSetup &iSetup) override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; struct ParamBlock { ParamBlock() : isSet_(false), usesCharge_(false) {} ParamBlock(const edm::ParameterSet &iConfig) @@ -44,8 +43,7 @@ class SeedClusterRemover : public edm::stream::EDProducer<> { usesCharge_(iConfig.exists("maxCharge")), usesSize_(iConfig.exists("maxSize")), maxChi2_(iConfig.getParameter("maxChi2")), - maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") - : 0), + maxCharge_(usesCharge_ ? iConfig.getParameter("maxCharge") : 0), maxSize_(usesSize_ ? iConfig.getParameter("maxSize") : 0) {} bool isSet_, usesCharge_, usesSize_; float maxChi2_, maxCharge_; @@ -56,10 +54,8 @@ class SeedClusterRemover : public edm::stream::EDProducer<> { bool doStrip_, doPixel_; bool mergeOld_; - typedef edm::ContainerMask > - PixelMaskContainer; - typedef edm::ContainerMask > - StripMaskContainer; + typedef edm::ContainerMask > PixelMaskContainer; + typedef edm::ContainerMask > StripMaskContainer; edm::EDGetTokenT > pixelClusters_; edm::EDGetTokenT > stripClusters_; edm::EDGetTokenT oldPxlMaskToken_; @@ -67,16 +63,20 @@ class SeedClusterRemover : public edm::stream::EDProducer<> { edm::EDGetTokenT trajectories_; ParamBlock pblocks_[NumberOfParamBlocks]; - void readPSet(const edm::ParameterSet &iConfig, const std::string &name, - int id1 = -1, int id2 = -1, int id3 = -1, int id4 = -1, - int id5 = -1, int id6 = -1); + void readPSet(const edm::ParameterSet &iConfig, + const std::string &name, + int id1 = -1, + int id2 = -1, + int id3 = -1, + int id4 = -1, + int id5 = -1, + int id6 = -1); std::vector pixels, strips; // avoid unneed alloc/dealloc of this edm::ProductID pixelSourceProdID, stripSourceProdID; // ProdIDs refs must point to (for consistency tests) - inline void process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg); + inline void process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg); inline void process(const OmniClusterRef &cluRef, uint32_t subdet); std::vector collectedStrips_; @@ -86,32 +86,32 @@ class SeedClusterRemover : public edm::stream::EDProducer<> { using namespace std; using namespace edm; -void SeedClusterRemover::readPSet(const edm::ParameterSet &iConfig, - const std::string &name, int id1, int id2, - int id3, int id4, int id5, int id6) { +void SeedClusterRemover::readPSet( + const edm::ParameterSet &iConfig, const std::string &name, int id1, int id2, int id3, int id4, int id5, int id6) { if (iConfig.exists(name)) { ParamBlock pblock(iConfig.getParameter(name)); if (id1 == -1) { fill(pblocks_, pblocks_ + NumberOfParamBlocks, pblock); } else { pblocks_[id1] = pblock; - if (id2 != -1) pblocks_[id2] = pblock; - if (id3 != -1) pblocks_[id3] = pblock; - if (id4 != -1) pblocks_[id4] = pblock; - if (id5 != -1) pblocks_[id5] = pblock; - if (id6 != -1) pblocks_[id6] = pblock; + if (id2 != -1) + pblocks_[id2] = pblock; + if (id3 != -1) + pblocks_[id3] = pblock; + if (id4 != -1) + pblocks_[id4] = pblock; + if (id5 != -1) + pblocks_[id5] = pblock; + if (id6 != -1) + pblocks_[id6] = pblock; } } } SeedClusterRemover::SeedClusterRemover(const ParameterSet &iConfig) : tTrackerGeom_(esConsumes()), - doStrip_(iConfig.existsAs("doStrip") - ? iConfig.getParameter("doStrip") - : true), - doPixel_(iConfig.existsAs("doPixel") - ? iConfig.getParameter("doPixel") - : true), + doStrip_(iConfig.existsAs("doStrip") ? iConfig.getParameter("doStrip") : true), + doPixel_(iConfig.existsAs("doPixel") ? iConfig.getParameter("doPixel") : true), mergeOld_(iConfig.exists("oldClusterRemovalInfo")) { fill(pblocks_, pblocks_ + NumberOfParamBlocks, ParamBlock()); readPSet(iConfig, "Common", -1); @@ -133,49 +133,38 @@ SeedClusterRemover::SeedClusterRemover(const ParameterSet &iConfig) bool usingCharge = false; for (size_t i = 0; i < NumberOfParamBlocks; ++i) { if (!pblocks_[i].isSet_) - throw cms::Exception("Configuration Error") - << "SeedClusterRemover: Missing configuration for detector with " - "subDetID = " - << (i + 1); + throw cms::Exception("Configuration Error") << "SeedClusterRemover: Missing configuration for detector with " + "subDetID = " + << (i + 1); if (pblocks_[i].usesCharge_ && !usingCharge) { - throw cms::Exception("Configuration Error") - << "SeedClusterRemover: Configuration for subDetID = " << (i + 1) - << " uses cluster charge, which is not enabled."; + throw cms::Exception("Configuration Error") << "SeedClusterRemover: Configuration for subDetID = " << (i + 1) + << " uses cluster charge, which is not enabled."; } } - trajectories_ = consumes( - iConfig.getParameter("trajectories")); + trajectories_ = consumes(iConfig.getParameter("trajectories")); if (doPixel_) - pixelClusters_ = consumes >( - iConfig.getParameter("pixelClusters")); + pixelClusters_ = consumes >(iConfig.getParameter("pixelClusters")); if (doStrip_) - stripClusters_ = consumes >( - iConfig.getParameter("stripClusters")); + stripClusters_ = consumes >(iConfig.getParameter("stripClusters")); if (mergeOld_) { - oldPxlMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldStrMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); + oldPxlMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldStrMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); } produces > >(); produces > >(); } -void SeedClusterRemover::process(OmniClusterRef const &ocluster, - uint32_t subdet) { +void SeedClusterRemover::process(OmniClusterRef const &ocluster, uint32_t subdet) { SiStripRecHit2D::ClusterRef cluster = ocluster.cluster_strip(); if (cluster.id() != stripSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemover: strip cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << stripSourceProdID << ")\n."; + << "SeedClusterRemover: strip cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << stripSourceProdID << ")\n."; assert(cluster.id() == stripSourceProdID); - if (pblocks_[subdet - 1].usesSize_ && - (cluster->amplitudes().size() > pblocks_[subdet - 1].maxSize_)) + if (pblocks_[subdet - 1].usesSize_ && (cluster->amplitudes().size() > pblocks_[subdet - 1].maxSize_)) return; strips[cluster.key()] = false; @@ -183,18 +172,19 @@ void SeedClusterRemover::process(OmniClusterRef const &ocluster, collectedStrips_[cluster.key()] = true; } -void SeedClusterRemover::process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg) { +void SeedClusterRemover::process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg) { DetId detid = hit->geographicalId(); uint32_t subdet = detid.subdetId(); assert((subdet > 0) && (subdet <= NumberOfParamBlocks)); // chi2 cut - if (chi2 > pblocks_[subdet - 1].maxChi2_) return; + if (chi2 > pblocks_[subdet - 1].maxChi2_) + return; if (GeomDetEnumerators::isTrackerPixel(tg->geomDetSubDetector(subdet))) { - if (!doPixel_) return; + if (!doPixel_) + return; // this is a pixel, and i *know* it is const SiPixelRecHit *pixelHit = static_cast(hit); @@ -202,10 +192,8 @@ void SeedClusterRemover::process(const TrackingRecHit *hit, float chi2, if (cluster.id() != pixelSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemover: pixel cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << pixelSourceProdID << ")\n."; + << "SeedClusterRemover: pixel cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << pixelSourceProdID << ")\n."; assert(cluster.id() == pixelSourceProdID); // DBG// cout << "HIT NEW PIXEL DETID = " << detid.rawId() << ", Cluster [ " @@ -213,8 +201,7 @@ void SeedClusterRemover::process(const TrackingRecHit *hit, float chi2, // endl; // if requested, cut on cluster size - if (pblocks_[subdet - 1].usesSize_ && - (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) + if (pblocks_[subdet - 1].usesSize_ && (cluster->pixels().size() > pblocks_[subdet - 1].maxSize_)) return; // mark as used @@ -224,31 +211,28 @@ void SeedClusterRemover::process(const TrackingRecHit *hit, float chi2, collectedPixels_[cluster.key()] = true; } else { // aka Strip - if (!doStrip_) return; + if (!doStrip_) + return; const type_info &hitType = typeid(*hit); if (hitType == typeid(SiStripRecHit2D)) { - const SiStripRecHit2D *stripHit = - static_cast(hit); + const SiStripRecHit2D *stripHit = static_cast(hit); // DBG// cout << "Plain RecHit 2D: " << endl; process(stripHit->omniClusterRef(), subdet); } else if (hitType == typeid(SiStripRecHit1D)) { const SiStripRecHit1D *hit1D = static_cast(hit); process(hit1D->omniClusterRef(), subdet); } else if (hitType == typeid(SiStripMatchedRecHit2D)) { - const SiStripMatchedRecHit2D *matchHit = - static_cast(hit); + const SiStripMatchedRecHit2D *matchHit = static_cast(hit); // DBG// cout << "Matched RecHit 2D: " << endl; process(matchHit->monoClusterRef(), subdet); process(matchHit->stereoClusterRef(), subdet); } else if (hitType == typeid(ProjectedSiStripRecHit2D)) { - const ProjectedSiStripRecHit2D *projHit = - static_cast(hit); + const ProjectedSiStripRecHit2D *projHit = static_cast(hit); // DBG// cout << "Projected RecHit 2D: " << endl; process(projHit->originalHit().omniClusterRef(), subdet); } else throw cms::Exception("NOT IMPLEMENTED") - << "Don't know how to handle " << hitType.name() << " on detid " - << detid.rawId() << "\n"; + << "Don't know how to handle " << hitType.name() << " on detid " << detid.rawId() << "\n"; } } @@ -284,9 +268,8 @@ void SeedClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { edm::Handle oldStrMask; iEvent.getByToken(oldPxlMaskToken_, oldPxlMask); iEvent.getByToken(oldStrMaskToken_, oldStrMask); - LogDebug("SeedClusterRemover") - << "to merge in, " << oldStrMask->size() << " strp and " - << oldPxlMask->size() << " pxl"; + LogDebug("SeedClusterRemover") << "to merge in, " << oldStrMask->size() << " strp and " << oldPxlMask->size() + << " pxl"; oldStrMask->copyMaskTo(collectedStrips_); oldPxlMask->copyMaskTo(collectedPixels_); assert(stripClusters->dataSize() >= collectedStrips_.size()); @@ -301,7 +284,8 @@ void SeedClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { for (auto const &seed : (*seeds)) { for (auto const &hit : seed.recHits()) { - if (!hit.isValid()) continue; + if (!hit.isValid()) + continue; process(&hit, 0., tgh); } } @@ -309,19 +293,15 @@ void SeedClusterRemover::produce(Event &iEvent, const EventSetup &iSetup) { pixels.clear(); strips.clear(); - LogDebug("SeedClusterRemover") - << "total strip to skip: " - << std::count(collectedStrips_.begin(), collectedStrips_.end(), true); - iEvent.put(std::make_unique( - edm::RefProd >(stripClusters), - collectedStrips_)); - - LogDebug("SeedClusterRemover") - << "total pxl to skip: " - << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); - iEvent.put(std::make_unique( - edm::RefProd >(pixelClusters), - collectedPixels_)); + LogDebug("SeedClusterRemover") << "total strip to skip: " + << std::count(collectedStrips_.begin(), collectedStrips_.end(), true); + iEvent.put(std::make_unique(edm::RefProd >(stripClusters), + collectedStrips_)); + + LogDebug("SeedClusterRemover") << "total pxl to skip: " + << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); + iEvent.put(std::make_unique(edm::RefProd >(pixelClusters), + collectedPixels_)); collectedStrips_.clear(); collectedPixels_.clear(); diff --git a/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemoverPhase2.cc b/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemoverPhase2.cc index 08a567c14d26c..72ceb86a3230e 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemoverPhase2.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/SeedClusterRemoverPhase2.cc @@ -27,22 +27,18 @@ // class SeedClusterRemoverPhase2 : public edm::stream::EDProducer<> { - public: +public: SeedClusterRemoverPhase2(const edm::ParameterSet &iConfig); void produce(edm::Event &iEvent, const edm::EventSetup &iSetup) override; - private: - edm::ESGetToken const - tTrackerGeom_; +private: + edm::ESGetToken const tTrackerGeom_; bool doOuterTracker_, doPixel_; bool mergeOld_; - typedef edm::ContainerMask > - PixelMaskContainer; - typedef edm::ContainerMask > - Phase2OTMaskContainer; + typedef edm::ContainerMask > PixelMaskContainer; + typedef edm::ContainerMask > Phase2OTMaskContainer; edm::EDGetTokenT > pixelClusters_; - edm::EDGetTokenT > - phase2OTClusters_; + edm::EDGetTokenT > phase2OTClusters_; edm::EDGetTokenT oldPxlMaskToken_; edm::EDGetTokenT oldPh2OTMaskToken_; edm::EDGetTokenT trajectories_; @@ -52,8 +48,7 @@ class SeedClusterRemoverPhase2 : public edm::stream::EDProducer<> { outerTrackerSourceProdID; // ProdIDs refs must point to (for consistency // tests) - inline void process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg); + inline void process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg); std::vector collectedOuterTrackers_; std::vector collectedPixels_; @@ -64,35 +59,25 @@ using namespace edm; SeedClusterRemoverPhase2::SeedClusterRemoverPhase2(const ParameterSet &iConfig) : tTrackerGeom_(esConsumes()), - doOuterTracker_(iConfig.existsAs("doOuterTracker") - ? iConfig.getParameter("doOuterTracker") - : true), - doPixel_(iConfig.existsAs("doPixel") - ? iConfig.getParameter("doPixel") - : true), + doOuterTracker_(iConfig.existsAs("doOuterTracker") ? iConfig.getParameter("doOuterTracker") : true), + doPixel_(iConfig.existsAs("doPixel") ? iConfig.getParameter("doPixel") : true), mergeOld_(iConfig.exists("oldClusterRemovalInfo")) { produces > >(); - produces< - edm::ContainerMask > >(); + produces > >(); - trajectories_ = consumes( - iConfig.getParameter("trajectories")); + trajectories_ = consumes(iConfig.getParameter("trajectories")); if (doPixel_) - pixelClusters_ = consumes >( - iConfig.getParameter("pixelClusters")); + pixelClusters_ = consumes >(iConfig.getParameter("pixelClusters")); if (doOuterTracker_) - phase2OTClusters_ = consumes >( - iConfig.getParameter("phase2OTClusters")); + phase2OTClusters_ = + consumes >(iConfig.getParameter("phase2OTClusters")); if (mergeOld_) { - oldPxlMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); - oldPh2OTMaskToken_ = consumes( - iConfig.getParameter("oldClusterRemovalInfo")); + oldPxlMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); + oldPh2OTMaskToken_ = consumes(iConfig.getParameter("oldClusterRemovalInfo")); } } -void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, - const TrackerGeometry *tg) { +void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, const TrackerGeometry *tg) { DetId detid = hit->geographicalId(); uint32_t subdet = detid.subdetId(); @@ -100,19 +85,17 @@ void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, const type_info &hitType = typeid(*hit); if (hitType == typeid(SiPixelRecHit)) { - if (!doPixel_) return; + if (!doPixel_) + return; const SiPixelRecHit *pixelHit = static_cast(hit); SiPixelRecHit::ClusterRef cluster = pixelHit->cluster(); - LogDebug("SeedClusterRemoverPhase2") - << "Plain Pixel RecHit in det " << detid.rawId(); + LogDebug("SeedClusterRemoverPhase2") << "Plain Pixel RecHit in det " << detid.rawId(); if (cluster.id() != pixelSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemoverPhase2: pixel cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << pixelSourceProdID << ")\n."; + << "SeedClusterRemoverPhase2: pixel cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << pixelSourceProdID << ")\n."; assert(cluster.id() == pixelSourceProdID); @@ -123,20 +106,17 @@ void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, collectedPixels_[cluster.key()] = true; } else if (hitType == typeid(Phase2TrackerRecHit1D)) { - if (!doOuterTracker_) return; + if (!doOuterTracker_) + return; - const Phase2TrackerRecHit1D *ph2OThit = - static_cast(hit); - LogDebug("SeedClusterRemoverPhase2") - << "Plain Phase2TrackerRecHit1D in det " << detid.rawId(); + const Phase2TrackerRecHit1D *ph2OThit = static_cast(hit); + LogDebug("SeedClusterRemoverPhase2") << "Plain Phase2TrackerRecHit1D in det " << detid.rawId(); Phase2TrackerRecHit1D::CluRef cluster = ph2OThit->cluster(); if (cluster.id() != outerTrackerSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << outerTrackerSourceProdID << ")\n."; + << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << outerTrackerSourceProdID << ")\n."; assert(cluster.id() == outerTrackerSourceProdID); @@ -145,20 +125,18 @@ void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, collectedOuterTrackers_[cluster.key()] = true; } else if (hitType == typeid(VectorHit)) { - if (!doOuterTracker_) return; + if (!doOuterTracker_) + return; const VectorHit *vhit = static_cast(hit); - LogDebug("SeedClusterRemoverPhase2") - << "Plain VectorHit in det " << detid.rawId(); + LogDebug("SeedClusterRemoverPhase2") << "Plain VectorHit in det " << detid.rawId(); // lower cluster Phase2TrackerRecHit1D::CluRef cluster = vhit->lowerCluster(); if (cluster.id() != outerTrackerSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << outerTrackerSourceProdID << ")\n."; + << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << outerTrackerSourceProdID << ")\n."; OTs[cluster.key()] = false; assert(collectedOuterTrackers_.size() > cluster.key()); @@ -168,23 +146,19 @@ void SeedClusterRemoverPhase2::process(const TrackingRecHit *hit, float chi2, cluster = vhit->upperCluster(); if (cluster.id() != outerTrackerSourceProdID) throw cms::Exception("Inconsistent Data") - << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " - << cluster.id() - << " does not match with source cluster collection (ID = " - << outerTrackerSourceProdID << ")\n."; + << "SeedClusterRemoverPhase2: strip cluster ref from Product ID = " << cluster.id() + << " does not match with source cluster collection (ID = " << outerTrackerSourceProdID << ")\n."; OTs[cluster.key()] = false; assert(collectedOuterTrackers_.size() > cluster.key()); collectedOuterTrackers_[cluster.key()] = true; } else - throw cms::Exception("NOT IMPLEMENTED") - << "I received a hit that was neither SiPixelRecHit nor " - "Phase2TrackerRecHit1D but " - << hitType.name() << " on detid " << detid.rawId() << "\n"; + throw cms::Exception("NOT IMPLEMENTED") << "I received a hit that was neither SiPixelRecHit nor " + "Phase2TrackerRecHit1D but " + << hitType.name() << " on detid " << detid.rawId() << "\n"; } -void SeedClusterRemoverPhase2::produce(Event &iEvent, - const EventSetup &iSetup) { +void SeedClusterRemoverPhase2::produce(Event &iEvent, const EventSetup &iSetup) { ProductID pixelOldProdID, stripOldProdID; const auto &tgh = &iSetup.getData(tTrackerGeom_); @@ -194,16 +168,14 @@ void SeedClusterRemoverPhase2::produce(Event &iEvent, iEvent.getByToken(pixelClusters_, pixelClusters); pixelSourceProdID = pixelClusters.id(); } - LogDebug("SeedClusterRemoverPhase2") - << "Read pixel with id " << pixelSourceProdID << std::endl; + LogDebug("SeedClusterRemoverPhase2") << "Read pixel with id " << pixelSourceProdID << std::endl; Handle > phase2OTClusters; if (doOuterTracker_) { iEvent.getByToken(phase2OTClusters_, phase2OTClusters); outerTrackerSourceProdID = phase2OTClusters.id(); } - LogDebug("SeedClusterRemoverPhase2") - << "Read OT cluster with id " << outerTrackerSourceProdID << std::endl; + LogDebug("SeedClusterRemoverPhase2") << "Read OT cluster with id " << outerTrackerSourceProdID << std::endl; if (doOuterTracker_) { OTs.resize(phase2OTClusters->dataSize()); @@ -218,9 +190,8 @@ void SeedClusterRemoverPhase2::produce(Event &iEvent, edm::Handle oldPh2OTMask; iEvent.getByToken(oldPxlMaskToken_, oldPxlMask); iEvent.getByToken(oldPh2OTMaskToken_, oldPh2OTMask); - LogDebug("SeedClusterRemoverPhase2") - << "to merge in, " << oldPh2OTMask->size() << " strp and " - << oldPxlMask->size() << " pxl"; + LogDebug("SeedClusterRemoverPhase2") << "to merge in, " << oldPh2OTMask->size() << " strp and " + << oldPxlMask->size() << " pxl"; oldPh2OTMask->copyMaskTo(collectedOuterTrackers_); oldPxlMask->copyMaskTo(collectedPixels_); assert(phase2OTClusters->dataSize() >= collectedOuterTrackers_.size()); @@ -236,7 +207,8 @@ void SeedClusterRemoverPhase2::produce(Event &iEvent, for (auto const &seed : (*seeds)) { for (auto const &hit : seed.recHits()) { - if (!hit.isValid()) continue; + if (!hit.isValid()) + continue; process(&hit, 0., tgh); } } @@ -244,21 +216,16 @@ void SeedClusterRemoverPhase2::produce(Event &iEvent, pixels.clear(); OTs.clear(); - LogDebug("SeedClusterRemoverPhase2") - << "total strip to skip: " - << std::count(collectedOuterTrackers_.begin(), - collectedOuterTrackers_.end(), true); + LogDebug("SeedClusterRemoverPhase2") << "total strip to skip: " + << std::count( + collectedOuterTrackers_.begin(), collectedOuterTrackers_.end(), true); iEvent.put(std::make_unique( - edm::RefProd >( - phase2OTClusters), - collectedOuterTrackers_)); - - LogDebug("SeedClusterRemoverPhase2") - << "total pxl to skip: " - << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); - iEvent.put(std::make_unique( - edm::RefProd >(pixelClusters), - collectedPixels_)); + edm::RefProd >(phase2OTClusters), collectedOuterTrackers_)); + + LogDebug("SeedClusterRemoverPhase2") << "total pxl to skip: " + << std::count(collectedPixels_.begin(), collectedPixels_.end(), true); + iEvent.put(std::make_unique(edm::RefProd >(pixelClusters), + collectedPixels_)); collectedOuterTrackers_.clear(); collectedPixels_.clear(); diff --git a/RecoLocalTracker/SubCollectionProducers/src/StripClusterSelectorTopBottom.cc b/RecoLocalTracker/SubCollectionProducers/src/StripClusterSelectorTopBottom.cc index 646ecdf62453f..489df988b22c7 100644 --- a/RecoLocalTracker/SubCollectionProducers/src/StripClusterSelectorTopBottom.cc +++ b/RecoLocalTracker/SubCollectionProducers/src/StripClusterSelectorTopBottom.cc @@ -2,8 +2,7 @@ #include "FWCore/Framework/interface/MakerMacros.h" -void StripClusterSelectorTopBottom::produce( - edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const { +void StripClusterSelectorTopBottom::produce(edm::StreamID, edm::Event& event, const edm::EventSetup& setup) const { edm::Handle> input; event.getByToken(token_, input); @@ -11,24 +10,19 @@ void StripClusterSelectorTopBottom::produce( auto output = std::make_unique>(); - for (edmNew::DetSetVector::const_iterator clustSet = - input->begin(); - clustSet != input->end(); ++clustSet) { + for (edmNew::DetSetVector::const_iterator clustSet = input->begin(); clustSet != input->end(); + ++clustSet) { edmNew::DetSet::const_iterator clustIt = clustSet->begin(); edmNew::DetSet::const_iterator end = clustSet->end(); DetId detIdObject(clustSet->detId()); - edmNew::DetSetVector::FastFiller spc(*output, - detIdObject.rawId()); - const StripGeomDetUnit* theGeomDet = - dynamic_cast(theTracker.idToDet(detIdObject)); - const StripTopology* topol = - dynamic_cast(&(theGeomDet->specificTopology())); + edmNew::DetSetVector::FastFiller spc(*output, detIdObject.rawId()); + const StripGeomDetUnit* theGeomDet = dynamic_cast(theTracker.idToDet(detIdObject)); + const StripTopology* topol = dynamic_cast(&(theGeomDet->specificTopology())); for (; clustIt != end; ++clustIt) { LocalPoint lpclust = topol->localPosition(clustIt->barycenter()); - GlobalPoint GPclust = theGeomDet->surface().toGlobal( - Local3DPoint(lpclust.x(), lpclust.y(), lpclust.z())); + GlobalPoint GPclust = theGeomDet->surface().toGlobal(Local3DPoint(lpclust.x(), lpclust.y(), lpclust.z())); double clustY = GPclust.y(); if ((clustY * y_) > 0) { spc.push_back(*clustIt);