Project
Loading...
Searching...
No Matches
GPUDisplayDraw.cxx
Go to the documentation of this file.
1// Copyright 2019-2020 CERN and copyright holders of ALICE O2.
2// See https://alice-o2.web.cern.ch/copyright for details of the copyright holders.
3// All rights not expressly granted are reserved.
4//
5// This software is distributed under the terms of the GNU General Public
6// License v3 (GPL Version 3), copied verbatim in the file "COPYING".
7//
8// In applying this license CERN does not waive the privileges and immunities
9// granted to it by virtue of its status as an Intergovernmental Organization
10// or submit itself to any jurisdiction.
11
14
15#ifndef GPUCA_NO_ROOT
16#include "Rtypes.h" // Include ROOT header first, to use ROOT and disable replacements
17#endif
18
19#include "GPUDisplay.h"
20#include "GPUTRDGeometry.h"
21#include "GPUO2DataTypes.h"
22#include "GPUTRDTracker.h"
23#include "GPUTRDTrackletWord.h"
24#include "GPUQA.h"
25#include "GPUTPCClusterData.h"
26#include "GPUTPCConvertImpl.h"
27#include "GPUTPCGMPropagator.h"
28#include "GPUTPCMCInfo.h"
29#include "GPUParam.inc"
30#include "GPUCommonMath.h"
31#include "GPUChainTracking.h"
32
33#include <type_traits>
34
39
40#include <oneapi/tbb.h>
41
42using namespace o2::gpu;
43
44#define GET_CID(sector, i) (mParam->par.earlyTpcTransform ? mIOPtrs->clusterData[sector][i].id : (mIOPtrs->clustersNative->clusterOffset[sector][0] + i))
45
46const GPUTRDGeometry* GPUDisplay::trdGeometry() { return (GPUTRDGeometry*)mCalib->trdGeometry; }
47const GPUTPCTracker& GPUDisplay::sectorTracker(int32_t iSector) { return mChain->GetProcessors()->tpcTrackers[iSector]; }
48
49inline void GPUDisplay::insertVertexList(std::pair<vecpod<int32_t>*, vecpod<uint32_t>*>& vBuf, size_t first, size_t last)
50{
51 if (first == last) {
52 return;
53 }
54 vBuf.first->emplace_back(first);
55 vBuf.second->emplace_back(last - first);
56}
57inline void GPUDisplay::insertVertexList(int32_t iSector, size_t first, size_t last)
58{
59 std::pair<vecpod<int32_t>*, vecpod<uint32_t>*> vBuf(mVertexBufferStart + iSector, mVertexBufferCount + iSector);
60 insertVertexList(vBuf, first, last);
61}
62
63inline void GPUDisplay::drawPointLinestrip(int32_t iSector, int32_t cid, int32_t id, int32_t id_limit)
64{
65 mVertexBuffer[iSector].emplace_back(mGlobalPos[cid].x, mGlobalPos[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPos[cid].z);
66 float curVal;
67 while ((curVal = mGlobalPos[cid].w) < id_limit) {
68 if (CAMath::AtomicCAS(&mGlobalPos[cid].w, curVal, (float)id)) {
69 break;
70 }
71 curVal = mGlobalPos[cid].w;
72 }
73}
74
75GPUDisplay::vboList GPUDisplay::DrawSpacePointsTRD(int32_t iSector, int32_t select, int32_t iCol)
76{
77 size_t startCount = mVertexBufferStart[iSector].size();
78 size_t startCountInner = mVertexBuffer[iSector].size();
79
80 if (iCol == 0) {
81 for (uint32_t i = 0; i < mIOPtrs->nTRDTracklets; i++) {
82 int32_t iSec = trdGeometry()->GetSector(mIOPtrs->trdTracklets[i].GetDetector());
83 bool draw = iSector == iSec && mGlobalPosTRD[i].w == select;
84 if (draw) {
85 mVertexBuffer[iSector].emplace_back(mGlobalPosTRD[i].x, mGlobalPosTRD[i].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTRD[i].z);
86 mVertexBuffer[iSector].emplace_back(mGlobalPosTRD2[i].x, mGlobalPosTRD2[i].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTRD2[i].z);
87 }
88 }
89 }
90
91 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
92 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
93}
94
95GPUDisplay::vboList GPUDisplay::DrawSpacePointsTOF(int32_t iSector, int32_t select, int32_t iCol)
96{
97 size_t startCount = mVertexBufferStart[iSector].size();
98 size_t startCountInner = mVertexBuffer[iSector].size();
99
100 if (iCol == 0 && iSector == 0) {
101 for (uint32_t i = 0; i < mIOPtrs->nTOFClusters; i++) {
102 mVertexBuffer[iSector].emplace_back(mGlobalPosTOF[i].x, mGlobalPosTOF[i].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTOF[i].z);
103 }
104 }
105
106 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
107 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
108}
109
110GPUDisplay::vboList GPUDisplay::DrawSpacePointsITS(int32_t iSector, int32_t select, int32_t iCol)
111{
112 size_t startCount = mVertexBufferStart[iSector].size();
113 size_t startCountInner = mVertexBuffer[iSector].size();
114
115 if (iCol == 0 && iSector == 0 && mIOPtrs->itsClusters) {
116 for (uint32_t i = 0; i < mIOPtrs->nItsClusters; i++) {
117 mVertexBuffer[iSector].emplace_back(mGlobalPosITS[i].x, mGlobalPosITS[i].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosITS[i].z);
118 }
119 }
120
121 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
122 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
123}
124
125GPUDisplay::vboList GPUDisplay::DrawClusters(int32_t iSector, int32_t select, uint32_t iCol)
126{
127 size_t startCount = mVertexBufferStart[iSector].size();
128 size_t startCountInner = mVertexBuffer[iSector].size();
129 if (mOverlayTFClusters.size() > 0 || iCol == 0 || mNCollissions) {
130 const int32_t firstCluster = (mOverlayTFClusters.size() > 1 && iCol > 0) ? mOverlayTFClusters[iCol - 1][iSector] : 0;
131 const int32_t lastCluster = (mOverlayTFClusters.size() > 1 && iCol + 1 < mOverlayTFClusters.size()) ? mOverlayTFClusters[iCol][iSector] : (mParam->par.earlyTpcTransform ? mIOPtrs->nClusterData[iSector] : mIOPtrs->clustersNative ? mIOPtrs->clustersNative->nClustersSector[iSector] : 0);
132 const bool checkClusterCollision = mQA && mNCollissions && mOverlayTFClusters.size() == 0 && mIOPtrs->clustersNative && mIOPtrs->clustersNative->clustersMCTruth;
133 for (int32_t cidInSector = firstCluster; cidInSector < lastCluster; cidInSector++) {
134 const int32_t cid = GET_CID(iSector, cidInSector);
135#ifdef GPUCA_TPC_GEOMETRY_O2
136 if (checkClusterCollision) {
137 const auto& labels = mIOPtrs->clustersNative->clustersMCTruth->getLabels(cid);
138 if (labels.size() ? (iCol != mQA->GetMCLabelCol(labels[0])) : (iCol != 0)) {
139 continue;
140 }
141 }
142#else
143 (void)checkClusterCollision;
144#endif
145 if (mCfgH.hideUnmatchedClusters && mQA && mQA->SuppressHit(cid)) {
146 continue;
147 }
148 bool draw = mGlobalPos[cid].w == select;
149
150 if (mCfgH.markAdjacentClusters) {
151 const int32_t attach = mIOPtrs->mergedTrackHitAttachment[cid];
152 if (attach) {
153 if (mCfgH.markAdjacentClusters >= 32) {
154 if (mQA && mQA->clusterRemovable(attach, mCfgH.markAdjacentClusters == 33)) {
155 draw = select == tMARKED;
156 }
157 } else if ((mCfgH.markAdjacentClusters & 2) && (attach & gputpcgmmergertypes::attachTube)) {
158 draw = select == tMARKED;
159 } else if ((mCfgH.markAdjacentClusters & 1) && (attach & (gputpcgmmergertypes::attachGood | gputpcgmmergertypes::attachTube)) == 0) {
160 draw = select == tMARKED;
161 } else if ((mCfgH.markAdjacentClusters & 4) && (attach & gputpcgmmergertypes::attachGoodLeg) == 0) {
162 draw = select == tMARKED;
163 } else if ((mCfgH.markAdjacentClusters & 16) && (attach & gputpcgmmergertypes::attachHighIncl)) {
164 draw = select == tMARKED;
165 } else if (mCfgH.markAdjacentClusters & 8) {
166 if (fabsf(mIOPtrs->mergedTracks[attach & gputpcgmmergertypes::attachTrackMask].GetParam().GetQPt()) > 20.f) {
167 draw = select == tMARKED;
168 }
169 }
170 }
171 } else if (mCfgH.markClusters) {
172 int16_t flags;
173 if (mParam->par.earlyTpcTransform) {
174 flags = mIOPtrs->clusterData[iSector][cidInSector].flags;
175 } else {
176 flags = mIOPtrs->clustersNative->clustersLinear[cid].getFlags();
177 }
178 const bool match = flags & mCfgH.markClusters;
179 draw = (select == tMARKED) ? (match) : (draw && !match);
180 } else if (mCfgH.markFakeClusters) {
181 const bool fake = (mQA->HitAttachStatus(cid));
182 draw = (select == tMARKED) ? (fake) : (draw && !fake);
183 }
184 if (draw) {
185 mVertexBuffer[iSector].emplace_back(mGlobalPos[cid].x, mGlobalPos[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPos[cid].z);
186 }
187 }
188 }
189 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
190 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
191}
192
193GPUDisplay::vboList GPUDisplay::DrawLinks(const GPUTPCTracker& tracker, int32_t id, bool dodown)
194{
195 int32_t iSector = tracker.ISector();
196 if (mCfgH.clustersOnly) {
197 return (vboList(0, 0, iSector));
198 }
199 size_t startCount = mVertexBufferStart[iSector].size();
200 size_t startCountInner = mVertexBuffer[iSector].size();
201 for (int32_t i = 0; i < GPUCA_ROW_COUNT; i++) {
202 const GPUTPCRow& row = tracker.Data().Row(i);
203
204 if (i < GPUCA_ROW_COUNT - 2) {
205 const GPUTPCRow& rowUp = tracker.Data().Row(i + 2);
206 for (int32_t j = 0; j < row.NHits(); j++) {
207 if (tracker.Data().HitLinkUpData(row, j) != CALINK_INVAL) {
208 const int32_t cid1 = GET_CID(iSector, tracker.Data().ClusterDataIndex(row, j));
209 const int32_t cid2 = GET_CID(iSector, tracker.Data().ClusterDataIndex(rowUp, tracker.Data().HitLinkUpData(row, j)));
210 drawPointLinestrip(iSector, cid1, id);
211 drawPointLinestrip(iSector, cid2, id);
212 }
213 }
214 }
215
216 if (dodown && i >= 2) {
217 const GPUTPCRow& rowDown = tracker.Data().Row(i - 2);
218 for (int32_t j = 0; j < row.NHits(); j++) {
219 if (tracker.Data().HitLinkDownData(row, j) != CALINK_INVAL) {
220 const int32_t cid1 = GET_CID(iSector, tracker.Data().ClusterDataIndex(row, j));
221 const int32_t cid2 = GET_CID(iSector, tracker.Data().ClusterDataIndex(rowDown, tracker.Data().HitLinkDownData(row, j)));
222 drawPointLinestrip(iSector, cid1, id);
223 drawPointLinestrip(iSector, cid2, id);
224 }
225 }
226 }
227 }
228 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
229 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
230}
231
232GPUDisplay::vboList GPUDisplay::DrawSeeds(const GPUTPCTracker& tracker)
233{
234 int32_t iSector = tracker.ISector();
235 if (mCfgH.clustersOnly) {
236 return (vboList(0, 0, iSector));
237 }
238 size_t startCount = mVertexBufferStart[iSector].size();
239 for (uint32_t i = 0; i < *tracker.NStartHits(); i++) {
240 const GPUTPCHitId& hit = tracker.TrackletStartHit(i);
241 size_t startCountInner = mVertexBuffer[iSector].size();
242 int32_t ir = hit.RowIndex();
243 calink ih = hit.HitIndex();
244 do {
245 const GPUTPCRow& row = tracker.Data().Row(ir);
246 const int32_t cid = GET_CID(iSector, tracker.Data().ClusterDataIndex(row, ih));
247 drawPointLinestrip(iSector, cid, tSEED);
248 ir += 2;
249 ih = tracker.Data().HitLinkUpData(row, ih);
250 } while (ih != CALINK_INVAL);
251 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
252 }
253 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
254}
255
256GPUDisplay::vboList GPUDisplay::DrawTracklets(const GPUTPCTracker& tracker)
257{
258 int32_t iSector = tracker.ISector();
259 if (mCfgH.clustersOnly) {
260 return (vboList(0, 0, iSector));
261 }
262 size_t startCount = mVertexBufferStart[iSector].size();
263 for (uint32_t i = 0; i < *tracker.NTracklets(); i++) {
264 const GPUTPCTracklet& tracklet = tracker.Tracklet(i);
265 size_t startCountInner = mVertexBuffer[iSector].size();
266 float4 oldpos;
267 for (int32_t j = tracklet.FirstRow(); j <= tracklet.LastRow(); j++) {
268 const calink rowHit = tracker.TrackletRowHits()[tracklet.FirstHit() + (j - tracklet.FirstRow())];
269 if (rowHit != CALINK_INVAL && rowHit != CALINK_DEAD_CHANNEL) {
270 const GPUTPCRow& row = tracker.Data().Row(j);
271 const int32_t cid = GET_CID(iSector, tracker.Data().ClusterDataIndex(row, rowHit));
272 oldpos = mGlobalPos[cid];
273 drawPointLinestrip(iSector, cid, tTRACKLET);
274 }
275 }
276 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
277 }
278 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
279}
280
281GPUDisplay::vboList GPUDisplay::DrawTracks(const GPUTPCTracker& tracker, int32_t global)
282{
283 int32_t iSector = tracker.ISector();
284 if (mCfgH.clustersOnly) {
285 return (vboList(0, 0, iSector));
286 }
287 size_t startCount = mVertexBufferStart[iSector].size();
288 for (uint32_t i = (global ? tracker.CommonMemory()->nLocalTracks : 0); i < (global ? *tracker.NTracks() : tracker.CommonMemory()->nLocalTracks); i++) {
289 GPUTPCTrack& track = tracker.Tracks()[i];
290 size_t startCountInner = mVertexBuffer[iSector].size();
291 for (int32_t j = 0; j < track.NHits(); j++) {
292 const GPUTPCHitId& hit = tracker.TrackHits()[track.FirstHitID() + j];
293 const GPUTPCRow& row = tracker.Data().Row(hit.RowIndex());
294 const int32_t cid = GET_CID(iSector, tracker.Data().ClusterDataIndex(row, hit.HitIndex()));
295 drawPointLinestrip(iSector, cid, tSECTORTRACK + global);
296 }
297 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
298 }
299 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
300}
301
302void GPUDisplay::DrawTrackITS(int32_t trackId, int32_t iSector)
303{
304 const auto& trk = mIOPtrs->itsTracks[trackId];
305 for (int32_t k = 0; k < trk.getNClusters(); k++) {
306 int32_t cid = mIOPtrs->itsTrackClusIdx[trk.getFirstClusterEntry() + k];
307 mVertexBuffer[iSector].emplace_back(mGlobalPosITS[cid].x, mGlobalPosITS[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosITS[cid].z);
308 mGlobalPosITS[cid].w = tITSATTACHED;
309 }
310}
311
312GPUDisplay::vboList GPUDisplay::DrawFinalITS()
313{
314 const int32_t iSector = 0;
315 size_t startCount = mVertexBufferStart[iSector].size();
316 for (uint32_t i = 0; i < mIOPtrs->nItsTracks; i++) {
317 if (mITSStandaloneTracks[i]) {
318 size_t startCountInner = mVertexBuffer[iSector].size();
319 DrawTrackITS(i, iSector);
320 insertVertexList(iSector, startCountInner, mVertexBuffer[iSector].size());
321 }
322 }
323 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
324}
325
326template <class T>
327void GPUDisplay::DrawFinal(int32_t iSector, int32_t /*iCol*/, const GPUTPCGMPropagator* prop, std::array<vecpod<int32_t>, 2>& trackList, threadVertexBuffer& threadBuffer)
328{
329 auto& vBuf = threadBuffer.vBuf;
330 auto& buffer = threadBuffer.buffer;
331 uint32_t nTracks = std::max(trackList[0].size(), trackList[1].size());
332 if (mCfgH.clustersOnly) {
333 nTracks = 0;
334 }
335 for (uint32_t ii = 0; ii < nTracks; ii++) {
336 int32_t i = 0;
337 const T* track = nullptr;
338 int32_t lastCluster = -1;
339 while (true) {
340 if (ii >= trackList[0].size()) {
341 break;
342 }
343 i = trackList[0][ii];
344 int32_t nClusters;
345 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
346 track = &mIOPtrs->mergedTracks[i];
347 nClusters = track->NClusters();
348 } else if constexpr (std::is_same_v<T, o2::tpc::TrackTPC>) {
349 track = &mIOPtrs->outputTracksTPCO2[i];
350 nClusters = track->getNClusters();
351 if (!mIOPtrs->clustersNative) {
352 break;
353 }
354 } else {
355 throw std::runtime_error("invalid type");
356 }
357
358 size_t startCountInner = mVertexBuffer[iSector].size();
359 bool drawing = false;
360
361 if constexpr (std::is_same_v<T, o2::tpc::TrackTPC>) {
362 if (!mCfgH.drawTracksAndFilter && !(mCfgH.drawTPCTracks || (mCfgH.drawITSTracks && mIOPtrs->tpcLinkITS && mIOPtrs->tpcLinkITS[i] != -1) || (mCfgH.drawTRDTracks && mIOPtrs->tpcLinkTRD && mIOPtrs->tpcLinkTRD[i] != -1) || (mCfgH.drawTOFTracks && mIOPtrs->tpcLinkTOF && mIOPtrs->tpcLinkTOF[i] != -1))) {
363 break;
364 }
365 if (mCfgH.drawTracksAndFilter && ((mCfgH.drawITSTracks && !(mIOPtrs->tpcLinkITS && mIOPtrs->tpcLinkITS[i] != -1)) || (mCfgH.drawTRDTracks && !(mIOPtrs->tpcLinkTRD && mIOPtrs->tpcLinkTRD[i] != -1)) || (mCfgH.drawTOFTracks && !(mIOPtrs->tpcLinkTOF && mIOPtrs->tpcLinkTOF[i] != -1)))) {
366 break;
367 }
368 }
369
370 if (mCfgH.trackFilter && !mTrackFilter[i]) {
371 break;
372 }
373
374 // Print TOF part of track
375 if constexpr (std::is_same_v<T, o2::tpc::TrackTPC>) {
376 if (mIOPtrs->tpcLinkTOF && mIOPtrs->tpcLinkTOF[i] != -1 && mIOPtrs->nTOFClusters) {
377 int32_t cid = mIOPtrs->tpcLinkTOF[i];
378 drawing = true;
379 mVertexBuffer[iSector].emplace_back(mGlobalPosTOF[cid].x, mGlobalPosTOF[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTOF[cid].z);
380 mGlobalPosTOF[cid].w = tTOFATTACHED;
381 }
382 }
383
384 // Print TRD part of track
385 auto tmpDoTRDTracklets = [&](const auto& trk) {
386 for (int32_t k = 5; k >= 0; k--) {
387 int32_t cid = trk.getTrackletIndex(k);
388 if (cid < 0) {
389 continue;
390 }
391 drawing = true;
392 mVertexBuffer[iSector].emplace_back(mGlobalPosTRD2[cid].x, mGlobalPosTRD2[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTRD2[cid].z);
393 mVertexBuffer[iSector].emplace_back(mGlobalPosTRD[cid].x, mGlobalPosTRD[cid].y * mYFactor, mCfgH.projectXY ? 0 : mGlobalPosTRD[cid].z);
394 mGlobalPosTRD[cid].w = tTRDATTACHED;
395 }
396 };
397 if (std::is_same_v<T, GPUTPCGMMergedTrack> || (!mIOPtrs->tpcLinkTRD && mIOPtrs->trdTracksO2)) {
398 if (mChain && ((int32_t)mConfig.showTPCTracksFromO2Format == (int32_t)mChain->GetProcessingSettings().trdTrackModelO2) && mTRDTrackIds[i] != -1 && mIOPtrs->nTRDTracklets) {
399 if (mIOPtrs->trdTracksO2) {
400 tmpDoTRDTracklets(mIOPtrs->trdTracksO2[mTRDTrackIds[i]]);
401 } else {
402 tmpDoTRDTracklets(mIOPtrs->trdTracks[mTRDTrackIds[i]]);
403 }
404 }
405 } else if constexpr (std::is_same_v<T, o2::tpc::TrackTPC>) {
406 if (mIOPtrs->tpcLinkTRD && mIOPtrs->tpcLinkTRD[i] != -1 && mIOPtrs->nTRDTracklets) {
407 if ((mIOPtrs->tpcLinkTRD[i] & 0x40000000) ? mIOPtrs->nTRDTracksITSTPCTRD : mIOPtrs->nTRDTracksTPCTRD) {
408 const auto* container = (mIOPtrs->tpcLinkTRD[i] & 0x40000000) ? mIOPtrs->trdTracksITSTPCTRD : mIOPtrs->trdTracksTPCTRD;
409 const auto& trk = container[mIOPtrs->tpcLinkTRD[i] & 0x3FFFFFFF];
410 tmpDoTRDTracklets(trk);
411 }
412 }
413 }
414
415 // Print TPC part of track
416 int32_t separateExtrapolatedTracksLimit = (mCfgH.separateExtrapolatedTracks ? tEXTRAPOLATEDTRACK : TRACK_TYPE_ID_LIMIT);
417 uint32_t lastSide = -1;
418 for (int32_t k = 0; k < nClusters; k++) {
419 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
420 if (mCfgH.hideRejectedClusters && (mIOPtrs->mergedTrackHits[track->FirstClusterRef() + k].state & GPUTPCGMMergedTrackHit::flagReject)) {
421 continue;
422 }
423 }
424 int32_t cid;
425 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
426 cid = mIOPtrs->mergedTrackHits[track->FirstClusterRef() + k].num;
427 } else {
428 cid = &track->getCluster(mIOPtrs->outputClusRefsTPCO2, k, *mIOPtrs->clustersNative) - mIOPtrs->clustersNative->clustersLinear;
429 }
430 int32_t w = mGlobalPos[cid].w;
431 if (drawing) {
432 if (mCfgH.splitCETracks && lastSide != (mGlobalPos[cid].z < 0)) {
433 insertVertexList(vBuf[0], startCountInner, mVertexBuffer[iSector].size());
434 drawing = false;
435 lastCluster = -1;
436 } else {
437 drawPointLinestrip(iSector, cid, tFINALTRACK, separateExtrapolatedTracksLimit);
438 }
439 }
440 if (w == separateExtrapolatedTracksLimit) {
441 if (drawing) {
442 insertVertexList(vBuf[0], startCountInner, mVertexBuffer[iSector].size());
443 }
444 drawing = false;
445 } else {
446 if (!drawing) {
447 startCountInner = mVertexBuffer[iSector].size();
448 if (lastCluster != -1 && (!mCfgH.splitCETracks || lastSide == (mGlobalPos[cid].z < 0))) {
449 int32_t lastcid;
450 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
451 lastcid = mIOPtrs->mergedTrackHits[track->FirstClusterRef() + lastCluster].num;
452 } else {
453 lastcid = &track->getCluster(mIOPtrs->outputClusRefsTPCO2, lastCluster, *mIOPtrs->clustersNative) - mIOPtrs->clustersNative->clustersLinear;
454 }
455 drawPointLinestrip(iSector, lastcid, tFINALTRACK, separateExtrapolatedTracksLimit);
456 }
457 drawPointLinestrip(iSector, cid, tFINALTRACK, separateExtrapolatedTracksLimit);
458 }
459 drawing = true;
460 }
461 lastCluster = k;
462 lastSide = mGlobalPos[cid].z < 0;
463 }
464
465 // Print ITS part of track
466 if constexpr (std::is_same_v<T, o2::tpc::TrackTPC>) {
467 if (mIOPtrs->tpcLinkITS && mIOPtrs->tpcLinkITS[i] != -1 && mIOPtrs->nItsTracks && mIOPtrs->nItsClusters) {
468 DrawTrackITS(mIOPtrs->tpcLinkITS[i], iSector);
469 }
470 }
471 insertVertexList(vBuf[0], startCountInner, mVertexBuffer[iSector].size());
472 break;
473 }
474
475 if (!mIOPtrs->clustersNative) {
476 continue;
477 }
478
479 // Propagate track paramters / plot MC tracks
480 for (int32_t iMC = 0; iMC < 2; iMC++) {
481 if (iMC) {
482 if (ii >= trackList[1].size()) {
483 continue;
484 }
485 i = trackList[1][ii];
486 } else {
487 if (track == nullptr) {
488 continue;
489 }
490 if (lastCluster == -1) {
491 continue;
492 }
493 }
494
495 size_t startCountInner = mVertexBuffer[iSector].size();
496 for (int32_t inFlyDirection = 0; inFlyDirection < 2; inFlyDirection++) {
498 float ZOffset = 0;
499 float x = 0;
500 float alphaOrg = 0;
501 if (iMC == 0) {
502 if (!inFlyDirection && mIOPtrs->tpcLinkITS && mIOPtrs->tpcLinkITS[i] != -1) {
503 continue;
504 }
505 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
506 trkParam.Set(track->GetParam());
507 alphaOrg = mParam->Alpha(iSector);
508 } else {
510 convertTrackParam(t, *track);
511 alphaOrg = track->getAlpha();
512 trkParam.Set(t);
513 }
514
515 if (mParam->par.earlyTpcTransform) {
516 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
517 x = mIOPtrs->mergedTrackHitsXYZ[track->FirstClusterRef() + lastCluster].x;
518 ZOffset = track->GetParam().GetTZOffset();
519 }
520 } else {
521 float y, z;
522 if constexpr (std::is_same_v<T, GPUTPCGMMergedTrack>) {
523 auto cl = mIOPtrs->mergedTrackHits[track->FirstClusterRef() + lastCluster];
524 const auto& cln = mIOPtrs->clustersNative->clustersLinear[cl.num];
525 GPUTPCConvertImpl::convert(*mCalib->fastTransform, *mParam, cl.sector, cl.row, cln.getPad(), cln.getTime(), x, y, z);
526 ZOffset = mCalib->fastTransformHelper->getCorrMap()->convVertexTimeToZOffset(iSector, track->GetParam().GetTZOffset(), mParam->continuousMaxTimeBin);
527 } else {
528 uint8_t sector, row;
529 auto cln = track->getCluster(mIOPtrs->outputClusRefsTPCO2, lastCluster, *mIOPtrs->clustersNative, sector, row);
530 GPUTPCConvertImpl::convert(*mCalib->fastTransform, *mParam, sector, row, cln.getPad(), cln.getTime(), x, y, z);
531 ZOffset = mCalib->fastTransformHelper->getCorrMap()->convVertexTimeToZOffset(sector, track->getTime0(), mParam->continuousMaxTimeBin);
532 }
533 }
534 } else {
535 const GPUTPCMCInfo& mc = mIOPtrs->mcInfosTPC[i];
536 if (mc.charge == 0.f) {
537 break;
538 }
539 if (mc.pid < 0) {
540 break;
541 }
542
543 alphaOrg = mParam->Alpha(iSector);
544 float c = cosf(alphaOrg);
545 float s = sinf(alphaOrg);
546 float mclocal[4];
547 x = mc.x;
548 float y = mc.y;
549 mclocal[0] = x * c + y * s;
550 mclocal[1] = -x * s + y * c;
551 float px = mc.pX;
552 float py = mc.pY;
553 mclocal[2] = px * c + py * s;
554 mclocal[3] = -px * s + py * c;
555 float charge = mc.charge > 0 ? 1.f : -1.f;
556
557 x = mclocal[0];
558#ifdef GPUCA_TPC_GEOMETRY_O2
559 trkParam.Set(mclocal[0], mclocal[1], mc.z, mclocal[2], mclocal[3], mc.pZ, -charge); // TODO: DR: unclear to me why we need -charge here
560 if (mParam->par.continuousTracking) {
561 ZOffset = fabsf(mCalib->fastTransformHelper->getCorrMap()->convVertexTimeToZOffset(0, mc.t0, mParam->continuousMaxTimeBin)) * (mc.z < 0 ? -1 : 1);
562 }
563#else
564 if (fabsf(mc.z) > GPUTPCGeometry::TPCLength()) {
565 ZOffset = mc.z > 0 ? (mc.z - GPUTPCGeometry::TPCLength()) : (mc.z + GPUTPCGeometry::TPCLength());
566 }
567 trkParam.Set(mclocal[0], mclocal[1], mc.z - ZOffset, mclocal[2], mclocal[3], mc.pZ, charge);
568#endif
569 }
570 float z0 = trkParam.Z();
571 if (iMC && inFlyDirection == 0) {
572 buffer.clear();
573 }
574 if (x < 1) {
575 break;
576 }
577 if (fabsf(trkParam.SinPhi()) > 1) {
578 break;
579 }
580 float alpha = alphaOrg;
581 vecpod<vtx>& useBuffer = iMC && inFlyDirection == 0 ? buffer : mVertexBuffer[iSector];
582 int32_t nPoints = 0;
583
584 while (nPoints++ < 5000) {
585 if ((inFlyDirection == 0 && x < 0) || (inFlyDirection && x * x + trkParam.Y() * trkParam.Y() > (iMC ? (450 * 450) : (300 * 300)))) {
586 break;
587 }
588 if (fabsf(trkParam.Z() + ZOffset) > mMaxClusterZ + (iMC ? 0 : 0)) {
589 break;
590 }
591 if (fabsf(trkParam.Z() - z0) > (iMC ? GPUTPCGeometry::TPCLength() : GPUTPCGeometry::TPCLength())) {
592 break;
593 }
594 if (inFlyDirection) {
595 if (fabsf(trkParam.SinPhi()) > 0.4f) {
596 float dalpha = asinf(trkParam.SinPhi());
597 trkParam.Rotate(dalpha);
598 alpha += dalpha;
599 }
600 x = trkParam.X() + 1.f;
601 if (!mCfgH.propagateLoopers) {
602 float diff = fabsf(alpha - alphaOrg) / (2.f * CAMath::Pi());
603 diff -= floor(diff);
604 if (diff > 0.25f && diff < 0.75f) {
605 break;
606 }
607 }
608 }
609 float B[3];
610 prop->GetBxByBz(alpha, trkParam.GetX(), trkParam.GetY(), trkParam.GetZ(), B);
611 float dLp = 0;
612 if (trkParam.PropagateToXBxByBz(x, B[0], B[1], B[2], dLp)) {
613 break;
614 }
615 if (fabsf(trkParam.SinPhi()) > 0.9f) {
616 break;
617 }
618 float sa = sinf(alpha), ca = cosf(alpha);
619 float drawX = trkParam.X() + mCfgH.xAdd;
620 useBuffer.emplace_back((ca * drawX - sa * trkParam.Y()) * GL_SCALE_FACTOR, (ca * trkParam.Y() + sa * drawX) * mYFactor * GL_SCALE_FACTOR, mCfgH.projectXY ? 0 : (trkParam.Z() + ZOffset) * GL_SCALE_FACTOR);
621 x += inFlyDirection ? 1 : -1;
622 }
623
624 if (inFlyDirection == 0) {
625 if (iMC) {
626 for (int32_t k = (int32_t)buffer.size() - 1; k >= 0; k--) {
627 mVertexBuffer[iSector].emplace_back(buffer[k]);
628 }
629 } else {
630 insertVertexList(vBuf[1], startCountInner, mVertexBuffer[iSector].size());
631 startCountInner = mVertexBuffer[iSector].size();
632 }
633 }
634 }
635 insertVertexList(vBuf[iMC ? 3 : 2], startCountInner, mVertexBuffer[iSector].size());
636 }
637 }
638}
639
640GPUDisplay::vboList GPUDisplay::DrawGrid(const GPUTPCTracker& tracker)
641{
642 int32_t iSector = tracker.ISector();
643 size_t startCount = mVertexBufferStart[iSector].size();
644 size_t startCountInner = mVertexBuffer[iSector].size();
645 for (int32_t i = 0; i < GPUCA_ROW_COUNT; i++) {
646 const GPUTPCRow& row = tracker.Data().Row(i);
647 for (int32_t j = 0; j <= (signed)row.Grid().Ny(); j++) {
648 float z1 = row.Grid().ZMin();
649 float z2 = row.Grid().ZMax();
650 float x = row.X() + mCfgH.xAdd;
651 float y = row.Grid().YMin() + (float)j / row.Grid().StepYInv();
652 float zz1, zz2, yy1, yy2, xx1, xx2;
653 mParam->Sector2Global(tracker.ISector(), x, y, z1, &xx1, &yy1, &zz1);
654 mParam->Sector2Global(tracker.ISector(), x, y, z2, &xx2, &yy2, &zz2);
655 if (iSector < 18) {
656 zz1 += mCfgH.zAdd;
657 zz2 += mCfgH.zAdd;
658 } else {
659 zz1 -= mCfgH.zAdd;
660 zz2 -= mCfgH.zAdd;
661 }
662 mVertexBuffer[iSector].emplace_back(xx1 * GL_SCALE_FACTOR, yy1 * GL_SCALE_FACTOR * mYFactor, zz1 * GL_SCALE_FACTOR);
663 mVertexBuffer[iSector].emplace_back(xx2 * GL_SCALE_FACTOR, yy2 * GL_SCALE_FACTOR * mYFactor, zz2 * GL_SCALE_FACTOR);
664 }
665 for (int32_t j = 0; j <= (signed)row.Grid().Nz(); j++) {
666 float y1 = row.Grid().YMin();
667 float y2 = row.Grid().YMax();
668 float x = row.X() + mCfgH.xAdd;
669 float z = row.Grid().ZMin() + (float)j / row.Grid().StepZInv();
670 float zz1, zz2, yy1, yy2, xx1, xx2;
671 mParam->Sector2Global(tracker.ISector(), x, y1, z, &xx1, &yy1, &zz1);
672 mParam->Sector2Global(tracker.ISector(), x, y2, z, &xx2, &yy2, &zz2);
673 if (iSector < 18) {
674 zz1 += mCfgH.zAdd;
675 zz2 += mCfgH.zAdd;
676 } else {
677 zz1 -= mCfgH.zAdd;
678 zz2 -= mCfgH.zAdd;
679 }
680 mVertexBuffer[iSector].emplace_back(xx1 * GL_SCALE_FACTOR, yy1 * GL_SCALE_FACTOR * mYFactor, zz1 * GL_SCALE_FACTOR);
681 mVertexBuffer[iSector].emplace_back(xx2 * GL_SCALE_FACTOR, yy2 * GL_SCALE_FACTOR * mYFactor, zz2 * GL_SCALE_FACTOR);
682 }
683 }
684 insertVertexList(tracker.ISector(), startCountInner, mVertexBuffer[iSector].size());
685 return (vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector));
686}
687
688GPUDisplay::vboList GPUDisplay::DrawGridTRD(int32_t sector)
689{
690 // TODO: tilted pads ignored at the moment
691 size_t startCount = mVertexBufferStart[sector].size();
692 size_t startCountInner = mVertexBuffer[sector].size();
693 auto* geo = trdGeometry();
694 if (geo) {
695 int32_t trdsector = NSECTORS / 2 - 1 - sector;
696 float alpha = geo->GetAlpha() / 2.f + geo->GetAlpha() * trdsector;
697 if (trdsector >= 9) {
698 alpha -= 2 * CAMath::Pi();
699 }
700 for (int32_t iLy = 0; iLy < GPUTRDTracker::EGPUTRDTracker::kNLayers; iLy++) {
701 for (int32_t iStack = 0; iStack < GPUTRDTracker::EGPUTRDTracker::kNStacks; iStack++) {
702 int32_t iDet = geo->GetDetector(iLy, iStack, trdsector);
703 auto matrix = geo->GetClusterMatrix(iDet);
704 if (!matrix) {
705 continue;
706 }
707 auto pp = geo->GetPadPlane(iDet);
708 for (int32_t i = 0; i < pp->GetNrows(); i++) {
709 float xyzLoc1[3];
710 float xyzLoc2[3];
711 float xyzGlb1[3];
712 float xyzGlb2[3];
713 xyzLoc1[0] = xyzLoc2[0] = geo->AnodePos();
714 xyzLoc1[1] = pp->GetCol0();
715 xyzLoc2[1] = pp->GetColEnd();
716 xyzLoc1[2] = xyzLoc2[2] = pp->GetRowPos(i) - pp->GetRowPos(pp->GetNrows() / 2);
717 matrix->LocalToMaster(xyzLoc1, xyzGlb1);
718 matrix->LocalToMaster(xyzLoc2, xyzGlb2);
719 float x1Tmp = xyzGlb1[0];
720 xyzGlb1[0] = xyzGlb1[0] * cosf(alpha) + xyzGlb1[1] * sinf(alpha);
721 xyzGlb1[1] = -x1Tmp * sinf(alpha) + xyzGlb1[1] * cosf(alpha);
722 float x2Tmp = xyzGlb2[0];
723 xyzGlb2[0] = xyzGlb2[0] * cosf(alpha) + xyzGlb2[1] * sinf(alpha);
724 xyzGlb2[1] = -x2Tmp * sinf(alpha) + xyzGlb2[1] * cosf(alpha);
725 mVertexBuffer[sector].emplace_back(xyzGlb1[0] * GL_SCALE_FACTOR, xyzGlb1[1] * GL_SCALE_FACTOR * mYFactor, xyzGlb1[2] * GL_SCALE_FACTOR);
726 mVertexBuffer[sector].emplace_back(xyzGlb2[0] * GL_SCALE_FACTOR, xyzGlb2[1] * GL_SCALE_FACTOR * mYFactor, xyzGlb2[2] * GL_SCALE_FACTOR);
727 }
728 for (int32_t j = 0; j < pp->GetNcols(); ++j) {
729 float xyzLoc1[3];
730 float xyzLoc2[3];
731 float xyzGlb1[3];
732 float xyzGlb2[3];
733 xyzLoc1[0] = xyzLoc2[0] = geo->AnodePos();
734 xyzLoc1[1] = xyzLoc2[1] = pp->GetColPos(j) + pp->GetColSize(j) / 2.f;
735 xyzLoc1[2] = pp->GetRow0() - pp->GetRowPos(pp->GetNrows() / 2);
736 xyzLoc2[2] = pp->GetRowEnd() - pp->GetRowPos(pp->GetNrows() / 2);
737 matrix->LocalToMaster(xyzLoc1, xyzGlb1);
738 matrix->LocalToMaster(xyzLoc2, xyzGlb2);
739 float x1Tmp = xyzGlb1[0];
740 xyzGlb1[0] = xyzGlb1[0] * cosf(alpha) + xyzGlb1[1] * sinf(alpha);
741 xyzGlb1[1] = -x1Tmp * sinf(alpha) + xyzGlb1[1] * cosf(alpha);
742 float x2Tmp = xyzGlb2[0];
743 xyzGlb2[0] = xyzGlb2[0] * cosf(alpha) + xyzGlb2[1] * sinf(alpha);
744 xyzGlb2[1] = -x2Tmp * sinf(alpha) + xyzGlb2[1] * cosf(alpha);
745 mVertexBuffer[sector].emplace_back(xyzGlb1[0] * GL_SCALE_FACTOR, xyzGlb1[1] * GL_SCALE_FACTOR * mYFactor, xyzGlb1[2] * GL_SCALE_FACTOR);
746 mVertexBuffer[sector].emplace_back(xyzGlb2[0] * GL_SCALE_FACTOR, xyzGlb2[1] * GL_SCALE_FACTOR * mYFactor, xyzGlb2[2] * GL_SCALE_FACTOR);
747 }
748 }
749 }
750 }
751 insertVertexList(sector, startCountInner, mVertexBuffer[sector].size());
752 return (vboList(startCount, mVertexBufferStart[sector].size() - startCount, sector));
753}
754
755size_t GPUDisplay::DrawGLScene_updateVertexList()
756{
757 for (int32_t i = 0; i < NSECTORS; i++) {
758 mVertexBuffer[i].clear();
759 mVertexBufferStart[i].clear();
760 mVertexBufferCount[i].clear();
761 }
762
763 for (int32_t i = 0; i < mCurrentClusters; i++) {
764 mGlobalPos[i].w = tCLUSTER;
765 }
766 for (int32_t i = 0; i < mCurrentSpacePointsTRD; i++) {
767 mGlobalPosTRD[i].w = tTRDCLUSTER;
768 }
769
770 for (int32_t iSector = 0; iSector < NSECTORS; iSector++) {
771 for (int32_t i = 0; i < N_POINTS_TYPE; i++) {
772 mGlDLPoints[iSector][i].resize(mNCollissions);
773 }
774 for (int32_t i = 0; i < N_FINAL_TYPE; i++) {
775 mGlDLFinal[iSector].resize(mNCollissions);
776 }
777 }
778 int32_t numThreads = getNumThreads();
779 tbb::task_arena(numThreads).execute([&] {
780 if (mChain && (mChain->GetRecoSteps() & GPUDataTypes::RecoStep::TPCSectorTracking)) {
781 tbb::parallel_for(0, NSECTORS, [&](int32_t iSector) {
782 GPUTPCTracker& tracker = (GPUTPCTracker&)sectorTracker(iSector);
783 tracker.SetPointersDataLinks(tracker.LinkTmpMemory());
784 mGlDLLines[iSector][tINITLINK] = DrawLinks(tracker, tINITLINK, true);
785 tracker.SetPointersDataLinks(mChain->rec()->Res(tracker.MemoryResLinks()).Ptr()); // clang-format off
786 }, tbb::simple_partitioner()); // clang-format on
787
788 tbb::parallel_for(0, NSECTORS, [&](int32_t iSector) {
789 const GPUTPCTracker& tracker = sectorTracker(iSector);
790
791 mGlDLLines[iSector][tLINK] = DrawLinks(tracker, tLINK);
792 mGlDLLines[iSector][tSEED] = DrawSeeds(tracker);
793 mGlDLLines[iSector][tTRACKLET] = DrawTracklets(tracker);
794 mGlDLLines[iSector][tSECTORTRACK] = DrawTracks(tracker, 0);
795 mGlDLGrid[iSector] = DrawGrid(tracker);
796 if (iSector < NSECTORS / 2) {
797 mGlDLGridTRD[iSector] = DrawGridTRD(iSector);
798 } // clang-format off
799 }, tbb::simple_partitioner()); // clang-format on
800
801 tbb::parallel_for(0, NSECTORS, [&](int32_t iSector) {
802 const GPUTPCTracker& tracker = sectorTracker(iSector);
803 mGlDLLines[iSector][tEXTRAPOLATEDTRACK] = DrawTracks(tracker, 1); // clang-format off
804 }, tbb::simple_partitioner()); // clang-format on
805 }
806 tbb::parallel_for(0, numThreads, [&](int32_t iThread) {
807 mThreadTracks[iThread].resize(mNCollissions);
808 for (int32_t i = 0; i < mNCollissions; i++) {
809 for (int32_t j = 0; j < NSECTORS; j++) {
810 for (int32_t k = 0; k < 2; k++) {
811 mThreadTracks[iThread][i][j][k].clear();
812 }
813 }
814 } // clang-format off
815 }, tbb::simple_partitioner()); // clang-format on
816 if (mConfig.showTPCTracksFromO2Format) {
817#ifdef GPUCA_TPC_GEOMETRY_O2
818 uint32_t col = 0;
819 tbb::parallel_for<uint32_t>(0, mIOPtrs->nOutputTracksTPCO2, [&](auto i) {
820 uint8_t sector, row;
821 if (mIOPtrs->clustersNative) {
822 mIOPtrs->outputTracksTPCO2[i].getCluster(mIOPtrs->outputClusRefsTPCO2, 0, *mIOPtrs->clustersNative, sector, row);
823 } else {
824 sector = 0;
825 }
826 if (mQA && mIOPtrs->outputTracksTPCO2MC) {
827 col = mQA->GetMCLabelCol(mIOPtrs->outputTracksTPCO2MC[i]);
828 }
829 mThreadTracks[GPUReconstruction::getHostThreadIndex()][col][sector][0].emplace_back(i);
830 });
831#endif
832 } else {
833 tbb::parallel_for<uint32_t>(0, mIOPtrs->nMergedTracks, [&](auto i) {
834 const GPUTPCGMMergedTrack* track = &mIOPtrs->mergedTracks[i];
835 if (track->NClusters() == 0) {
836 return;
837 }
838 if (mCfgH.hideRejectedTracks && !track->OK()) {
839 return;
840 }
841 int32_t sector = mIOPtrs->mergedTrackHits[track->FirstClusterRef() + track->NClusters() - 1].sector;
842 uint32_t col = 0;
843 if (mQA) {
844 const auto& label = mQA->GetMCTrackLabel(i);
845#ifdef GPUCA_TPC_GEOMETRY_O2
846 col = mQA->GetMCLabelCol(label);
847#else
848 while (label.isValid() && col < mOverlayTFClusters.size() && mOverlayTFClusters[col][NSECTORS] < label.track) {
849 col++;
850 }
851#endif
852 }
853 mThreadTracks[GPUReconstruction::getHostThreadIndex()][col][sector][0].emplace_back(i);
854 });
855 }
856 for (uint32_t col = 0; col < mIOPtrs->nMCInfosTPCCol; col++) {
857 tbb::parallel_for(mIOPtrs->mcInfosTPCCol[col].first, mIOPtrs->mcInfosTPCCol[col].first + mIOPtrs->mcInfosTPCCol[col].num, [&](uint32_t i) {
858 const GPUTPCMCInfo& mc = mIOPtrs->mcInfosTPC[i];
859 if (mc.charge == 0.f) {
860 return;
861 }
862 if (mc.pid < 0) {
863 return;
864 }
865
866 float alpha = atan2f(mc.y, mc.x);
867 if (alpha < 0) {
868 alpha += 2 * CAMath::Pi();
869 }
870 int32_t sector = alpha / (2 * CAMath::Pi()) * 18;
871 if (mc.z < 0) {
872 sector += 18;
873 }
874 mThreadTracks[GPUReconstruction::getHostThreadIndex()][col][sector][1].emplace_back(i);
875 });
876 }
877
879 prop.SetMaxSinPhi(.999);
880 prop.SetMaterialTPC();
881 prop.SetPolynomialField(&mParam->polynomialField);
882
883 tbb::parallel_for(0, NSECTORS, [&](int32_t iSector) {
884 int32_t numThread = GPUReconstruction::getHostThreadIndex();
885 for (int32_t iCol = 0; iCol < mNCollissions; iCol++) {
886 mThreadBuffers[numThread].clear();
887 for (int32_t iSet = 0; iSet < numThreads; iSet++) {
888 if (mConfig.showTPCTracksFromO2Format) {
889 DrawFinal<o2::tpc::TrackTPC>(iSector, iCol, &prop, mThreadTracks[iSet][iCol][iSector], mThreadBuffers[numThread]);
890 } else {
891 DrawFinal<GPUTPCGMMergedTrack>(iSector, iCol, &prop, mThreadTracks[iSet][iCol][iSector], mThreadBuffers[numThread]);
892 }
893 }
894 vboList* list = &mGlDLFinal[iSector][iCol][0];
895 for (int32_t i = 0; i < N_FINAL_TYPE; i++) {
896 size_t startCount = mVertexBufferStart[iSector].size();
897 for (uint32_t j = 0; j < mThreadBuffers[numThread].start[i].size(); j++) {
898 mVertexBufferStart[iSector].emplace_back(mThreadBuffers[numThread].start[i][j]);
899 mVertexBufferCount[iSector].emplace_back(mThreadBuffers[numThread].count[i][j]);
900 }
901 list[i] = vboList(startCount, mVertexBufferStart[iSector].size() - startCount, iSector);
902 }
903 } // clang-format off
904 }, tbb::simple_partitioner()); // clang-format on
905
906 tbb::parallel_for(0, NSECTORS, [&](int32_t iSector) {
907 for (int32_t i = 0; i < N_POINTS_TYPE_TPC; i++) {
908 for (int32_t iCol = 0; iCol < mNCollissions; iCol++) {
909 mGlDLPoints[iSector][i][iCol] = DrawClusters(iSector, i, iCol);
910 }
911 } // clang-format off
912 }, tbb::simple_partitioner()); // clang-format on
913 });
914 // End omp parallel
915
916 mGlDLFinalITS = DrawFinalITS();
917
918 for (int32_t iSector = 0; iSector < NSECTORS; iSector++) {
919 for (int32_t i = N_POINTS_TYPE_TPC; i < N_POINTS_TYPE_TPC + N_POINTS_TYPE_TRD; i++) {
920 for (int32_t iCol = 0; iCol < mNCollissions; iCol++) {
921 mGlDLPoints[iSector][i][iCol] = DrawSpacePointsTRD(iSector, i, iCol);
922 }
923 }
924 }
925
926 for (int32_t iSector = 0; iSector < NSECTORS; iSector++) {
927 for (int32_t i = N_POINTS_TYPE_TPC + N_POINTS_TYPE_TRD; i < N_POINTS_TYPE_TPC + N_POINTS_TYPE_TRD + N_POINTS_TYPE_TOF; i++) {
928 for (int32_t iCol = 0; iCol < mNCollissions; iCol++) {
929 mGlDLPoints[iSector][i][iCol] = DrawSpacePointsTOF(iSector, i, iCol);
930 }
931 }
932 break; // TODO: Only sector 0 filled for now
933 }
934
935 for (int32_t iSector = 0; iSector < NSECTORS; iSector++) {
936 for (int32_t i = N_POINTS_TYPE_TPC + N_POINTS_TYPE_TRD + N_POINTS_TYPE_TOF; i < N_POINTS_TYPE_TPC + N_POINTS_TYPE_TRD + N_POINTS_TYPE_TOF + N_POINTS_TYPE_ITS; i++) {
937 for (int32_t iCol = 0; iCol < mNCollissions; iCol++) {
938 mGlDLPoints[iSector][i][iCol] = DrawSpacePointsITS(iSector, i, iCol);
939 }
940 }
941 break; // TODO: Only sector 0 filled for now
942 }
943
944 mUpdateVertexLists = 0;
945 size_t totalVertizes = 0;
946 for (int32_t i = 0; i < NSECTORS; i++) {
947 totalVertizes += mVertexBuffer[i].size();
948 }
949 if (totalVertizes > 0xFFFFFFFF) {
950 throw std::runtime_error("Display vertex count exceeds 32bit uint32_t counter");
951 }
952 size_t needMultiVBOSize = mBackend->needMultiVBO();
953 mUseMultiVBO = needMultiVBOSize && (totalVertizes * sizeof(mVertexBuffer[0][0]) >= needMultiVBOSize);
954 if (!mUseMultiVBO) {
955 size_t totalYet = mVertexBuffer[0].size();
956 mVertexBuffer[0].resize(totalVertizes);
957 for (int32_t i = 1; i < GPUCA_NSECTORS; i++) {
958 for (uint32_t j = 0; j < mVertexBufferStart[i].size(); j++) {
959 mVertexBufferStart[i][j] += totalYet;
960 }
961 memcpy(&mVertexBuffer[0][totalYet], &mVertexBuffer[i][0], mVertexBuffer[i].size() * sizeof(mVertexBuffer[i][0]));
962 totalYet += mVertexBuffer[i].size();
963 mVertexBuffer[i].clear();
964 }
965 }
966 mBackend->loadDataToGPU(totalVertizes);
967 for (int32_t i = 0; i < (mUseMultiVBO ? GPUCA_NSECTORS : 1); i++) {
968 mVertexBuffer[i].clear();
969 }
970 return totalVertizes;
971}
A const (ready only) version of MCTruthContainer.
int16_t charge
Definition RawEventData.h:5
int32_t i
#define GET_CID(sector, i)
#define CALINK_DEAD_CHANNEL
Definition GPUTPCDef.h:22
#define CALINK_INVAL
Definition GPUTPCDef.h:21
#define GPUCA_NSECTORS
#define GPUCA_ROW_COUNT
Online TRD tracker based on extrapolated TPC tracks.
TRD Tracklet word for GPU tracker - 32bit tracklet info + half chamber ID + index.
uint32_t j
Definition RawData.h:0
uint32_t col
Definition RawData.h:4
uint32_t c
Definition RawData.h:2
Definition of the ITS track.
int nClusters
Definition B.h:16
gsl::span< const TruthElement > getLabels(uint32_t dataindex) const
GPUReconstruction::RecoStepField GetRecoSteps() const
Definition GPUChain.h:70
const GPUSettingsProcessing & GetProcessingSettings() const
Definition GPUChain.h:75
const GPUConstantMem * GetProcessors() const
Definition GPUChain.h:67
bool clusterRemovable(int32_t attach, bool prot) const
Definition GPUQA.h:52
int32_t HitAttachStatus(int32_t iHit) const
Definition GPUQA.h:50
bool SuppressHit(int32_t iHit) const
Definition GPUQA.h:49
bool match(const std::vector< std::string > &queries, const char *pattern)
Definition dcs-ccdb.cxx:229
GLfloat GLfloat GLfloat alpha
Definition glcorearb.h:279
GLint GLenum GLint x
Definition glcorearb.h:403
GLint GLsizei count
Definition glcorearb.h:399
GLuint buffer
Definition glcorearb.h:655
GLuint GLfloat GLfloat GLfloat GLfloat y1
Definition glcorearb.h:5034
GLsizeiptr size
Definition glcorearb.h:659
GLint y
Definition glcorearb.h:270
GLuint GLsizei const GLchar * label
Definition glcorearb.h:2519
typedef void(APIENTRYP PFNGLCULLFACEPROC)(GLenum mode)
GLbitfield flags
Definition glcorearb.h:1570
GLuint start
Definition glcorearb.h:469
GLubyte GLubyte GLubyte GLubyte w
Definition glcorearb.h:852
GLdouble GLdouble GLdouble z
Definition glcorearb.h:843
uint8_t itsSharedClusterMap uint8_t
Node par(int index)
Parameters.
std::vector< InputSpec > select(char const *matcher="")
uint32_t calink
Definition GPUTPCDef.h:30
float float float float float z2
Definition MathUtils.h:44
float float float float z1
Definition MathUtils.h:44
const float3 float float float y2
Definition MathUtils.h:42
if(!okForPhiMin(phi0, phi1))
Definition list.h:40
S< o2::trd::GeometryFlat >::type * trdGeometry
S< TPCFastTransform >::type * fastTransform
S< CorrectionMapsHelper >::type * fastTransformHelper
GPUTPCTracker tpcTrackers[GPUCA_NSECTORS]
const GPUTPCGMMergedTrackHitXYZ * mergedTrackHitsXYZ
const o2::tpc::ClusterNativeAccess * clustersNative
const GPUTPCMCInfo * mcInfosTPC
const o2::BaseCluster< float > * itsClusters
const uint32_t * mergedTrackHitAttachment
const GPUTRDTrackGPU * trdTracks
const GPUTRDTrackletWord * trdTracklets
const GPUTPCClusterData * clusterData[NSECTORS]
const o2::its::TrackITS * itsTracks
const o2::MCCompLabel * outputTracksTPCO2MC
const o2::tpc::TrackTPC * outputTracksTPCO2
const GPUTPCGMMergedTrackHit * mergedTrackHits
const GPUTRDTrack * trdTracksITSTPCTRD
const GPUTPCGMMergedTrack * mergedTracks
const o2::dataformats::ConstMCTruthContainerView< o2::MCCompLabel > * clustersMCTruth
const ClusterNative * clustersLinear
o2::InteractionRecord ir(0, 0)
std::vector< int > row
typename std::vector< T, vecpod_allocator< T > > vecpod
Definition vecpod.h:31