CMS 3D CMS Logo

 All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Properties Friends Macros Groups Pages
Typedefs | Functions | Variables
CAHitNtupletGeneratorKernelsImpl.h File Reference
#include <cmath>
#include <cstdint>
#include <limits>
#include <cuda_runtime.h>
#include "HeterogeneousCore/CUDAUtilities/interface/cudaCheck.h"
#include "HeterogeneousCore/CUDAUtilities/interface/cuda_assert.h"
#include "RecoLocalTracker/SiPixelRecHits/interface/pixelCPEforGPU.h"
#include "CAConstants.h"
#include "CAHitNtupletGeneratorKernels.h"
#include "GPUCACell.h"
#include "gpuFishbone.h"
#include "gpuPixelDoublets.h"

Go to the source code of this file.

Typedefs

using HitContainer = pixelTrack::HitContainer
 
using HitsOnCPU = TrackingRecHit2DGPU
 
using HitsOnGPU = TrackingRecHit2DSOAView
 
using HitToTuple = caConstants::HitToTuple
 
using Quality = pixelTrack::Quality
 
using TkSoA = pixelTrack::TrackSoA
 
using TupleMultiplicity = caConstants::TupleMultiplicity
 

Functions

 __attribute__ ((always_inline)) void kernel_checkOverflows(HitContainer const *foundNtuplets
 
 assert (nCells)
 
 for (int idx=first, nt=(*nCells);idx< nt;idx+=gridDim.x *blockDim.x)
 
 if (0==first)
 
tracks setNTracks (ntracks)
 

Variables

caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
apc
 
auto & c = *counters
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const * 
cellNeighbors
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__ 
cells
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const
gpuPixelDoublets::CellTracksVector
const * 
cellTracks
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const
gpuPixelDoublets::CellTracksVector
const
GPUCACell::OuterHitOfCell
const int32_t uint32_t
CAHitNtupletGeneratorKernelsGPU::Counters
counters
 
TkSoA const *__restrict__
CAHitNtupletGeneratorKernelsGPU::QualityCuts 
cuts
 
uint32_t const *__restrict__
TkSoA const *__restrict__
Quality bool 
dupPassThrough
 
auto first = threadIdx.x + blockIdx.x * blockDim.x
 
HitContainer const
*__restrict__ TkSoA const
*__restrict__ Quality const
*__restrict__
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const *__restrict__ int32_t 
firstPrint
 
auto const & foundNtuplets = *ptuples
 
auto const good = pixelTrack::Quality::strict
 min quality of good More...
 
auto const & hh = *hhp
 
TrackingRecHit2DSOAView const
*__restrict__ 
hhp
 
TrackingRecHit2DSOAView const
*__restrict__ HitContainer
*__restrict__ 
hitDetIndices
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const * 
hitToTuple
 
HitContainer const
*__restrict__ TkSoA const
*__restrict__ Quality const
*__restrict__
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const *__restrict__ int32_t
int32_t int 
iev
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const
gpuPixelDoublets::CellTracksVector
const
GPUCACell::OuterHitOfCell
const 
isOuterHitOfCell
 
int l1end = hh.hitsLayerStart()[1]
 
HitContainer const
*__restrict__ TkSoA const
*__restrict__ Quality const
*__restrict__
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const *__restrict__ int32_t
int32_t 
lastPrint
 
auto const longTqual = pixelTrack::Quality::highPurity
 
constexpr auto loose = pixelTrack::Quality::loose
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const
gpuPixelDoublets::CellTracksVector
const
GPUCACell::OuterHitOfCell
const int32_t uint32_t 
maxNumberOfDoublets
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__ 
nCells
 
caConstants::TupleMultiplicity
const
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const
cms::cuda::AtomicPairCounter
GPUCACell const *__restrict__
uint32_t const *__restrict__
gpuPixelDoublets::CellNeighborsVector
const
gpuPixelDoublets::CellTracksVector
const
GPUCACell::OuterHitOfCell
const int32_t 
nHits
 
auto nhits = hh.nHits()
 
Quality *__restrict__ uint16_t nmin
 
auto ntracks = apc->get().m
 
HitContainer const
*__restrict__ Quality const
*__restrict__
CAHitNtupletGeneratorKernelsGPU::HitToTuple
const *__restrict__ 
phitToTuple
 
uint32_t const *__restrict__
TkSoA const *__restrict__ 
ptracks
 
HitContainer const *__restrict__ ptuples
 
uint32_t const *__restrict__
Quality
quality
 
auto const reject = dupPassThrough ? loose : dup
 
auto const & tracks = *ptracks
 cannot be loose More...
 
caConstants::TupleMultiplicity
const * 
tupleMultiplicity
 
HitContainer const *__restrict__ tuples
 

Typedef Documentation

Definition at line 32 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 25 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 24 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 27 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 30 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 31 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 28 of file CAHitNtupletGeneratorKernelsImpl.h.

Function Documentation

__attribute__ ( (always_inline)  ) const
inline
assert ( nCells  )
for ( int  idx = first)

Definition at line 102 of file CAHitNtupletGeneratorKernelsImpl.h.

References cms::cudacompat::atomicAdd(), c, and gpuVertexFinder::printf().

102  {
103  auto const &thisCell = cells[idx];
104  if (thisCell.hasFishbone() && !thisCell.isKilled())
105  atomicAdd(&c.nFishCells, 1);
106  if (thisCell.outerNeighbors().full()) //++tooManyNeighbors[thisCell.theLayerPairId];
107  printf("OuterNeighbors overflow %d in %d\n", idx, thisCell.layerPairId());
108  if (thisCell.tracks().full()) //++tooManyTracks[thisCell.theLayerPairId];
109  printf("Tracks overflow %d in %d\n", idx, thisCell.layerPairId());
110  if (thisCell.isKilled())
111  atomicAdd(&c.nKilledCells, 1);
112  if (!thisCell.unused())
113  atomicAdd(&c.nEmptyCells, 1);
114  if ((0 == hitToTuple->size(thisCell.inner_hit_id())) && (0 == hitToTuple->size(thisCell.outer_hit_id())))
115  atomicAdd(&c.nZeroTrackCells, 1);
116  }
const edm::EventSetup & c
caConstants::TupleMultiplicity const CAHitNtupletGeneratorKernelsGPU::HitToTuple const * hitToTuple
printf("params %d %f %f %f\n", minT, eps, errmax, chi2max)
caConstants::TupleMultiplicity const CAHitNtupletGeneratorKernelsGPU::HitToTuple const cms::cuda::AtomicPairCounter GPUCACell const *__restrict__ cells
T1 atomicAdd(T1 *a, T2 b)
Definition: cudaCompat.h:61
if ( = first)

Definition at line 58 of file CAHitNtupletGeneratorKernelsImpl.h.

References cms::cudacompat::atomicAdd(), c, cms::cuda::AtomicPairCounter::get(), and cms::cuda::AtomicPairCounter::Counters::m.

58  {
59  atomicAdd(&c.nEvents, 1);
60  atomicAdd(&c.nHits, nHits);
61  atomicAdd(&c.nCells, *nCells);
62  atomicAdd(&c.nTuples, apc->get().m);
63  atomicAdd(&c.nFitTracks, tupleMultiplicity->size());
64  }
const edm::EventSetup & c
caConstants::TupleMultiplicity const *__restrict__ tupleMultiplicity
caConstants::TupleMultiplicity const CAHitNtupletGeneratorKernelsGPU::HitToTuple const cms::cuda::AtomicPairCounter * apc
caConstants::TupleMultiplicity const CAHitNtupletGeneratorKernelsGPU::HitToTuple const cms::cuda::AtomicPairCounter GPUCACell const *__restrict__ uint32_t const *__restrict__ gpuPixelDoublets::CellNeighborsVector const gpuPixelDoublets::CellTracksVector const GPUCACell::OuterHitOfCell const int32_t nHits
caConstants::TupleMultiplicity const CAHitNtupletGeneratorKernelsGPU::HitToTuple const cms::cuda::AtomicPairCounter GPUCACell const *__restrict__ uint32_t const *__restrict__ nCells
T1 atomicAdd(T1 *a, T2 b)
Definition: cudaCompat.h:61
tracks setNTracks ( ntracks  )

Variable Documentation

Initial value:
{
auto &tracks = *ptracks
uint32_t const *__restrict__ TkSoA const *__restrict__ ptracks
auto const & tracks
cannot be loose

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

auto& c = *counters

Definition at line 56 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

Referenced by DTVDriftCalibration::cellInfo::add(), FWHGCalTriggerCellProxyBuilder::build(), FWHGCalTriggerClusterProxyBuilder::build(), HGCScintillatorDetId::detectorCells(), CaloSubdetectorGeometry::getCellSet(), HGCalTriggerGeometryV9Imp2::getCellsFromModule(), HGCalTriggerGeometryV9Imp3::getCellsFromModule(), HcalDDDRecConstants::getHFCellParameters(), HGCalTriggerGeometryV9Imp2::getOrderedCellsFromModule(), HGCalTriggerGeometryV9Imp3::getOrderedCellsFromModule(), HcalDDDRecConstants::HcalCellTypes(), HGCSiliconDetIdToROC::HGCSiliconDetIdToROC(), Grid1D::index(), HcalHardcodeGeometryLoader::makeHECells_H2(), HcalHardcodeGeometryLoader::makeHFCells(), HcalFlexiHardcodeGeometryLoader::makeHFCells(), HcalHardcodeGeometryLoader::makeHOCells(), HGCalDDDConstants::maxCells(), HGCalGeometry::newCell(), spr::newECALIdEW(), spr::newECALIdNS(), HGCalDDDConstants::numberCells(), CastorHardcodeCalibrations::produceChannelQuality(), HcalHardcodeCalibrations::produceChannelQuality(), HcalHardcodeCalibrations::produceElectronicsMap(), HcalHardcodeCalibrations::produceFlagHFDigiTimeParams(), HcalHardcodeCalibrations::produceFrontEndMap(), CastorHardcodeCalibrations::produceGains(), HcalHardcodeCalibrations::produceGains(), CastorHardcodeCalibrations::produceGainWidths(), HcalHardcodeCalibrations::produceGainWidths(), HcalHardcodeCalibrations::produceL1TriggerObjects(), HcalHardcodeCalibrations::produceLongRecoParams(), HcalHardcodeCalibrations::produceLUTCorrs(), HcalHardcodeCalibrations::produceLutMetadata(), HcalHardcodeCalibrations::produceMCParams(), CastorHardcodeCalibrations::producePedestals(), HcalHardcodeCalibrations::producePedestals_(), CastorHardcodeCalibrations::producePedestalWidths(), HcalHardcodeCalibrations::producePedestalWidths_(), HcalHardcodeCalibrations::producePFCorrs(), CastorHardcodeCalibrations::produceQIEData(), HcalHardcodeCalibrations::produceQIEData(), HcalHardcodeCalibrations::produceQIETypes(), CastorHardcodeCalibrations::produceRecoParams(), HcalHardcodeCalibrations::produceRecoParams(), HcalHardcodeCalibrations::produceRespCorrs(), CastorHardcodeCalibrations::produceSaturationCorrs(), HcalHardcodeCalibrations::produceSiPMParameters(), HcalHardcodeCalibrations::produceTimeCorrs(), HcalHardcodeCalibrations::produceTimingParams(), HcalHardcodeCalibrations::produceTPChannelParameters(), HcalHardcodeCalibrations::produceValidationCorrs(), HcalHardcodeCalibrations::produceZDCLowGainFractions(), HcalHardcodeCalibrations::produceZSThresholds(), HGCalCLUEAlgoT< TILE >::reset(), HGCalTriggerGeometryV9Imp2::validTriggerCellFromCells(), and HGCalTriggerGeometryV9Imp3::validTriggerCellFromCells().

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

Quality const *__restrict__ CAHitNtupletGeneratorKernelsGPU::Counters * counters
Quality *__restrict__ uint16_t bool dupPassThrough
Initial value:

Definition at line 144 of file CAHitNtupletGeneratorKernelsImpl.h.

Definition at line 127 of file CAHitNtupletGeneratorKernelsImpl.h.

Referenced by gpuPixelRecHits::__attribute__(), cms::cuda::__attribute__(), cms::cuda::OneToManyAssoc< hindex_type, S+1, 5 *S >::__attribute__(), optutl::CommandLineParser::_setVariablesFromFile(), HGCDigitizer::accumulate(), TkTrackingRegionsMargin< float >::add(), ReferenceTrajectory::addMaterialEffectsBrl(), PrintGeomSummary::addName(), BaseProtonTransport::addPartToHepMC(), StripCPEgeometric::WrappedCluster::addSuppressedEdgeStrip(), SiStripGainFromCalibTree::algoEndRun(), algorithm(), CalibrationAlgorithm::analyse(), VpspScanAlgorithm::analyse(), PFCandConnector::analyseNuclearWPrim(), PFCandConnector::analyseNuclearWSec(), EcalDQMonitorTask::analyze(), SiStripDetVOffTrendPlotter::analyze(), BxTiming::analyze(), HcalQIEDataCheck::analyze(), TopMonitor::analyze(), HcalPedestalWidthsCheck::analyze(), TestOutliers::analyze(), TestTrackHits::analyze(), HeavyFlavorValidation::analyze(), HLTScalers::analyze(), SiPixelErrorsDigisToCalibDigis::analyze(), EcalTPGParamBuilder::analyze(), TauTagValidation::analyze(), IsolatedTracksNxN::analyze(), EcalSelectiveReadoutValidation::analyzeEB(), EcalSelectiveReadoutValidation::analyzeEE(), CaloSimHitAnalysis::analyzeHits(), L1GtAnalyzer::analyzeL1GtUtilsCore(), SiStripBadAPVAlgorithmFromClusterOccupancy::AnalyzeOccupancy(), SiStripBadAPVandHotStripAlgorithmFromClusterOccupancy::AnalyzeOccupancy(), HGCalTBAnalyzer::analyzeRecHits(), HGCalTimingAnalyzer::analyzeSimHits(), HGCalTBAnalyzer::analyzeSimHits(), AlignmentProducerBase::applyDB(), SurveyDataConverter::applyFineSurveyInfo(), ticl::assignPCAtoTracksters(), TrackerHitAssociator::associateHit(), TrackerHitAssociator::associateSimpleRecHitCluster(), iterator_pair_as_a_range< I >::begin(), edm::beginGlobalTransitionAsync(), TkAlCaSkimTreeMerger::beginJob(), PickEvents::beginJob(), edm::EventProcessor::beginLumiAsync(), EcalMonitorPrescaler::beginRun(), IsoTrackCalib::beginRun(), edm::EventProcessor::beginRun(), edm::beginStreamTransitionAsync(), SiStripGainFromCalibTree::bookDQMHistos(), QIE11Task::bookHistograms(), BxTiming::bookHistograms(), SiStripGainsCalibTreeWorker::bookHistograms(), HeavyFlavorValidation::bookHistograms(), SiStripGainsPCLWorker::bookHistograms(), MVAJetPuId::bookReader(), SiStripBaseCondObjDQM::bookSummaryCumulMEs(), SiStripBaseCondObjDQM::bookSummaryMEs(), SiStripBaseCondObjDQM::bookSummaryProfileMEs(), HcalSiPM::BorelCDF(), magneticfield::bRod::bRod(), magneticfield::bSector::bSector(), CSCGeometryParsFromDD::build(), magneticfield::MagGeoBuilder::build(), MagGeoBuilderFromDDD::build(), RealisticSimClusterMapper::buildClusters(), SiStripDetVOffBuilder::BuildDetVOffObj(), RunRangeDependentPedeLabeler::buildRunRangeDependencyMap(), SeedFromGenericPairOrTriplet::buildSeed(), SiTrackerMultiRecHitUpdator::calcParameters(), FWInvMassDialog::Calculate(), JetPartonMatching::calculate(), ClusterShapeAlgo::Calculate_BarrelBasketEnergyFraction(), gpuCalibPixel::calibDigis(), gpuCalibPixel::calibDigisPhase2(), PedeLabelerBase::calibrationLabel(), StripCPEgeometric::WrappedCluster::centroid(), HiggsValidation::MonitoredDecays::channel(), HBHENegativeEFilter::checkPassFilter(), CommonHcalNoiseRBXData::CheckPassFilter(), HBHEPulseShapeFlagSetter::CheckPassFilter(), TIBRing::checkPeriodicity(), TrackTransformer::checkRecHitsOrdering(), TrackExtenderWithMTDT< TrackCollection >::checkRecHitsOrdering(), SiStripAPVRestorer::cleaner_LocalMinimumAdder(), TrackCleaner::cleanTracks(), PixelTrackCleanerBySharedHits::cleanTracks(), EcalClusterToolsT< noZS >::cluster2ndMoments(), HGCalHistoClusteringImpl::clusterSeedMulticluster(), SimplePlan1RechitCombiner::combine(), AlignableObjectId::commonGeometry(), L1MuGMTExtendedCand::compareRank(), DynamicTruncation::compatibleDets(), MuonSeedOrcaPatternRecognition::complete(), L1TLSBlock::computeErrorFromRange(), GaussianSumUtilities1D::computeMode(), SiStripFecCabling::connections(), DDHGCalTBModuleX::constructBlocks(), DDHCalBarrelAlgo::constructGeneralVolume(), DDHCalEndcapAlgo::constructGeneralVolume(), DDHCalBarrelAlgo::constructInsideDetectors(), HcalBarrelAlgo::constructInsideDetectors(), DDHCalBarrelAlgo::constructInsideLayers(), DDHCalEndcapModuleAlgo::constructInsideModule(), DDHCalEndcapAlgo::constructInsideModule(), HCalEndcapAlgo::constructInsideModule(), DDHCalEndcapModuleAlgo::constructInsideModule0(), DDHCalEndcapAlgo::constructInsideModule0(), HCalEndcapAlgo::constructInsideModule0(), DDHCalBarrelAlgo::constructInsideSector(), DDHCalEndcapAlgo::constructInsideSector(), DDAHcalModuleAlgo::constructLayers(), DDHGCalTBModule::constructLayers(), DDHGCalTBModuleX::constructLayers(), DDHGCalModule::constructLayers(), DDHGCalModuleAlgo::constructLayers(), DDHGCalEEAlgo::constructLayers(), DDHGCalHEAlgo::constructLayers(), DDHGCalHEFileAlgo::constructLayers(), DDHGCalEEFileAlgo::constructLayers(), DDHGCalMixLayer::constructLayers(), DDHGCalSiliconModule::constructLayers(), DDHCalBarrelAlgo::constructMidLayer(), DDHCalEndcapModuleAlgo::constructScintLayer(), DDHCalEndcapAlgo::constructScintLayer(), DDHCalBarrelAlgo::constructSideLayer(), convertHB(), FFTJetCorrectorSequence< Jet, InitialConverter, FinalConverter >::correct(), EcalHitMaker::cracksPads(), PFAlgo::createCandidatesHCAL(), RPCStripsRing::createRefConnections(), RPCSeedPattern::createSeed(), RPCReadOutMapping::dccNumberRange(), DEutils< T >::de_find(), MuonOffsetFromDD::debugParameters(), GenericTruncatedAverageDeDxEstimator::dedx(), StripCPEgeometric::WrappedCluster::deformed(), HFSimpleTimeCheck::determineAnodeStatus(), CSCOfflineMonitor::doEfficiencies(), CSCValidation::doEfficiencies(), edm::SubProcess::doEndProcessBlockAsync(), VectorDoublet< Vector3D, Vector3D >::dot(), EmissionVetoHook1::doVetoMPIStep(), MultiTrackValidatorGenPs::dqmAnalyze(), MultiTrackValidator::dqmAnalyze(), DTDCSByLumiTask::dqmEndLuminosityBlock(), CorrPCCProducer::dqmEndRunProduce(), drawMap(), StripCPEgeometric::WrappedCluster::dropSmallerEdgeStrip(), triggerExpression::L1uGTReader::dump(), triggerExpression::PathReader::dump(), MuScleFit::duringFastLoop(), InvRingCalib::duringLoop(), EcalEleCalibLooper::duringLoop(), ZeeCalibration::duringLoop(), EcalClusterToolsT< noZS >::e1x3(), EcalClusterToolsT< noZS >::e1x5(), EcalClusterToolsT< noZS >::e2nd(), EcalClusterToolsT< noZS >::e2x2(), EcalClusterToolsT< noZS >::e2x5Bottom(), EcalClusterToolsT< noZS >::e2x5Left(), EcalClusterToolsT< noZS >::e2x5Max(), EcalClusterToolsT< noZS >::e2x5Right(), EcalClusterToolsT< noZS >::e2x5Top(), EcalClusterToolsT< noZS >::e3x1(), EcalClusterToolsT< noZS >::e3x2(), EcalClusterToolsT< noZS >::e3x3(), EcalClusterToolsT< noZS >::e4x4(), EcalClusterToolsT< noZS >::e5x1(), EcalClusterToolsT< noZS >::e5x5(), EcalClusterToolsT< noZS >::eBottom(), EcalMonitorPrescaler::EcalMonitorPrescaler(), EcalClusterToolsT< noZS >::eLeft(), TRange< int >::empty(), PixelRecoRange< float >::empty(), muonisolation::Range< float >::empty(), edm::endGlobalTransitionAsync(), FilterOutLowPt::endJob(), DTT0Calibration::endJob(), SplitVertexResolution::endJob(), IsoTrig::endJob(), IsolatedTracksNxN::endJob(), GeneralPurposeTrackAnalyzer::endJob(), DMRChecker::endJob(), SimG4FluxProducer::endOfEvent(), HGCPassive::endOfEvent(), AlcaBeamMonitorClient::endRun(), edm::EventProcessor::endRun(), edm::endStreamTransitionAsync(), EcalClusterToolsT< noZS >::energyBasketFractionEta(), EcalClusterToolsT< noZS >::energyBasketFractionPhi(), cond::serialization::access< T[N]>::equal_(), MillePedeMonitor::equidistLogBins(), CosmicGenFilterHelix::equidistLogBins(), EcalClusterToolsT< noZS >::eRight(), magneticfield::eSector::eSector(), StripCPEgeometric::WrappedCluster::eta(), EcalClusterToolsT< noZS >::eTop(), GenericMVAComputer::eval(), trklet::TrackletEventProcessor::event(), TtEvent::eventHypo(), DDHGCalCell::execute(), DDHGCalWafer8::execute(), trklet::PurgeDuplicate::execute(), DDTOBRodAlgo::execute(), DDTrackerZPosAlgo::execute(), DDTrackerXYZPosAlgo::execute(), DDTIDModulePosAlgo::execute(), DDHGCalWaferP::execute(), DDTrackerAngular::execute(), DDTrackerLinear::execute(), DDTrackerPhiAlgo::execute(), DDTrackerPhiAltAlgo::execute(), DDHCalLinearXY::execute(), DDTIDRingAlgo::execute(), DDPixFwdDiskAlgo::execute(), DDTECOptoHybAlgo::execute(), DDTECPhiAlgo::execute(), DDTECPhiAltAlgo::execute(), DDTIDAxialCableAlgo::execute(), DDTIDModuleAlgo::execute(), DDTOBAxCableAlgo::execute(), DDTOBRadCableAlgo::execute(), DDPixPhase1FwdDiskAlgo::execute(), DDHCalTBZposAlgo::execute(), DDHGCalWaferF::execute(), DDHGCalWaferFullRotated::execute(), DDHCalForwardAlgo::execute(), DDPixBarLayerAlgo::execute(), DDPixBarLayerUpgradeAlgo::execute(), DDTIBLayerAlgo::execute(), DDBHMAngular::execute(), DDHCalTBCableAlgo::execute(), DDTECCoolAlgo::execute(), DDHCalFibreBundle::execute(), DDTECModuleAlgo::execute(), DDTrackerIrregularRingAlgo::execute(), DDTrackerRingAlgo::execute(), DDPixFwdBlades::execute(), HistogramManager::executeExtend(), PropagateToMuon::extrapolate(), pos::PixelCalibConfiguration::fedCardsAndChannels(), CSCValHists::fill2DProfile(), SiPixelGainCalibrationRejectNoisyAndDead::fillDatabase(), DQMProvInfo::fillDcsBitsFromDcsStatusCollection(), pat::PATElectronProducer::fillElectron(), pat::PATElectronProducer::fillElectron2(), EcalEBTrigPrimTestAlgo::fillMap(), EcalTrigPrimFunctionalAlgo::fillMap(), SiStripThresholdDQM::fillMEsForLayer(), SimTrackManager::fillMotherList(), pat::PATMuonProducer::fillMuon(), MsgTools::fillNames(), LumiCorrectionSource::fillparamcache(), SiPixelUtility::fillPaveText(), ThroughputServiceClient::fillSummaryPlots(), ticl::ClusterFilterByAlgoAndSizeAndLayerRange::filter(), ProbeTreeProducer::filter(), JetHTJetPlusHOFilter::filter(), HBHELinearMap::find(), mkfit::MkBuilder::find_tracks_in_layers(), EcalTBCrystalMap::findCrystalAngles(), reco::findMethod(), FFTJetPFPileupCleaner::findSomeVertexWFakes(), HoughGrouping::findTheMaxima(), mkfit::MkBuilder::findTracksStandard(), SymmetryFit::findUsableMinMax(), FWDetailViewManager::findViewersFor(), Alignable::firstCompsWithParams(), mkfit::MkBuilder::fit_cands(), alpgen::fixEventHiggsTTbar(), alpgen::fixEventSingleTop(), alpgen::fixEventTTbar(), gpuClustering::for(), FWEveViewManager::FWEveViewManager(), SiStripGainsPCLHarvester::gainQualityMonitor(), EcalHitMaker::gapsLifting(), models::generate(), gen::Pythia6JetGun::generateEvent(), SuepShower::generateShower(), IterativeHelixExtrapolatorToLine::genericPathLength(), l1t::L1TGlobalUtil::getAlgNameFromBit(), NoisyChannel::getAverage(), cond::getDbCredentials(), SiStripPsuDetIdMap::getDcuId(), JME::getDefinitionLine(), SiStripPsuDetIdMap::getDetectorLocation(), EcalClusterLazyToolsBase::getEcalRecHitCollection(), hcaldqm::quantity::getEid_FED(), hcaldqm::quantity::getEid_FEDuTCA(), HcalDDDRecConstants::getEtaBins(), l1t::L1TGlobalUtil::getFinalDecisionByBit(), EcalClusterToolsT< noZS >::getFraction(), l1t::L1TGlobalUtil::getInitialDecisionByBit(), l1t::L1TGlobalUtil::getIntermDecisionByBit(), CSCStripElectronicsSim::getKeyStripsFromMC(), TtHadEvtSolution::getLRJetCombObsVal(), TtSemiEvtSolution::getLRJetCombObsVal(), TtDilepEvtSolution::getLRSignalEvtObsVal(), TtHadEvtSolution::getLRSignalEvtObsVal(), TtSemiEvtSolution::getLRSignalEvtObsVal(), l1t::L1TGlobalUtil::getMaskByBit(), EcalClusterToolsT< noZS >::getMaximum(), ForwardName::getName(), FillInfoPopConSourceHandler::getNewObjects(), trigger::TriggerEventWithRefs::getObjects(), HcalDDDRecConstants::getOneEtaBin(), SimTrackManager::getOrCreateVertex(), edm::ParameterSet::getParameterNames(), HcalDDDRecConstants::getPhis(), l1t::L1TGlobalUtil::getPrescaleByBit(), SimpleL1MuGMTCand::getRank(), OMTFConfiguration::getRegionNumberFromMap(), HcalDDDRecConstants::getRZ(), HGCalTriggerGeometryV9Imp2::getTriggerCellsFromModule(), ZCountingTrigger::TTrigger::getTriggerObjectBit(), cond::persistency::KeyList::getUsingIndex(), EcalClusterCrackCorrection::getValue(), ZIterativeAlgorithmWithFit::getWeight(), CSCValidation::getWidth(), hgcal::ClusterTools::getWidths(), GenWeightsTableProducer::globalBeginRun(), edm::EventProcessor::globalEndLumiAsync(), AlcaBeamMonitor::globalEndLuminosityBlock(), CachingSeedCleanerBySharedInput::good(), graph_combine(), HcalIndexLookup::hasDuplicateIds(), HcalDDDRecConstants::HcalCellTypes(), HcalPiecewiseLinearFunctor::HcalPiecewiseLinearFunctor(), HcalTopology::HcalTopology(), CSCMake2DRecHit::hitFromStripAndWire(), PixelTripletLargeTipGenerator::hitTriplets(), HLTEgammaEtFilterPairs::hltFilter(), l1tVertexFinder::VertexFinder::HPV(), SimTrackManager::idSavedTrack(), L1TPhase2CorrelatorOffline::InCone::InCone(), MixCollection< T >::init(), init_filter(), HcalLayerDepthMap::initialize(), L1GtUtils::LogicalExpressionL1Results::initialize(), SiPixelFedCablingMap::initializeRocs(), SimpleVFATFrameCollection::InsertEmptyFrame(), muonisolation::Range< float >::inside(), TRange< int >::inside(), PixelRecoRange< float >::inside(), EcalTBDaqFormatter::interpretRawData(), EcalTB07DaqFormatter::interpretRawData(), TtGenEvent::isFullLeptonic(), RPCFw::isMajor(), npstat::isNonDecreasing(), npstat::isNonIncreasing(), npstat::isStrictlyDecreasing(), AbsHcalFunctor::isStrictlyDecreasing(), npstat::isStrictlyIncreasing(), AbsHcalFunctor::isStrictlyIncreasing(), L1CaloJetProducer::L1CaloJetProducer(), Alignable::lastCompsWithParams(), HGVHistoProducerAlgo::layerClusters_to_CaloParticles(), HGVHistoProducerAlgo::layerClusters_to_SimClusters(), CastorLedAnalysis::LedCastorHists(), HcalLedAnalysis::LedHBHEHists(), HcalLedAnalysis::LedHFHists(), HcalLedAnalysis::LedHOHists(), TkTrackingRegionsMargin< float >::left(), fftjetcms::LinInterpolatedTable1D::LinInterpolatedTable1D(), SimpleNavigationSchool::linkOuterGroup(), CalorimetryManager::loadFromEcalBarrel(), CalorimetryManager::loadFromEcalEndcap(), CalorimetryManager::loadFromHcal(), CalorimetryManager::loadFromPreshower(), CTPPSRPAlignmentCorrectionsMethods::loadFromXML(), HGCalGeomParameters::loadGeometryHexagon(), MaterialBudgetVolume::loadLV(), l1t::L1TGlobalUtil::loadPrescalesAndMasks(), egammaTools::localEcalClusterCoordsEB(), egammaTools::localEcalClusterCoordsEE(), L1GtUtils::LogicalExpressionL1Results::logicalExpressionRunUpdate(), ls_cert(), IsoDepositVetoFactory::make(), MuonRPCDetLayerGeometryBuilder::makeBarrelLayers(), MuonRPCDetLayerGeometryBuilder::makeBarrelRods(), HcalFlexiHardcodeGeometryLoader::makeHECells(), align::makeNonOverlappingRunRanges(), DQMImplNet< DQMNet::Object >::makeObject(), ticl::PatternRecognitionbyCA< TILES >::makeTracksters(), pat::Flags::maskToString(), cms::DDFilteredView::match(), JetPartonMatching::matchingTotalMinDist(), TRange< int >::mean(), PixelRecoRange< float >::mean(), muonisolation::Range< float >::mean(), EcalClusterToolsT< noZS >::meanClusterPosition(), trklet::MatchCalculator::mergeMatches(), StripCPEgeometric::WrappedCluster::middle(), TRange< int >::min(), muonisolation::Range< float >::min(), PixelRecoRange< float >::min(), HGCalDDDConstants::modules(), EcalClusterToolsT< noZS >::n5x5(), ShallowClustersProducer::NearDigis::NearDigis(), L1TGlobalPrescalesVetosOnlineProd::newObject(), lhef::LHEReader::next(), pos::PixelCalibConfiguration::nextFECState(), DDValuePair::operator const std::string &(), DDValuePair::operator std::string &(), CastorDbASCIIIO::DetIdLess::operator()(), EcalBasicClusterLocalContCorrection::operator()(), helper::SelectionFirstPointerAdder< StoreContainer >::operator()(), reco::tau::RecoTauPiZeroStripPlugin::operator()(), helper::SelectionFirstRefAdder< StoreContainer >::operator()(), reco::tau::RecoTauPiZeroStripPlugin2::operator()(), reco::tau::RecoTauPiZeroStripPlugin3::operator()(), HLTScalersClient::CountLS_t::operator<(), sistrip::RawToDigiUnpacker::Registry::operator<(), operator<<(), HLTScalersClient::CountLS_t::operator==(), HGCal3DClustering::organizeByLayer(), ClusterClusterMapping::overlap(), optutl::CommandLineParser::parseArguments(), reco::modules::TrackerTrackHitFilter::parseStoN(), SiStripDbParams::partitionNames(), PartitionGenerator::partitions(), L1MuBMAssignmentUnit::PhiAU(), L1MuDTAssignmentUnit::PhiAU(), PhiBorderFinder::PhiBorderFinder(), pos::PixelCalibConfiguration::PixelCalibConfiguration(), HiggsValidation::MonitoredDecays::position(), SiStripRegionCabling::position(), DDHGCalHEAlgo::positionMix(), DDHGCalHEFileAlgo::positionMix(), DDHGCalMixLayer::positionMix(), DDAHcalModuleAlgo::positionSensitive(), DDHGCalTBModule::positionSensitive(), DDHGCalEEAlgo::positionSensitive(), DDHGCalTBModuleX::positionSensitive(), DDHGCalEEFileAlgo::positionSensitive(), DDHGCalSiliconModule::positionSensitive(), DDHGCalHEFileAlgo::positionSensitive(), DDHGCalHEAlgo::positionSensitive(), precomputed_value_sort(), BPHMonitor::Prescale(), PrimaryVertexPlots(), HGCSiliconDetIdToROC::print(), TtFullLeptonicEvent::print(), UCTCTP7RawData::print(), mkfit::LayerOfHits::printBins(), SiStripConfigDb::printFedConnections(), SiStripConfigDb::printFedDescriptions(), sipixelobjects::PixelFEDLink::printForMap(), PrintGeomSummary::printSummary(), HcalSimpleReconstructor::process(), edm::SubProcess::processAsync(), cscdqm::EventProcessor::processCSC(), edm::EventProcessor::processEventAsyncImpl(), TtFullLepKinSolutionProducer::produce(), CandOneToManyDeltaRMatcher::produce(), CTPPSInterpolatedOpticalFunctionsESSource::produce(), CalibratedPhotonProducerRun2T< T >::produce(), CalibratedElectronProducerRun2T< T >::produce(), JetCoreClusterSplitter::produce(), reco::modules::MatcherBase< C1, C2, M >::produce(), reco::modulesNew::Matcher< C1, C2, S, D >::produce(), ConvertObjectMapRecord::produce(), EcalBarrelClusterFastTimer::produce(), cms::DigitizerFP420::produce(), CleanAndMergeProducer::produce(), TrackListMerger::produce(), pat::PATGenericParticleProducer::produce(), MkFitEventOfHitsProducer::produce(), DTFakeT0ESProducer::produce(), L1ExtraParticleMapProd::produce(), TrackMCQuality::produce(), SimPFProducer::produce(), UnifiedSCCollectionProducer::produce(), pat::PATPhotonProducer::produce(), InterestingDetIdFromSuperClusterProducer::produce(), HcalHitReconstructor::produce(), pat::PATTauProducer::produce(), HLTScoutingMuonProducer::produce(), ConvBremSeedProducer::produce(), TrackingRecHitProducer::produce(), HFPreReconstructor::produce(), ClusterTPAssociationProducer::produce(), ClusterSummaryProducer::produce(), BTagProbabilityToDiscriminator::produce(), pat::PATElectronProducer::produce(), NuclearTrackCorrector::produce(), TTStubBuilder< T >::produce(), DeepCoreSeedGenerator::produce(), reco::PFCluster::pruneUsing(), PixelDigiCollection::put(), DigiCollectionFP420::put(), TrackCollectionFP420::put(), RecoCollectionFP420::put(), ClusterCollectionFP420::put(), CompositeLogicalTrajectoryFilter::QF(), RBorderFinder::RBorderFinder(), pos::PixelROCDACSettings::read(), l1t::XmlConfigParser::readContext(), popcon::RPCEMapSourceHandler::readEMap1(), popcon::L1RPCHwConfigSourceHandler::readHwConfig1(), PixelToFEDAssociateFromAscii::readRange(), FedRawDataInputSource::readWorker(), RealQuadEquation::RealQuadEquation(), CSCEfficiency::recHitSegment_Efficiencies(), EcalTBHodoscopeRecInfoAlgo::reconstruct(), HcalSimpleRecAlgo::reconstruct(), SiStripDetVOffBuilder::reduce(), CandidateSeededTrackingRegionsProducer::regions(), PointSeededTrackingRegionsProducer::regions(), L1MuonSeededTrackingRegionsProducer::regions(), DDName::registerName(), mkfit::ConfigJsonPatcher::replace(), pat::eventhypothesis::Looper< T >::reset(), l1t::L1TGlobalUtil::resetDecisionVectors(), l1t::L1TGlobalUtil::resetMaskVectors(), l1t::L1TGlobalUtil::resetPrescaleVectors(), SiStripPsuDetIdMap::retrieveDcuDeviceAddresses(), l1t::L1TGlobalUtil::retrieveL1Event(), l1t::L1TGlobalUtil::retrieveL1Setup(), EcalClusterToolsT< noZS >::roundnessBarrelSuperClusters(), EcalClusterToolsT< noZS >::roundnessBarrelSuperClustersUserExtended(), reco::modules::TrackerTrackHitFilter::Rule::Rule(), PixelCPEClusterRepair::Rule::Rule(), FP420TrackMain::run(), FP420RecoMain::run(), FP420ClusterMain::run(), FP420DigiMain::run(), L1MuDTERS::run(), L1MuBMERS::run(), EcalEBTrigPrimTestAlgo::run(), DTTSS::run(), DTTSM::run(), DTSC::run(), EcalTrigPrimFunctionalAlgo::run_part2(), ecaldqm::RecoSummaryTask::runOnBasicClusters(), ContentsXRange::runTest(), ContentsYRange::runTest(), DeadChannel::runTest(), NoisyChannel::runTest(), DTTSPhi::runTSPhi(), JanAlignmentAlgorithm::saveDiagnostics(), EcalClusterToolsT< noZS >::scLocalCovariances(), EcalSelectiveReadoutValidation::selectFedsForLog(), mkfit::MkFinder::selectHitIndices(), MuScleFit::selectMuons(), cond::payloadInspector::serialize(), cond::payloadInspector::serializeAnnotations(), pos::PixelPortCardConfig::setAOHGain(), egHLT::ComCodes::setCode(), egHLT::TrigCodes::setCode(), pos::PixelPortCardConfig::setDataBaseAOHGain(), HcalFEDList::setListOfFEDs(), DTTracoTrigData::setPV(), DTTracoChip::setPV(), ConvBremSeedProducer::sharedHits(), FWECALCaloDataDetailViewBuilder::showSuperCluster(), FWECALDetailViewBuilder::showSuperCluster(), StripCPEgeometric::WrappedCluster::sign(), SingleTopTChannelLeptonDQM::SingleTopTChannelLeptonDQM(), SingleTopTChannelLeptonDQM_miniAOD::SingleTopTChannelLeptonDQM_miniAOD(), SiStripGainsCalibTreeWorker::SiStripGainsCalibTreeWorker(), SiStripGainsPCLWorker::SiStripGainsPCLWorker(), TrackTransformerForCosmicMuons::SlopeSum(), StripCPEgeometric::WrappedCluster::smallerEdgeStrip(), JanAlignmentAlgorithm::solve(), muonisolation::Range< float >::sort(), PixelRecoRange< float >::sort(), TRange< int >::sort(), HFPreReconstructor::sortDataByPmt(), split(), trklet::MatchEngineUnit::step(), SiPixelStatusHarvester::stepIOV(), AlCaHcalIsotrkProducer::storeEnergy(), HGCPassive::storeInfo(), hgcal::EGammaPCAHelper::storeRecHits(), StringBasedNTupler::StringBasedNTupler(), SuepShower::SuepShower(), TrackTransformerForCosmicMuons::SumDy(), MagGeoBuilderFromDDD::summary(), StripCPEgeometric::WrappedCluster::sumQ(), PixelRecoRange< float >::swap(), muon::tevOptimized(), HcalSimParameters::timeSmearRMS(), TopDiLeptonOfflineDQM::TopDiLeptonOfflineDQM(), TopSingleLeptonDQM::TopSingleLeptonDQM(), TopSingleLeptonDQM_miniAOD::TopSingleLeptonDQM_miniAOD(), fwlite::MultiChainEvent::toSec(), CmsMTDStringToEnum::type(), EcalHitMaker::unbalancedDirection(), FastFedCablingHistosUsingDb::update(), SimG4FluxProducer::update(), HGCPassive::update(), SiTrackerMultiRecHitUpdator::update(), o2o_db_cfgmap.DbManagerDAQ::update_hashmap(), StraightTrackAlignment::updateDiagnosticHistograms(), HBHENegativeEFilter::validate(), FWConfiguration::valueForKey(), viewNameFrom(), HGCalDDDConstants::waferType(), HGCalDDDConstants::waferTypeRotation(), HGCalWaferMask::waferXY(), popcon::PopCon::write(), MuonAssociatorByHitsHelper::write_matched_simtracks(), pos::PixelCalibConfiguration::writeASCII(), trklet::TrackletConfigBuilder::writeASMemories(), trklet::TrackletConfigBuilder::writeILMemories(), VirtualJetProducer::writeJets(), edm::SubProcess::writeLumiAsync(), edm::Schedule::writeLumiAsync(), edm::EventProcessor::writeLumiAsync(), edm::SubProcess::writeProcessBlockAsync(), edm::Schedule::writeProcessBlockAsync(), edm::EventProcessor::writeProcessBlockAsync(), edm::SubProcess::writeRunAsync(), edm::Schedule::writeRunAsync(), edm::EventProcessor::writeRunAsync(), and trklet::TrackletConfigBuilder::writeSPMemories().

HitContainer const* __restrict__ TkSoA const* __restrict__ Quality const* __restrict__ CAHitNtupletGeneratorKernelsGPU::HitToTuple const* __restrict__ int32_t firstPrint

Definition at line 863 of file CAHitNtupletGeneratorKernelsImpl.h.

auto const & foundNtuplets = *ptuples
auto const good = pixelTrack::Quality::strict
auto const & hh = *hhp
TrackingRecHit2DSOAView const* __restrict__ hhp

Definition at line 546 of file CAHitNtupletGeneratorKernelsImpl.h.

TrackingRecHit2DSOAView const* __restrict__ HitContainer* __restrict__ hitDetIndices
Initial value:
{
int first = blockDim.x * blockIdx.x + threadIdx.x
const dim3 threadIdx
Definition: cudaCompat.h:29
const dim3 blockDim
Definition: cudaCompat.h:30
const dim3 blockIdx
Definition: cudaCompat.h:32

Definition at line 547 of file CAHitNtupletGeneratorKernelsImpl.h.

auto & hitToTuple
Initial value:
{
int first = blockDim.x * blockIdx.x + threadIdx.x
const dim3 threadIdx
Definition: cudaCompat.h:29
const dim3 blockDim
Definition: cudaCompat.h:30
const dim3 blockIdx
Definition: cudaCompat.h:32

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const* __restrict__ TkSoA const* __restrict__ Quality const* __restrict__ CAHitNtupletGeneratorKernelsGPU::HitToTuple const* __restrict__ int32_t int32_t int iev

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

int l1end = hh.hitsLayerStart()[1]

Definition at line 723 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const* __restrict__ TkSoA const* __restrict__ Quality const* __restrict__ CAHitNtupletGeneratorKernelsGPU::HitToTuple const* __restrict__ int32_t int32_t lastPrint

Definition at line 863 of file CAHitNtupletGeneratorKernelsImpl.h.

auto const longTqual = pixelTrack::Quality::highPurity

Definition at line 717 of file CAHitNtupletGeneratorKernelsImpl.h.

constexpr auto loose = pixelTrack::Quality::loose

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

uint32_t const * nCells

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

Referenced by DTMeantimerPatternReco::addHits(), SeedFinder::addHitSelector(), CAHitNtupletGeneratorKernels< TTraits >::allocateOnGPU(), SiPixelPhase1MonitorTrackSoA::analyze(), DTOccupancyEfficiency::analyze(), DTSegmentsTask::analyze(), DTSegmentAnalysisTask::analyze(), BtlLocalRecoValidation::analyze(), PhotonValidator::analyze(), HcalHitValidation::analyzeLayer(), RecoMuonValidator::bookHistograms(), RecoMuonValidator::MuonME::bookHistos(), FWTracksterLayersProxyBuilder::build(), PtAssignmentEngine2017::calculate_address(), PtAssignmentEngine2017::calculate_pt_xml(), RPCSeedOverlapper::CheckOverlap(), mkfit::Event::clean_cms_seedtracks(), mkfit::StdSeq::clean_cms_seedtracks_iter(), DAFTrackProducerAlgorithm::collectHits(), OMTFResult::empty(), SimG4HcalValidation::fetchHits(), L2TauNNProducer::fillPatatracks(), TrackingNtuple::fillSeeds(), CastorShowerLibraryMaker::FillShowerEvent(), TrackingNtuple::fillTracks(), gpuVertexFinder::for(), for(), DTChamberEfficiencyTask::getBestSegment(), MuonSeedCleaner::GroupSeeds(), MillePedeMonitor::init(), DTChamberEfficiencyTask::isGoodSegment(), PFDisplacedVertexHelper::isTrackSelected(), main(), SCEnergyCorrectorDRN::makeInput(), BTLUncalibRecHitAlgo::makeRecHit(), LowPtConversion::match(), CovarianceParameterization::meanValue(), SETSeedFinder::pre_prune(), CtfSpecialSeedGenerator::preliminaryCheck(), PixelTemplateSmearerBase::process(), SeedGeneratorFromProtoTracksEDProducer::produce(), QualityFilter::produce(), TrackListCombiner::produce(), PixelTracksProducer::produce(), TSGFromL1Muon::produce(), HiBadParticleCleaner::produce(), PixelTrackSoAFromCUDA::produce(), SeedProducerFromSoA::produce(), PixelTrackProducerFromSoA::produce(), PixelVertexProducerCUDA::produceOnCPU(), CSCSegAlgoShowering::pruneFromResidual(), CSCSegAlgoDF::pruneFromResidual(), mkfit::StdSeq::qfilter_pixelLessBkwd(), mkfit::StdSeq::qfilter_pixelLessFwd(), SeedFromGenericPairOrTriplet::qualityFilter(), PFAlgo::recoTracksNotHCAL(), PixelTrackReconstruction::run(), ConversionTrackPairFinder::run(), CSCSegAlgoDF::run(), MuonSeedCleaner::SeedCandidates(), SeedFromProtoTrack::SeedFromProtoTrack(), DTCombinatorialPatternReco4D::segmentSpecialZed(), L2TauNNProducer::selectGoodTracksAndVertices(), CSCBaseElectronicsSim::simulate(), storeTracks(), mkfit::Track::Track(), and hi::EPCuts::trackQuality_Pixel().

auto nhits = hh.nHits()

Definition at line 555 of file CAHitNtupletGeneratorKernelsImpl.h.

Quality *__restrict__ uint16_t nmin
auto ntracks = apc->get().m
HitContainer const *__restrict__ TkSoA const *__restrict__ Quality const *__restrict__ CAHitNtupletGeneratorKernelsGPU::HitToTuple const *__restrict__ phitToTuple
Initial value:

Definition at line 591 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const *__restrict__ TkSoA const *__restrict__ ptracks

Definition at line 141 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const *__restrict__ ptuples

Definition at line 589 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const* __restrict__ TkSoA const* __restrict__ Quality const* __restrict__ quality
Initial value:

Definition at line 124 of file CAHitNtupletGeneratorKernelsImpl.h.

auto const reject = dupPassThrough ? loose : dup
auto const & tracks = *ptracks

cannot be loose

Definition at line 148 of file CAHitNtupletGeneratorKernelsImpl.h.

Referenced by GPUCACell::__attribute__(), RecoTrackAccumulator::accumulateEvent(), CaloTruthAccumulator::accumulateEvent(), edm::DataMixingGeneralTrackWorker::addGeneralTrackPileups(), edm::DataMixingGeneralTrackWorker::addGeneralTrackSignals(), PFMuonAlgo::addMissingMuons(), PreMixingTrackingParticleWorker::addSignals(), l1tVertexFinder::VertexFinder::agglomerativeHierarchicalClustering(), SiStripGainCosmicCalculator::algoAnalyze(), ElectronGeneralAnalyzer::analyze(), ElectronTagProbeAnalyzer::analyze(), TrackTypeMonitor::analyze(), CSCTFAnalyzer::analyze(), PixelTrackDumpCUDA::analyze(), PatTrackAnalyzer::analyze(), TrackingRecoMaterialAnalyser::analyze(), ElasticPlotDQMSource::analyze(), EwkMuLumiMonitorDQM::analyze(), EfficiencyAnalyzer::analyze(), AlcaBeamMonitor::analyze(), CTPPSTrackDistributionPlotter::analyze(), ElectronAnalyzer::analyze(), PPSAlignmentWorker::analyze(), TotemRPDQMSource::analyze(), CTPPSProtonReconstructionEfficiencyEstimatorMC::analyze(), PhotonsWithConversionsAnalyzer::analyze(), CSCTFanalyzer::analyze(), BeamMonitor::analyze(), SiStripLAProfileBooker::analyze(), StandaloneTrackMonitor::analyze(), LhcTrackAnalyzer::analyze(), cms::ProducerAnalyzer::analyze(), SiPixelTrackResidualSource::analyze(), CaloParticleDebugger::analyze(), DiamondSampicDQMSource::analyze(), TestResolution::analyze(), TkConvValidator::analyze(), DTChamberEfficiency::analyze(), QcdUeDQM::analyze(), PhotonValidator::analyze(), CosmicSplitterValidation::analyze(), L1TdeCSCTF::analyze(), TestCorrection::analyze(), L1TCSCTF::analyze(), SiPixelErrorEstimation::analyze(), EopTreeWriter::analyze(), DuplicateRecHits::analyze(), SiPixelTrackingRecHitsValid::analyze(), TrackCount::analyze(), SiPixelPhase1Analyzer::analyze(), CTPPSDiamondDQMSource::analyze(), EcalCosmicsHists::analyze(), PrimaryVertexValidation::analyze(), CosmicRateAnalyzer::analyze(), SplitVertexResolution::analyze(), PhotonAnalyzer::analyze(), PFAnalysis::analyze(), SiStripTrackingRecHitsValid::analyze(), PackedCandidateTrackValidator::analyze(), TrackingNtuple::analyze(), CaloSimHitAnalysis::analyzePassiveHits(), TrackerRemapper::analyzeRechits(), PixelClusterShapeExtractor::analyzeRecTracks(), GlobalTrackerMuonAlignment::analyzeTrackTrack(), GlobalTrackerMuonAlignment::analyzeTrackTrajectory(), FWTrackProxyBuilderFullFramework::build(), tthelpers::buildTT(), FWElectronProxyBuilder::buildViewType(), PhotonIsolationCalculator::calculateTrackIso(), EgammaL1TkIsolation::calIsol(), BestTrackSelection::cancel_one_bx(), AlignmentGlobalTrackSelector::checkJetCount(), PFMuonAlgo::cleanMismeasured(), TrackCleaner::cleanTracks(), GapClusterizerInZ::clusterize(), TrackMVAClassifier< MVA, EventCache >::computeMVA(), TemplatedJetProbabilityComputer< Container, Base >::discriminator(), TemplatedJetBProbabilityComputer< Container, Base >::discriminator(), SiStripGainsPCLWorker::dqmAnalyze(), KalmanSmoothedVertexChi2Estimator< N >::estimate(), GsfVertexTrackCompatibilityEstimator::estimate(), KalmanVertexTrackCompatibilityEstimator< N >::estimate(), BoostedDoubleSVProducer::etaRelToTauAxis(), MuonCosmicCompatibilityFiller::eventActivity(), JetPlusTrackCorrector::excludeJta(), tmtt::MuxHToutputs::exec(), CombinedSVComputer::fillCommonVariables(), SiStripMonitorTrack::fillControlViewHistos(), pat::GenericParticle::fillInFrom(), Phase2ITValidateTrackingRecHit::fillITHistos(), Phase2OTValidateTrackingRecHit::fillOTHistos(), TrackerValidationVariables::fillTrackQuantities(), TrackingFailureFilter::filter(), MuScleFitFilter::filter(), FilterScrapingPixelProbability::filter(), JetVertexChecker::filter(), CSCDigiValidator::filter(), CSCSkim::filter(), LeptonSkimming::filter(), reco::tau::PFRecoTauChargedHadronFromGenericTrackPlugin< TrackClass >::filterTrack(), shallow::findTrackIndex(), AdaptiveVertexFitter::fit(), SequentialVertexFitter< N >::fit(), EcalTBHodoscopeRecInfoAlgo::fitLine(), CrossingPtBasedLinearizationPointFinder::getBestTracks(), getBestVertex(), getBestVertex_withError(), PFMuonAlgo::getMinMaxMET2(), l1tVertexFinder::getPrimaryVertex(), l1t::MicroGMTCancelOutUnit::getTrackAddrCancelBits(), reco::JetTrackMatch< JetC >::getTracks(), HLTSingleVertexPixelTrackFilter::hltFilter(), HLTDisplacedEgammaFilter::hltFilter(), HLTPixlMBFilt::hltFilter(), HLTPixlMBForAlignmentFilter::hltFilter(), GeneralTracksImporter::importToBlock(), mkfit::MkFitter::inputTracksAndHitIdx(), tmtt::HTrphi::killTracksBusySec(), L1TMuonProducer::L1TMuonProducer(), QualityCutsAnalyzer::LoopOverJetTracksAssociation(), CAHitNtupletGeneratorOnGPU::makeTuples(), CAHitNtupletGeneratorOnGPU::makeTuplesAsync(), L1EmulBias::ModifyCollection(), Cluster1DMerger< T >::operator()(), pixeltemp::Cluster1DMerger< T >::operator()(), KalmanTrackToTrackCovCalculator< N >::operator()(), reco::tau::PFRecoTauChargedHadronFromGenericTrackPlugin< TrackClass >::operator()(), PromptTrackCountingComputer::orderedSignificances(), TemplatedTrackCountingComputer< Container, Base >::orderedSignificances(), reco::TrackJet::print(), AngleCalculation::process(), PrimitiveMatching::process(), __class__Worker::process(), PixelVertexProducerMedian::produce(), IsolationProducerForTracks::produce(), TrackExtraRekeyer::produce(), L3MuonCleaner::produce(), JetChargeProducer::produce(), ShallowTracksProducer::produce(), HIPixelMedianVtxProducer::produce(), JetSignalVertexCompatibility::produce(), CSCTFCandidateProducer::produce(), SiStripOnTrackClusterTableProducerBase::produce(), ShallowSimTracksProducer::produce(), IsolatedTrackCleaner::produce(), PixelTracksProducer::produce(), MuIsoDepositProducer::produce(), LowPtGSFToTrackLinker::produce(), CaloMuonMerger::produce(), FastTrackerRecHitMaskProducer::produce(), GenTrackMatcher::produce(), MCTrackMatcher::produce(), TrackFromPVSelector::produce(), l1tpf::PFTrackProducerFromL1Tracks::produce(), TSGFromL1Muon::produce(), pat::PATLostTracks::produce(), LowPtGSFToPackedCandidateLinker::produce(), TrackstersMergeProducer::produce(), cms::JetVertexAssociation::produce(), MTDTrackQualityMVAProducer::produce(), TOFPIDProducer::produce(), ShallowTrackClustersProducer::produce(), TauJetSelectorForHLTTrackSeeding::produce(), CTPPSDiamondLocalTrackFitter::produce(), CSCTFPacker::produce(), L3TkMuonProducer::produce(), CSCTFUnpacker::produce(), TrackSelectorByRegion::produce(), L2MuonCandidateProducer::produce(), CosmicsMuonIdProducer::produce(), MkFitProducer::produce(), TrackMVAClassifierBase::produce(), HGCalTrackCollectionProducer::produce(), HITrackClusterRemover::produce(), PATTracksToPackedCandidates::produce(), L3MuonCandidateProducer::produce(), PixelVertexProducer::produce(), HLTTrackMETProducer::produce(), PackedCandidateGenAssociationProducer::produce(), SiStripFineDelayHit::produce(), CalibrationTrackSelectorFromDetIdList::produce(), TrackingMaterialProducer::produce(), ExtraFromSeeds::produce(), InterestingTrackEcalDetIdProducer::produce(), PixelTrackProducerFromSoA::produce(), MuonErrorMatrixAdjuster::produce(), TracksToTrajectories::produce(), PixelTrackProducer::produce(), TrackFromSeedProducer::produce(), reco::CentralityProducer::produce(), PixelJetPuId::produce(), ShallowGainCalibration::produce(), CSCOverlapsTrackPreparation::produce(), DeepFlavourTagInfoProducer::produce(), L1TPFProducer::produce(), reco::modules::TrackerTrackHitFilter::produce(), reco::modules::CosmicTrackSplitter::produce(), TemplatedVertexArbitrator< InputContainer, VTX >::produce(), SoftLepton::produce(), TemplatedInclusiveVertexFinder< InputContainer, VTX >::produce(), IPProducer< Container, Base, Helper >::produce(), L1FPGATrackProducer::produce(), PixelVertexProducerCUDA::produceOnCPU(), PixelVertexProducerCUDA::produceOnGPU(), MuonCosmicCompatibilityFiller::pvMatches(), BeamFitter::readEvent(), JetPlusTrackCorrector::rebuildJta(), ProtonReconstructionAlgorithm::reconstructFromMultiRP(), MuonTrackingRegionByPtBuilder::regions(), MuonTrackingRegionBuilder::regions(), HITRegionalPixelSeedGenerator::regions(), HIPAlignmentAlgorithm::run(), MillePedeAlignmentAlgorithm::run(), HITrackFilterForPVFinding::select(), CalibrationTrackSelector::select(), AlignmentTwoBodyDecayTrackSelector::select(), AlignmentGlobalTrackSelector::select(), AlignmentTrackSelector::select(), MuScleFitMuonSelector::selectMuons(), CandCommonVertexFitterBase::set(), L1DummyProducer::SimpleDigi(), GsfVertexSmoother::smooth(), PhotonCoreProducer::solveAmbiguity(), storeTracks(), ConeIsolationAlgorithm::tag(), ImpactParameterAlgorithm::tag(), DivisiveClusterizer1D< T >::takeTracks(), pixeltemp::DivisiveClusterizer1D< T >::takeTracks(), CalibrationTrackSelector::theNHighestPtTracks(), AlignmentTrackSelector::theNHighestPtTracks(), MuonTrackAnalyzer::tracksAnalysis(), reco::JetTracksAssociation::tracksP4(), StoreSecondary::update(), CheckSecondary::update(), DAClusterizerInZ_vect::update(), DAClusterizerInZT_vect::update(), AdaptiveVertexFitter::vertex(), ConfigurableTrimmedVertexFinder::vertexCandidates(), MultiVertexFitter::vertices(), and vtxP4().

Quality const* __restrict__ caConstants::TupleMultiplicity* tupleMultiplicity
Initial value:
{
auto first = blockIdx.x * blockDim.x + threadIdx.x
const dim3 threadIdx
Definition: cudaCompat.h:29
const dim3 blockDim
Definition: cudaCompat.h:30
const dim3 blockIdx
Definition: cudaCompat.h:32

Definition at line 43 of file CAHitNtupletGeneratorKernelsImpl.h.

HitContainer const* __restrict__ tuples

Definition at line 625 of file CAHitNtupletGeneratorKernelsImpl.h.

Referenced by HelixFitOnGPU::allocateOnGPU().