1 #include <cuda_runtime.h> 28 #undef PIXVERTEX_DEBUG_PRODUCE 30 template <
typename TrackerTraits>
61 template <
typename TrackerTraits>
63 : onGPU_(conf.getParameter<
bool>(
"onGPU")),
64 gpuAlgo_(conf.getParameter<
bool>(
"oneKernel"),
65 conf.getParameter<
bool>(
"useDensity"),
66 conf.getParameter<
bool>(
"useDBSCAN"),
67 conf.getParameter<
bool>(
"useIterative"),
68 conf.getParameter<
int>(
"minT"),
69 conf.getParameter<double>(
"eps"),
70 conf.getParameter<double>(
"errmax"),
71 conf.getParameter<double>(
"chi2max")),
72 ptMin_(conf.getParameter<double>(
"PtMin")),
73 ptMax_(conf.getParameter<double>(
"PtMax"))
84 template <
typename TrackerTraits>
90 desc.add<
bool>(
"onGPU",
true);
91 desc.add<
bool>(
"oneKernel",
true);
92 desc.add<
bool>(
"useDensity",
true);
93 desc.add<
bool>(
"useDBSCAN",
false);
94 desc.add<
bool>(
"useIterative",
false);
96 desc.add<
int>(
"minT", 2);
97 desc.add<
double>(
"eps", 0.07);
98 desc.add<
double>(
"errmax", 0.01);
99 desc.add<
double>(
"chi2max", 9.);
101 desc.add<
double>(
"PtMin", 0.5);
102 desc.add<
double>(
"PtMax", 75.);
108 template <
typename TrackerTraits>
113 auto hTracks =
iEvent.getHandle(tokenGPUTrack_);
116 auto&
tracks = ctx.get(*hTracks);
118 ctx.emplace(
iEvent, tokenGPUVertex_, gpuAlgo_.makeAsync(ctx.stream(),
tracks.view(), ptMin_, ptMax_));
121 template <
typename TrackerTraits>
127 #ifdef PIXVERTEX_DEBUG_PRODUCE 128 auto const& tsoa = *
tracks;
133 for (int32_t it = 0; it <
maxTracks; ++it) {
140 std::cout <<
"found " <<
nt <<
" tracks in cpu SoA for Vertexing at " <<
tracks << std::endl;
141 #endif // PIXVERTEX_DEBUG_PRODUCE 143 iEvent.emplace(tokenCPUVertex_, gpuAlgo_.make(
tracks.view(), ptMin_, ptMax_));
146 template <
typename TrackerTraits>
151 produceOnGPU(streamID,
iEvent, iSetup);
153 produceOnCPU(streamID,
iEvent, iSetup);
void addWithDefaultLabel(ParameterSetDescription const &psetDescription)
T getParameter(std::string const &) const
edm::EDGetTokenT< TracksSoAHost > tokenCPUTrack_
void produceOnGPU(edm::StreamID streamID, edm::Event &iEvent, const edm::EventSetup &iSetup) const
~PixelVertexProducerCUDAT() override=default
edm::EDPutTokenT< cms::cuda::Product< ZVertexSoADevice > > tokenGPUVertex_
EDGetTokenT< ProductType > consumes(edm::InputTag const &tag)
edm::EDPutTokenT< ZVertexSoAHost > tokenCPUVertex_
#define DEFINE_FWK_MODULE(type)
void produceOnCPU(edm::StreamID streamID, edm::Event &iEvent, const edm::EventSetup &iSetup) const
edm::EDGetTokenT< cms::cuda::Product< TracksSoADevice > > tokenGPUTrack_
static constexpr __host__ __device__ int nHits(const TrackSoAConstView &tracks, int i)
static void fillDescriptions(edm::ConfigurationDescriptions &descriptions)
PixelVertexProducerCUDAT(const edm::ParameterSet &iConfig)
void produce(edm::StreamID streamID, edm::Event &iEvent, const edm::EventSetup &iSetup) const override
TupleMultiplicity< TrackerTraits > const *__restrict__ uint32_t nHits