Merge 016da34 into sapling-pr-archive-ktf

Author: ktf

DataFormats/Headers/include/Headers/Stack.h

@@ -14,10 +14,7 @@
 #include "MemoryResources/MemoryResources.h"
 #include "Headers/DataHeader.h"
 
-namespace o2
-{
-
-namespace header
+namespace o2::header
 {
 //__________________________________________________________________________________________________
 /// @struct Stack
@@ -45,20 +42,20 @@ struct Stack {
   };
 
  public:
-  using allocator_type = boost::container::pmr::polymorphic_allocator<std::byte>;
+  using allocator_type = fair::mq::pmr::polymorphic_allocator<std::byte>;
   using value_type = std::byte;
   using BufferType = std::unique_ptr<value_type[], freeobj>; //this gives us proper default move semantics for free
 
   Stack() = default;
   Stack(Stack&&) = default;
   Stack(Stack&) = delete;
   Stack& operator=(Stack&) = delete;
-  Stack& operator=(Stack&&) = default;
+  Stack& operator=(Stack&&) = delete;
 
-  value_type* data() const { return buffer.get(); }
-  size_t size() const { return bufferSize; }
+  [[nodiscard]] value_type* data() const { return buffer.get(); }
+  [[nodiscard]] size_t size() const { return bufferSize; }
   allocator_type get_allocator() const { return allocator; }
-  const BaseHeader* first() const { return reinterpret_cast<const BaseHeader*>(this->data()); }
+  [[nodiscard]] const BaseHeader* first() const { return reinterpret_cast<const BaseHeader*>(this->data()); }
   static const BaseHeader* firstHeader(std::byte const* buf) { return BaseHeader::get(buf); }
   static const BaseHeader* lastHeader(std::byte const* buf)
   {
@@ -90,9 +87,9 @@ struct Stack {
   /// all headers must derive from BaseHeader, in addition also other stacks can be passed to ctor.
   template <typename FirstArgType, typename... Headers,
             typename std::enable_if_t<
-              !std::is_convertible<FirstArgType, boost::container::pmr::polymorphic_allocator<std::byte>>::value, int> = 0>
+              !std::is_convertible<FirstArgType, fair::mq::pmr::polymorphic_allocator<std::byte>>::value, int> = 0>
   Stack(FirstArgType&& firstHeader, Headers&&... headers)
-    : Stack(boost::container::pmr::new_delete_resource(), std::forward<FirstArgType>(firstHeader),
+    : Stack(fair::mq::pmr::new_delete_resource(), std::forward<FirstArgType>(firstHeader),
             std::forward<Headers>(headers)...)
   {
   }
@@ -143,7 +140,7 @@ struct Stack {
   constexpr static size_t calculateSize() { return 0; }
 
  private:
-  allocator_type allocator{boost::container::pmr::new_delete_resource()};
+  allocator_type allocator{fair::mq::pmr::new_delete_resource()};
   size_t bufferSize{0};
   BufferType buffer{nullptr, freeobj{allocator.resource()}};
 
@@ -231,7 +228,7 @@ struct Stack {
   }
 };
 
-} // namespace header
-} // namespace o2
+} // namespace o2::header
+
 
 #endif // HEADERS_STACK_H

DataFormats/Headers/test/testDataHeader.cxx

@@ -314,7 +314,7 @@ BOOST_AUTO_TEST_CASE(headerStack_test)
   BOOST_CHECK(h3->secret == 42);
 
   //test constructing from a buffer and an additional header
-  using namespace boost::container::pmr;
+  using namespace fair::mq::pmr;
   Stack s5(new_delete_resource(), s1.data(), Stack{}, meta);
   BOOST_CHECK(s5.size() == s1.size() + sizeof(meta));
   // check if we can find the header even though there was an empty stack in the middle

DataFormats/MemoryResources/include/MemoryResources/MemoryResources.h

@@ -115,7 +115,7 @@ class MessageResource : public FairMQMemoryResource
 // A spectator pmr memory resource which only watches the memory of the underlying buffer, does not
 // carry out real allocation. It owns the underlying buffer which is destroyed on deallocation.
 template <typename BufferType>
-class SpectatorMemoryResource : public boost::container::pmr::memory_resource
+class SpectatorMemoryResource : public fair::mq::pmr::memory_resource
 {
  public:
   using buffer_type = BufferType;
@@ -183,10 +183,10 @@ class SpectatorMemoryResource : public boost::container::pmr::memory_resource
 // This in general (as in STL) is a bad idea, but here it is safe to inherit from an allocator since we
 // have no additional data and only override some methods so we don't get into slicing and other problems.
 template <typename T>
-class SpectatorAllocator : public boost::container::pmr::polymorphic_allocator<T>
+class SpectatorAllocator : public fair::mq::pmr::polymorphic_allocator<T>
 {
  public:
-  using boost::container::pmr::polymorphic_allocator<T>::polymorphic_allocator;
+  using fair::mq::pmr::polymorphic_allocator<T>::polymorphic_allocator;
   using propagate_on_container_move_assignment = std::true_type;
 
   // skip default construction of empty elements
@@ -243,7 +243,7 @@ class OwningMessageSpectatorAllocator
     return OwningMessageSpectatorAllocator();
   }
 
-  boost::container::pmr::memory_resource* resource() { return &mResource; }
+  fair::mq::pmr::memory_resource* resource() { return &mResource; }
 
   // skip default construction of empty elements
   // this is important for two reasons: one: it allows us to adopt an existing buffer (e.g. incoming message) and
@@ -269,14 +269,14 @@ class OwningMessageSpectatorAllocator
 
 // The NoConstructAllocator behaves like the normal pmr vector but does not call constructors / destructors
 template <typename T>
-class NoConstructAllocator : public boost::container::pmr::polymorphic_allocator<T>
+class NoConstructAllocator : public fair::mq::pmr::polymorphic_allocator<T>
 {
  public:
-  using boost::container::pmr::polymorphic_allocator<T>::polymorphic_allocator;
+  using fair::mq::pmr::polymorphic_allocator<T>::polymorphic_allocator;
   using propagate_on_container_move_assignment = std::true_type;
 
   template <typename... Args>
-  NoConstructAllocator(Args&&... args) : boost::container::pmr::polymorphic_allocator<T>(std::forward<Args>(args)...)
+  NoConstructAllocator(Args&&... args) : fair::mq::pmr::polymorphic_allocator<T>(std::forward<Args>(args)...)
   {
   }
 
@@ -302,9 +302,9 @@ class NoConstructAllocator : public boost::container::pmr::polymorphic_allocator
 //__________________________________________________________________________________________________
 
 using ByteSpectatorAllocator = SpectatorAllocator<std::byte>;
-using BytePmrAllocator = boost::container::pmr::polymorphic_allocator<std::byte>;
+using BytePmrAllocator = fair::mq::pmr::polymorphic_allocator<std::byte>;
 template <class T>
-using vector = std::vector<T, o2::pmr::polymorphic_allocator<T>>;
+using vector = std::vector<T, fair::mq::pmr::polymorphic_allocator<T>>;
 
 //__________________________________________________________________________________________________
 /// Return a std::vector spanned over the contents of the message, takes ownership of the message

DataFormats/MemoryResources/test/testMemoryResources.cxx

@@ -60,7 +60,7 @@ BOOST_AUTO_TEST_CASE(transportallocatormap_test)
   BOOST_CHECK(_tmp == allocZMQ);
 }
 
-using namespace boost::container::pmr;
+using namespace fair::mq::pmr;
 
 BOOST_AUTO_TEST_CASE(allocator_test)
 {

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

@@ -19,6 +19,7 @@
 namespace o2::its
 {
 class CellSeed;
+class ExternalAllocator;
 namespace gpu
 {
 #ifdef GPUCA_GPUCODE // GPUg() global kernels must only when compiled by GPU compiler
@@ -178,7 +179,8 @@ void computeCellNeighboursHandler(CellSeed** cellsLayersDevice,
 
 int filterCellNeighboursHandler(gpuPair<int, int>*,
                                 int*,
-                                unsigned int);
+                                unsigned int,
+                                o2::its::ExternalAllocator* = nullptr);
 
 template <int nLayers = 7>
 void processNeighboursHandler(const int startLayer,
@@ -191,6 +193,7 @@ void processNeighboursHandler(const int startLayer,
                               gsl::span<int*> neighboursDeviceLUTs,
                               const TrackingFrameInfo** foundTrackingFrameInfo,
                               bounded_vector<CellSeed>& seedsHost,
+                              o2::its::ExternalAllocator*,
                               const float bz,
                               const float MaxChi2ClusterAttachment,
                               const float maxChi2NDF,

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

@@ -18,6 +18,7 @@
 #include "ITStrackingGPU/TrackerTraitsGPU.h"
 #include "ITStrackingGPU/TrackingKernels.h"
 #include "ITStracking/TrackingConfigParam.h"
+
 namespace o2::its
 {
 constexpr int UnusedIndex{-1};
@@ -209,7 +210,8 @@ void TrackerTraitsGPU<nLayers>::findCellsNeighbours(const int iteration)
 
     filterCellNeighboursHandler(mTimeFrameGPU->getDeviceNeighbourPairs(iLayer),
                                 mTimeFrameGPU->getDeviceNeighbours(iLayer),
-                                nNeigh);
+                                nNeigh,
+                                mTimeFrameGPU->getExternalAllocator());
   }
   mTimeFrameGPU->createNeighboursDeviceArray();
   mTimeFrameGPU->unregisterRest();
@@ -236,6 +238,7 @@ void TrackerTraitsGPU<nLayers>::findRoads(const int iteration)
                                         mTimeFrameGPU->getDeviceNeighboursLUTs(),
                                         mTimeFrameGPU->getDeviceArrayTrackingFrameInfo(),
                                         trackSeeds,
+                                        mTimeFrameGPU->getExternalAllocator(),
                                         this->mBz,
                                         this->mTrkParams[0].MaxChi2ClusterAttachment,
                                         this->mTrkParams[0].MaxChi2NDF,

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

@@ -28,15 +28,14 @@
 #include "ITStracking/Constants.h"
 #include "ITStracking/IndexTableUtils.h"
 #include "ITStracking/MathUtils.h"
+#include "ITStracking/ExternalAllocator.h"
 #include "DataFormatsITS/TrackITS.h"
 #include "ReconstructionDataFormats/Vertex.h"
 
 #include "ITStrackingGPU/TrackerTraitsGPU.h"
 #include "ITStrackingGPU/TrackingKernels.h"
 #include "ITStrackingGPU/Utils.h"
 
-#include "GPUCommonHelpers.h"
-
 #ifndef __HIPCC__
 #define THRUST_NAMESPACE thrust::cuda
 #else
@@ -64,6 +63,37 @@ GPUdii() float Sq(float v)
 namespace gpu
 {
 
+template <typename T>
+class TypedAllocator : public thrust::device_allocator<T>
+{
+ public:
+  using value_type = T;
+  using pointer = T*;
+
+  template <typename U>
+  struct rebind {
+    using other = TypedAllocator<U>;
+  };
+
+  explicit TypedAllocator(ExternalAllocator* allocPtr)
+    : mInternalAllocator(allocPtr) {}
+
+  T* allocate(size_t n)
+  {
+    return reinterpret_cast<T*>(mInternalAllocator->allocate(n * sizeof(T)));
+  }
+
+  void deallocate(T* p, size_t n)
+  {
+    char* raw_ptr = reinterpret_cast<char*>(p);
+    size_t bytes = n * sizeof(T);
+    mInternalAllocator->deallocate(raw_ptr, bytes); // redundant as internal dealloc is no-op.
+  }
+
+ private:
+  ExternalAllocator* mInternalAllocator;
+};
+
 GPUd() const int4 getBinsRect(const Cluster& currentCluster, const int layerIndex,
                               const o2::its::IndexTableUtils& utils,
                               const float z1, const float z2, float maxdeltaz, float maxdeltaphi)
@@ -1117,7 +1147,8 @@ void computeCellNeighboursHandler(CellSeed** cellsLayersDevice,
 
 int filterCellNeighboursHandler(gpuPair<int, int>* cellNeighbourPairs,
                                 int* cellNeighbours,
-                                unsigned int nNeigh)
+                                unsigned int nNeigh,
+                                o2::its::ExternalAllocator* allocator)
 {
   thrust::device_ptr<gpuPair<int, int>> neighVectorPairs(cellNeighbourPairs);
   thrust::device_ptr<int> validNeighs(cellNeighbours);
@@ -1140,6 +1171,7 @@ void processNeighboursHandler(const int startLayer,
                               gsl::span<int*> neighboursDeviceLUTs,
                               const TrackingFrameInfo** foundTrackingFrameInfo,
                               bounded_vector<CellSeed>& seedsHost,
+                              o2::its::ExternalAllocator* allocator,
                               const float bz,
                               const float maxChi2ClusterAttachment,
                               const float maxChi2NDF,
@@ -1148,8 +1180,10 @@ void processNeighboursHandler(const int startLayer,
                               const int nBlocks,
                               const int nThreads)
 {
-  thrust::device_vector<int> foundSeedsTable(nCells[startLayer] + 1); // Shortcut: device_vector skips central memory management, we are relying on the contingency.
-                                                                      // TODO: fix this.
+  auto allocInt = gpu::TypedAllocator<int>(allocator);
+  auto allocCellSeed = gpu::TypedAllocator<CellSeed>(allocator);
+  thrust::device_vector<int, gpu::TypedAllocator<int>> foundSeedsTable(nCells[startLayer] + 1, 0, allocInt); // Shortcut: device_vector skips central memory management, we are relying on the contingency.
+                                                                                                             // TODO: fix this.
 
   gpu::processNeighboursKernel<true><<<o2::gpu::CAMath::Min(nBlocks, GPU_BLOCKS),
                                        o2::gpu::CAMath::Min(nThreads, GPU_THREADS)>>>(
@@ -1172,8 +1206,8 @@ void processNeighboursHandler(const int startLayer,
     matCorrType);
   gpu::cubExclusiveScanInPlace(foundSeedsTable, nCells[startLayer] + 1);
 
-  thrust::device_vector<int> updatedCellId(foundSeedsTable.back());
-  thrust::device_vector<CellSeed> updatedCellSeed(foundSeedsTable.back());
+  thrust::device_vector<int, gpu::TypedAllocator<int>> updatedCellId(foundSeedsTable.back(), 0, allocInt);
+  thrust::device_vector<CellSeed, gpu::TypedAllocator<CellSeed>> updatedCellSeed(foundSeedsTable.back(), allocCellSeed);
   gpu::processNeighboursKernel<false><<<o2::gpu::CAMath::Min(nBlocks, GPU_BLOCKS),
                                         o2::gpu::CAMath::Min(nThreads, GPU_THREADS)>>>(
     startLayer,
@@ -1195,13 +1229,13 @@ void processNeighboursHandler(const int startLayer,
     matCorrType);
 
   int level = startLevel;
-  thrust::device_vector<int> lastCellId;
-  thrust::device_vector<CellSeed> lastCellSeed;
+  thrust::device_vector<int, gpu::TypedAllocator<int>> lastCellId(allocInt);
+  thrust::device_vector<CellSeed, gpu::TypedAllocator<CellSeed>> lastCellSeed(allocCellSeed);
   for (int iLayer{startLayer - 1}; iLayer > 0 && level > 2; --iLayer) {
     lastCellSeed.swap(updatedCellSeed);
     lastCellId.swap(updatedCellId);
-    thrust::device_vector<CellSeed>().swap(updatedCellSeed);
-    thrust::device_vector<int>().swap(updatedCellId);
+    thrust::device_vector<CellSeed, gpu::TypedAllocator<CellSeed>>(allocCellSeed).swap(updatedCellSeed);
+    thrust::device_vector<int, gpu::TypedAllocator<int>>(allocInt).swap(updatedCellId);
     auto lastCellSeedSize{lastCellSeed.size()};
     foundSeedsTable.resize(lastCellSeedSize + 1);
     thrust::fill(foundSeedsTable.begin(), foundSeedsTable.end(), 0);
@@ -1253,8 +1287,7 @@ void processNeighboursHandler(const int startLayer,
       propagator,
       matCorrType);
   }
-
-  thrust::device_vector<CellSeed> outSeeds(updatedCellSeed.size());
+  thrust::device_vector<CellSeed, gpu::TypedAllocator<CellSeed>> outSeeds(updatedCellSeed.size(), allocCellSeed);
   auto end = thrust::copy_if(updatedCellSeed.begin(), updatedCellSeed.end(), outSeeds.begin(), gpu::seed_selector(1.e3, maxChi2NDF * ((startLevel + 2) * 2 - 5)));
   auto s{end - outSeeds.begin()};
   seedsHost.reserve(seedsHost.size() + s);
@@ -1367,6 +1400,7 @@ template void processNeighboursHandler<7>(const int startLayer,
                                           gsl::span<int*> neighboursDeviceLUTs,
                                           const TrackingFrameInfo** foundTrackingFrameInfo,
                                           bounded_vector<CellSeed>& seedsHost,
+                                          o2::its::ExternalAllocator*,
                                           const float bz,
                                           const float maxChi2ClusterAttachment,
                                           const float maxChi2NDF,

Detectors/ITSMFT/ITS/tracking/include/ITStracking/Cell.h

@@ -67,7 +67,7 @@ class CellSeed final : public o2::track::TrackParCovF
 {
  public:
   GPUhdDefault() CellSeed() = default;
-  GPUhd() CellSeed(int innerL, int cl0, int cl1, int cl2, int trkl0, int trkl1, o2::track::TrackParCovF& tpc, float chi2) : o2::track::TrackParCovF{tpc}, mLevel{1}, mChi2{chi2}
+  GPUhd() CellSeed(int innerL, int cl0, int cl1, int cl2, int trkl0, int trkl1, o2::track::TrackParCovF& tpc, float chi2) : o2::track::TrackParCovF(tpc), mChi2(chi2), mLevel(1)
   {
     setUserField(innerL);
     mClusters[innerL + 0] = cl0;

Detectors/ITSMFT/ITS/tracking/include/ITStracking/ExternalAllocator.h

@@ -23,8 +23,8 @@ class ExternalAllocator
 {
  public:
   virtual void* allocate(size_t) = 0;
+  virtual void deallocate(char*, size_t) = 0;
 };
-
 } // namespace o2::its
 
 #endif

Detectors/ITSMFT/ITS/tracking/include/ITStracking/IOUtils.h

@@ -72,10 +72,6 @@ inline static const o2::itsmft::ChipMappingITS& getChipMappingITS()
   return MP;
 }
 
-std::vector<std::unordered_map<int, Label>> loadLabels(const int, const std::string&);
-void writeRoadsReport(std::ofstream&, std::ofstream&, std::ofstream&, const std::vector<std::vector<Road<5>>>&,
-                      const std::unordered_map<int, Label>&);
-
 template <class iterator, typename T>
 o2::math_utils::Point3D<T> extractClusterData(const itsmft::CompClusterExt& c, iterator& iter, const itsmft::TopologyDictionary* dict, T& sig2y, T& sig2z)
 {