Skip to content

Commit 4b54376

Browse files
f3schf3sch
committed
ITS: separate into cell and track seed class
Signed-off-by: Felix Schlepper <felix.schlepper@cern.ch>
1 parent 238c87c commit 4b54376

10 files changed

Lines changed: 174 additions & 108 deletions

File tree

‎Detectors/ITSMFT/ITS/tracking/GPU/ITStrackingGPU/TimeFrameGPU.h‎

Lines changed: 4 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -32,6 +32,7 @@ class TimeFrameGPU final : public TimeFrame<NLayers>
3232
using typename TimeFrame<NLayers>::ROFOverlapTableN;
3333
using typename TimeFrame<NLayers>::ROFVertexLookupTableN;
3434
using typename TimeFrame<NLayers>::ROFMaskTableN;
35+
using typename TimeFrame<NLayers>::TrackSeedN;
3536

3637
public:
3738
TimeFrameGPU() = default;
@@ -72,7 +73,7 @@ class TimeFrameGPU final : public TimeFrame<NLayers>
7273
void loadCellsLUTDevice();
7374
void loadTrackSeedsDevice();
7475
void loadTrackSeedsChi2Device();
75-
void loadTrackSeedsDevice(bounded_vector<CellSeedN>&);
76+
void loadTrackSeedsDevice(bounded_vector<TrackSeedN>&);
7677
void createTrackletsBuffers(const int);
7778
void createTrackletsBuffersArray(const int);
7879
void createCellsBuffers(const int);
@@ -137,7 +138,7 @@ class TimeFrameGPU final : public TimeFrame<NLayers>
137138
int** getDeviceArrayCellsLUT() const { return mCellsLUTDeviceArray; }
138139
int** getDeviceArrayNeighboursCellLUT() const { return mNeighboursCellLUTDeviceArray; }
139140
CellSeedN** getDeviceArrayCells() { return mCellsDeviceArray; }
140-
CellSeedN* getDeviceTrackSeeds() { return mTrackSeedsDevice; }
141+
TrackSeedN* getDeviceTrackSeeds() { return mTrackSeedsDevice; }
141142
int* getDeviceTrackSeedsLUT() { return mTrackSeedsLUTDevice; }
142143
auto getNTrackSeeds() const { return mNTracks; }
143144
o2::track::TrackParCovF** getDeviceArrayTrackSeeds() { return mCellSeedsDeviceArray; }
@@ -206,7 +207,7 @@ class TimeFrameGPU final : public TimeFrame<NLayers>
206207
std::array<CellSeedN*, NLayers - 2> mCellsDevice;
207208
CellSeedN** mCellsDeviceArray;
208209
std::array<int*, NLayers - 3> mNeighboursIndexTablesDevice;
209-
CellSeedN* mTrackSeedsDevice{nullptr};
210+
TrackSeedN* mTrackSeedsDevice{nullptr};
210211
int* mTrackSeedsLUTDevice{nullptr};
211212
unsigned int mNTracks{0};
212213
std::array<o2::track::TrackParCovF*, NLayers - 2> mCellSeedsDevice;

‎Detectors/ITSMFT/ITS/tracking/GPU/ITStrackingGPU/TrackingKernels.h‎

Lines changed: 11 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -24,8 +24,9 @@
2424

