Merge 84ca389 into sapling-pr-archive-ktf

Author: ktf

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

@@ -40,6 +40,7 @@ GPUg() void fitTrackSeedsKernel(
   CellSeed* trackSeeds,
   const TrackingFrameInfo** foundTrackingFrameInfo,
   o2::its::TrackITSExt* tracks,
+  const float* minPts,
   const unsigned int nSeeds,
   const float Bz,
   const int startLevel,
@@ -182,6 +183,7 @@ void filterCellNeighboursHandler(std::vector<int>&,
 void trackSeedHandler(CellSeed* trackSeeds,
                       const TrackingFrameInfo** foundTrackingFrameInfo,
                       o2::its::TrackITSExt* tracks,
+                      std::vector<float>& minPtsHost,
                       const unsigned int nSeeds,
                       const float Bz,
                       const int startLevel,

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

@@ -278,6 +278,9 @@ void TrackerTraitsGPU<nLayers>::findRoads(const int iteration)
     const int minimumLayer{startLevel - 1};
     std::vector<CellSeed> trackSeeds;
     for (int startLayer{mTrkParams[iteration].CellsPerRoad() - 1}; startLayer >= minimumLayer; --startLayer) {
+      if ((mTrkParams[iteration].StartLayerMask & (1 << (startLayer + 2))) == 0) {
+        continue;
+      }
       std::vector<int> lastCellId, updatedCellId;
       std::vector<CellSeed> lastCellSeed, updatedCellSeed;
 
@@ -308,6 +311,7 @@ void TrackerTraitsGPU<nLayers>::findRoads(const int iteration)
     trackSeedHandler(mTimeFrameGPU->getDeviceTrackSeeds(),             // CellSeed* trackSeeds,
                      mTimeFrameGPU->getDeviceArrayTrackingFrameInfo(), // TrackingFrameInfo** foundTrackingFrameInfo,
                      mTimeFrameGPU->getDeviceTrackITSExt(),            // o2::its::TrackITSExt* tracks,
+                     mTrkParams[iteration].MinPt,                      // std::vector<float>& minPtsHost,
                      trackSeeds.size(),                                // const size_t nSeeds,
                      mBz,                                              // const float Bz,
                      startLevel,                                       // const int startLevel,

Detectors/ITSMFT/ITS/tracking/GPU/cuda/TrackingKernels.cu

@@ -267,6 +267,7 @@ GPUg() void fitTrackSeedsKernel(
   CellSeed* trackSeeds,
   const TrackingFrameInfo** foundTrackingFrameInfo,
   o2::its::TrackITSExt* tracks,
+  const float* minPts,
   const unsigned int nSeeds,
   const float Bz,
   const int startLevel,
@@ -317,7 +318,7 @@ GPUg() void fitTrackSeedsKernel(
                           foundTrackingFrameInfo,   // TrackingFrameInfo** trackingFrameInfo,
                           propagator,               // const o2::base::Propagator* propagator,
                           matCorrType);             // o2::base::PropagatorF::MatCorrType matCorrType
-    if (!fitSuccess) {
+    if (!fitSuccess || temporaryTrack.getPt() < minPts[nLayers - temporaryTrack.getNClusters()]) {
       continue;
     }
     tracks[iCurrentTrackSeedIndex] = temporaryTrack;
@@ -1089,6 +1090,7 @@ void filterCellNeighboursHandler(std::vector<int>& neighHost,
 void trackSeedHandler(CellSeed* trackSeeds,
                       const TrackingFrameInfo** foundTrackingFrameInfo,
                       o2::its::TrackITSExt* tracks,
+                      std::vector<float>& minPtsHost,
                       const unsigned int nSeeds,
                       const float Bz,
                       const int startLevel,
@@ -1099,17 +1101,19 @@ void trackSeedHandler(CellSeed* trackSeeds,
                       const int nBlocks,
                       const int nThreads)
 {
+  thrust::device_vector<float> minPts(minPtsHost);
   gpu::fitTrackSeedsKernel<<<nBlocks, nThreads>>>(
-    trackSeeds,               // CellSeed*
-    foundTrackingFrameInfo,   // TrackingFrameInfo**
-    tracks,                   // TrackITSExt*
-    nSeeds,                   // const unsigned int
-    Bz,                       // const float
-    startLevel,               // const int
-    maxChi2ClusterAttachment, // float
-    maxChi2NDF,               // float
-    propagator,               // const o2::base::Propagator*
-    matCorrType);             // o2::base::PropagatorF::MatCorrType
+    trackSeeds,                           // CellSeed*
+    foundTrackingFrameInfo,               // TrackingFrameInfo**
+    tracks,                               // TrackITSExt*
+    thrust::raw_pointer_cast(&minPts[0]), // const float* minPts,
+    nSeeds,                               // const unsigned int
+    Bz,                                   // const float
+    startLevel,                           // const int
+    maxChi2ClusterAttachment,             // float
+    maxChi2NDF,                           // float
+    propagator,                           // const o2::base::Propagator*
+    matCorrType);                         // o2::base::PropagatorF::MatCorrType
 
   gpuCheckError(cudaPeekAtLastError());
   gpuCheckError(cudaDeviceSynchronize());

Framework/AnalysisSupport/src/AODJAlienReaderHelpers.cxx

@@ -10,10 +10,12 @@
 // or submit itself to any jurisdiction.
 
 #include "AODJAlienReaderHelpers.h"
+#include <memory>
 #include "Framework/TableTreeHelpers.h"
 #include "Framework/AnalysisHelpers.h"
 #include "Framework/DataProcessingStats.h"
 #include "Framework/RootTableBuilderHelpers.h"
+#include "Framework/RootArrowFilesystem.h"
 #include "Framework/AlgorithmSpec.h"
 #include "Framework/ConfigParamRegistry.h"
 #include "Framework/ControlService.h"
@@ -41,6 +43,8 @@
 #include <arrow/io/interfaces.h>
 #include <arrow/table.h>
 #include <arrow/util/key_value_metadata.h>
+#include <arrow/dataset/dataset.h>
+#include <arrow/dataset/file_base.h>
 
 using namespace o2;
 using namespace o2::aod;
@@ -272,11 +276,13 @@ AlgorithmSpec AODJAlienReaderHelpers::rootFileReaderCallback(ConfigContext const
             // Origin file name for derived output map
             auto o2 = Output(TFFileNameHeader);
             auto fileAndFolder = didir->getFileFolder(dh, fcnt, ntf);
-            std::string currentFilename(fileAndFolder.file->GetName());
-            if (strcmp(fileAndFolder.file->GetEndpointUrl()->GetProtocol(), "file") == 0 && fileAndFolder.file->GetEndpointUrl()->GetFile()[0] != '/') {
+            auto rootFS = std::dynamic_pointer_cast<TFileFileSystem>(fileAndFolder.filesystem());
+            auto* f = dynamic_cast<TFile*>(rootFS->GetFile());
+            std::string currentFilename(f->GetFile()->GetName());
+            if (strcmp(f->GetEndpointUrl()->GetProtocol(), "file") == 0 && f->GetEndpointUrl()->GetFile()[0] != '/') {
               // This is not an absolute local path. Make it absolute.
               static std::string pwd = gSystem->pwd() + std::string("/");
-              currentFilename = pwd + std::string(fileAndFolder.file->GetName());
+              currentFilename = pwd + std::string(f->GetName());
             }
             outputs.make<std::string>(o2) = currentFilename;
           }
@@ -312,7 +318,9 @@ AlgorithmSpec AODJAlienReaderHelpers::rootFileReaderCallback(ConfigContext const
       auto concrete = DataSpecUtils::asConcreteDataMatcher(firstRoute.matcher);
       auto dh = header::DataHeader(concrete.description, concrete.origin, concrete.subSpec);
       auto fileAndFolder = didir->getFileFolder(dh, fcnt, ntf);
-      if (!fileAndFolder.file) {
+
+      // In case the filesource is empty, move to the next one.
+      if (fileAndFolder.path().empty()) {
         fcnt += 1;
         ntf = 0;
         if (didir->atEnd(fcnt)) {