2525
namespace o2::its
2626
{
27-
template <int>
2827
class CellSeed;
28+
template <int>
29+
class TrackSeed;
2930
class TrackingFrameInfo;
3031
class Tracklet;
3132
template <int>
@@ -101,7 +102,7 @@ void countCellsHandler(const Cluster** sortedClusters,
101102
int** trackletsLUT,
102103
const int nTracklets,
103104
const int layer,
104-
CellSeed<NLayers>* cells,
105+
CellSeed* cells,
105106
int** cellsLUTsDeviceArray,
106107
int* cellsLUTsHost,
107108
const float bz,
@@ -119,7 +120,7 @@ void computeCellsHandler(const Cluster** sortedClusters,
119120
int** trackletsLUT,
120121
const int nTracklets,
121122
const int layer,
122-
CellSeed<NLayers>* cells,
123+
CellSeed* cells,
123124
int** cellsLUTsDeviceArray,
124125
int* cellsLUTsHost,
125126
const float bz,
@@ -129,7 +130,7 @@ void computeCellsHandler(const Cluster** sortedClusters,
129130
gpu::Streams& streams);
130131

131132
template <int NLayers>
132-
void countCellNeighboursHandler(CellSeed<NLayers>** cellsLayersDevice,
133+
void countCellNeighboursHandler(CellSeed** cellsLayersDevice,
133134
int* neighboursLUTs,
134135
int** cellsLUTs,
135136
gpuPair<int, int>* cellNeighbours,
@@ -145,7 +146,7 @@ void countCellNeighboursHandler(CellSeed<NLayers>** cellsLayersDevice,
145146
gpu::Stream& stream);
146147

147148
template <int NLayers>
148-
void computeCellNeighboursHandler(CellSeed<NLayers>** cellsLayersDevice,
149+
void computeCellNeighboursHandler(CellSeed** cellsLayersDevice,
149150
int* neighboursLUTs,
150151
int** cellsLUTs,
151152
gpuPair<int, int>* cellNeighbours,
@@ -168,14 +169,14 @@ int filterCellNeighboursHandler(gpuPair<int, int>*,
168169
template <int NLayers = 7>
169170
void processNeighboursHandler(const int startLayer,
170171
const int startLevel,
171-
CellSeed<NLayers>** allCellSeeds,
172-
CellSeed<NLayers>* currentCellSeeds,
172+
CellSeed** allCellSeeds,
173+
CellSeed* currentCellSeeds,
173174
std::array<int, NLayers - 2>& nCells,
174175
const unsigned char** usedClusters,
175176
std::array<int*, NLayers - 2>& neighbours,
176177
gsl::span<int*> neighboursDeviceLUTs,
177178
const TrackingFrameInfo** foundTrackingFrameInfo,
178-
bounded_vector<CellSeed<NLayers>>& seedsHost,
179+
bounded_vector<TrackSeed<NLayers>>& seedsHost,
179180
const float bz,
180181
const float MaxChi2ClusterAttachment,
181182
const float maxChi2NDF,
@@ -184,7 +185,7 @@ void processNeighboursHandler(const int startLayer,
184185
o2::its::ExternalAllocator* alloc);
185186

186187
template <int NLayers = 7>
187-
void countTrackSeedHandler(CellSeed<NLayers>* trackSeeds,
188+
void countTrackSeedHandler(TrackSeed<NLayers>* trackSeeds,
188189
const TrackingFrameInfo** foundTrackingFrameInfo,
189190
const Cluster** unsortedClusters,
190191
int* seedLUT,
@@ -203,7 +204,7 @@ void countTrackSeedHandler(CellSeed<NLayers>* trackSeeds,
203204
o2::its::ExternalAllocator* alloc);
204205

205206
template <int NLayers = 7>
206-
void computeTrackSeedHandler(CellSeed<NLayers>* trackSeeds,
207+
void computeTrackSeedHandler(TrackSeed<NLayers>* trackSeeds,
207208
const TrackingFrameInfo** foundTrackingFrameInfo,
208209
const Cluster** unsortedClusters,
209210
o2::its::TrackITSExt* tracks,

‎Detectors/ITSMFT/ITS/tracking/GPU/cuda/TimeFrameGPU.cu‎

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -495,12 +495,12 @@ void TimeFrameGPU<NLayers>::loadCellsLUTDevice()
495495
}
496496

497497
template <int NLayers>
498-
void TimeFrameGPU<NLayers>::loadTrackSeedsDevice(bounded_vector<CellSeedN>& seeds)
498+
void TimeFrameGPU<NLayers>::loadTrackSeedsDevice(bounded_vector<TrackSeedN>& seeds)
499499
{
500500
GPUTimer timer("loading track seeds");
501-
GPULog("gpu-transfer: loading {} track seeds, for {:.2f} MB.", seeds.size(), seeds.size() * sizeof(CellSeedN) / constants::MB);
502-
allocMem(reinterpret_cast<void**>(&mTrackSeedsDevice), seeds.size() * sizeof(CellSeedN), this->hasFrameworkAllocator(), (o2::gpu::GPUMemoryResource::MEMORY_GPU | o2::gpu::GPUMemoryResource::MEMORY_STACK));
503-
GPUChkErrS(cudaMemcpy(mTrackSeedsDevice, seeds.data(), seeds.size() * sizeof(CellSeedN), cudaMemcpyHostToDevice));
501+
GPULog("gpu-transfer: loading {} track seeds, for {:.2f} MB.", seeds.size(), seeds.size() * sizeof(TrackSeedN) / constants::MB);
502+
allocMem(reinterpret_cast<void**>(&mTrackSeedsDevice), seeds.size() * sizeof(TrackSeedN), this->hasFrameworkAllocator(), (o2::gpu::GPUMemoryResource::MEMORY_GPU | o2::gpu::GPUMemoryResource::MEMORY_STACK));
503+
GPUChkErrS(cudaMemcpy(mTrackSeedsDevice, seeds.data(), seeds.size() * sizeof(TrackSeedN), cudaMemcpyHostToDevice));
504504
GPULog("gpu-transfer: creating {} track seeds LUT, for {:.2f} MB.", seeds.size() + 1, (seeds.size() + 1) * sizeof(int) / constants::MB);
505505
allocMem(reinterpret_cast<void**>(&mTrackSeedsLUTDevice), (seeds.size() + 1) * sizeof(int), this->hasFrameworkAllocator(), (o2::gpu::GPUMemoryResource::MEMORY_GPU | o2::gpu::GPUMemoryResource::MEMORY_STACK));
506506
GPUChkErrS(cudaMemset(mTrackSeedsLUTDevice, 0, (seeds.size() + 1) * sizeof(int)));

‎Detectors/ITSMFT/ITS/tracking/GPU/cuda/TrackerTraitsGPU.cxx‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -265,7 +265,7 @@ void TrackerTraitsGPU<NLayers>::findRoads(const int iteration)
265265
{
266266
for (int startLevel{this->mTrkParams[iteration].CellsPerRoad()}; startLevel >= this->mTrkParams[iteration].CellMinimumLevel(); --startLevel) {
267267
const int minimumLayer{startLevel - 1};
268-
bounded_vector<CellSeed<NLayers>> trackSeeds(this->getMemoryPool().get());
268+
bounded_vector<TrackSeed<NLayers>> trackSeeds(this->getMemoryPool().get());
269269
for (int startLayer{this->mTrkParams[iteration].CellsPerRoad() - 1}; startLayer >= minimumLayer; --startLayer) {
270270
if ((this->mTrkParams[iteration].StartLayerMask & (1 << (startLayer + 2))) == 0) {
271271
continue;

0 commit comments

Comments
 (0)