VecGeom/navigation/NavStatePath.h

0001 /// \file NavStatePath.h
0002 /// \author Sandro Wenzel (sandro.wenzel@cern.ch)
0003 /// \date 12.03.2014
0004
0005 #ifndef VECGEOM_NAVIGATION_NAVSTATEPATH_H_
0006 #define VECGEOM_NAVIGATION_NAVSTATEPATH_H_
0007
0008 #include "VecGeom/base/Config.h"
0009
0010 #include "VecGeom/base/VariableSizeObj.h"
0011 #include "VecGeom/base/Transformation3D.h"
0012 #include "VecGeom/volumes/PlacedVolume.h"
0013 #include "VecGeom/management/GeoManager.h"
0014 #ifdef VECGEOM_ENABLE_CUDA
0015 #include "VecGeom/management/CudaManager.h"
0016 #endif
0017 #include "VecGeom/base/Global.h"
0018
0019 #include <iostream>
0020 #include <string>
0021
0022 class TGeoBranchArray;
0023
0024 // gcc 4.8.2's -Wnon-virtual-dtor is broken and turned on by -Weffc++, we
0025 // need to disable it for SOA3D
0026
0027 #if __GNUC__ < 3 || (__GNUC__ == 4 && __GNUC_MINOR__ <= 8)
0028
0029 #pragma GCC diagnostic push
0030 #pragma GCC diagnostic ignored "-Wnon-virtual-dtor"
0031 #pragma GCC diagnostic ignored "-Weffc++"
0032 #define GCC_DIAG_POP_NEEDED
0033 #endif
0034
0035 namespace vecgeom {
0036 inline namespace VECGEOM_IMPL_NAMESPACE {
0037
0038 // the NavStateIndex type determines is used
0039 // to calculate addresses of PlacedVolumes
0040 // a short type should be used in case the number of PlacedVolumes can
0041 // be counted with 16bits
0042 // TODO: consider putting uint16 + uint32 types
0043 #ifdef VECGEOM_USE_INDEXEDNAVSTATES
0044 // typedef unsigned short NavStateIndex_t;
0045 typedef unsigned long NavStateIndex_t;
0046 #else
0047 typedef VPlacedVolume const *NavStateIndex_t;
0048 #endif
0049
0050 // helper functionality to convert from NavStateIndex_t to *PlacedVolumes and back
0051 // the template abstraction also allows to go back to pointers as NavStateIndex_t
0052 // via a template specialization
0053 // T stands for NavStateIndex_t
0054 template <typename T>
0055 struct Index2PVolumeConverter {
0056   VECCORE_ATT_HOST_DEVICE
0057   VECGEOM_FORCE_INLINE
0058   static VPlacedVolume const *ToPlacedVolume(T index)
0059   {
0060 #ifdef VECCORE_CUDA_DEVICE_COMPILATION
0061     // checking here for NVCC_DEVICE since the global variable globaldevicegeomgata::gCompact...
0062     // is marked __device__ and can only be compiled within device compiler passes
0063     assert(vecgeom::globaldevicegeomdata::gCompactPlacedVolBuffer != nullptr);
0064     return &vecgeom::globaldevicegeomdata::gCompactPlacedVolBuffer[index];
0065 #else
0066 #ifndef VECCORE_CUDA
0067     assert(vecgeom::GeoManager::gCompactPlacedVolBuffer == nullptr ||
0068            vecgeom::GeoManager::gCompactPlacedVolBuffer[index].id() == index);
0069     return &vecgeom::GeoManager::gCompactPlacedVolBuffer[index];
0070 #else
0071     // this is the case when we compile with nvcc for host side
0072     // (failed previously due to undefined symbol vecgeom::cuda::GeoManager::gCompactPlacedVolBuffer)
0073     assert(false && "reached unimplement code");
0074     (void)index; // avoid unused parameter warning.
0075     return nullptr;
0076 #endif
0077 #endif
0078   }
0079
0080   VECCORE_ATT_HOST_DEVICE
0081   VECGEOM_FORCE_INLINE
0082   static T ToIndex(VPlacedVolume const *pvol) { return pvol->id(); }
0083 };
0084
0085 // template specialization when we directly save VPlacedVolume pointers into the NavStates
0086 template <>
0087 struct Index2PVolumeConverter<VPlacedVolume const *> {
0088   VECCORE_ATT_HOST_DEVICE
0089   VECGEOM_FORCE_INLINE
0090   static VPlacedVolume const *ToPlacedVolume(VPlacedVolume const *pvol) { return pvol; }
0091   VECCORE_ATT_HOST_DEVICE
0092   VECGEOM_FORCE_INLINE
0093   static VPlacedVolume const *ToIndex(VPlacedVolume const *pvol) { return pvol; }
0094 };
0095
0096 /**
0097  * A class describing a current geometry state
0098  * likely there will be such an object for each particle/track currently treated.
0099  */
0100 class NavStatePath : protected VariableSizeObjectInterface<NavStatePath, NavStateIndex_t>,
0101                      private Index2PVolumeConverter<NavStateIndex_t> {
0102 public:
0103   using Value_t        = NavStateIndex_t;
0104   using Base_t         = VariableSizeObjectInterface<NavStatePath, Value_t>;
0105   using VariableData_t = VariableSizeObj<Value_t>;
0106
0107 private:
0108   friend Base_t;
0109
0110   // Required by VariableSizeObjectInterface
0111   VECCORE_ATT_HOST_DEVICE
0112   VariableData_t &GetVariableData() { return fPath; }
0113   VECCORE_ATT_HOST_DEVICE
0114   const VariableData_t &GetVariableData() const { return fPath; }
0115
0116   unsigned char
0117       fCurrentLevel; // value indicating the next free slot in the fPath array ( ergo the current geometry depth )
0118   // we choose unsigned char in order to save memory, thus supporting geometry depths up to 255 which seems large enough
0119
0120   // a member to cache some state information across state usages
0121   // one particular example could be a calculated index for this state
0122   // if fCache == -1 it means the we have to recalculate it; otherwise we don't
0123   short fCache;
0124
0125   bool fOnBoundary; // flag indicating whether track is on boundary of the "Top()" placed volume
0126
0127 #ifdef VECGEOM_CACHED_TRANS
0128   // Cached transformation for top level
0129   bool fCacheM = true; // flag indicating whether the global matrix for the state is cached
0130   Transformation3D fTopTrans;
0131 #endif
0132
0133   // pointer data follows; has to be last
0134   VariableSizeObj<Value_t> fPath;
0135
0136   // constructors and assignment operators are private
0137   // states have to be constructed using MakeInstance() function
0138   VECGEOM_FORCE_INLINE
0139   VECCORE_ATT_HOST_DEVICE
0140   NavStatePath(size_t nvalues);
0141
0142   VECGEOM_FORCE_INLINE
0143   VECCORE_ATT_HOST_DEVICE
0144   NavStatePath(size_t new_size, NavStatePath &other)
0145       : fCurrentLevel(other.fCurrentLevel), fCache(-1), fOnBoundary(other.fOnBoundary),
0146 #ifdef VECGEOM_CACHED_TRANS
0147         fCacheM(other.fCacheM), fTopTrans(other.fTopTrans),
0148 #endif
0149         fPath(new_size, other.fPath)
0150   {
0151     // Raw memcpy of the content to another existing state.
0152     //
0153     // in case NavStatePath was a virtual class: change to
0154     // std::memcpy(other->DataStart(), DataStart(), DataSize());
0155
0156     if (new_size > other.fPath.fN) {
0157       memset(fPath.GetValues() + other.fPath.fN, 0, new_size - other.fPath.fN);
0158     }
0159   }
0160
0161   // some private management methods
0162   VECGEOM_FORCE_INLINE
0163   VECCORE_ATT_HOST_DEVICE
0164   void InitInternalStorage();
0165
0166 private:
0167   // The data start should point to the address of the first data member,
0168   // after the virtual table
0169   // the purpose is probably for the Copy function
0170   const void *DataStart() const { return (const void *)&fCurrentLevel; }
0171   const void *ObjectStart() const { return (const void *)this; }
0172   void *DataStart() { return (void *)&fCurrentLevel; }
0173   void *ObjectStart() { return (void *)this; }
0174
0175   // The actual size of the data for an instance, excluding the virtual table
0176   size_t DataSize() const { return SizeOf() + (size_t)ObjectStart() - (size_t)DataStart(); }
0177
0178 public:
0179   // replaces the volume pointers from CPU volumes in fPath
0180   // to the equivalent pointers on the GPU
0181   // uses the CudaManager to do so
0182   void ConvertToGPUPointers();
0183
0184   // replaces the pointers from GPU volumes in fPath
0185   // to the equivalent pointers on the CPU
0186   // uses the CudaManager to do so
0187   void ConvertToCPUPointers();
0188
0189   // Enumerate the part of the private interface, we want to expose.
0190   using Base_t::MakeCopy;
0191   using Base_t::MakeCopyAt;
0192   using Base_t::ReleaseInstance;
0193   using Base_t::SizeOf;
0194   using Base_t::SizeOfAlignAware;
0195
0196   // Enumerate functions from converter which we want to use
0197   // ( without retyping of the struct name )
0198   using Index2PVolumeConverter<NavStateIndex_t>::ToIndex;
0199   using Index2PVolumeConverter<NavStateIndex_t>::ToPlacedVolume;
0200
0201   // produces a compact navigation state object of a certain depth
0202   // the caller can give a memory address where the object will
0203   // be placed
0204   // the caller has to make sure that the size of the external memory
0205   // is >= sizeof(NavStatePath) + sizeof(VPlacedVolume*)*maxlevel
0206   //
0207   // Methods MakeInstance(), MakeInstanceAt(), MakeCopy() and MakeCopyAt() are provided by
0208   // VariableSizeObjectInterface
0209
0210   VECCORE_ATT_HOST_DEVICE
0211   static NavStatePath *MakeInstance(int maxlevel)
0212   {
0213     // MaxLevel is 'zero' based (i.e. maxlevel==0 requires one value)
0214     return Base_t::MakeInstance(maxlevel + 1);
0215   }
0216
0217   VECCORE_ATT_HOST_DEVICE
0218   static NavStatePath *MakeInstanceAt(int maxlevel, void *addr)
0219   {
0220     // MaxLevel is 'zero' based (i.e. maxlevel==0 requires one value)
0221     return Base_t::MakeInstanceAt(maxlevel + 1, addr);
0222   }
0223
0224   // returns the size in bytes of a NavStatePath object with internal
0225   // path depth maxlevel
0226   VECCORE_ATT_HOST_DEVICE
0227   static size_t SizeOfInstance(int maxlevel)
0228   {
0229     // MaxLevel is 'zero' based (i.e. maxlevel==0 requires one value)
0230     return VariableSizeObjectInterface::SizeOf(maxlevel + 1);
0231   }
0232
0233   // returns the size in bytes of a NavStatePath object with internal
0234   // path depth maxlevel -- including space needed for padding to next aligned object
0235   // of same kind
0236   VECCORE_ATT_HOST_DEVICE
0237   static size_t SizeOfInstanceAlignAware(int maxlevel)
0238   {
0239     // MaxLevel is 'zero' based (i.e. maxlevel==0 requires one value)
0240     return VariableSizeObjectInterface::SizeOfAlignAware(maxlevel + 1);
0241   }
0242
0243   VECCORE_ATT_HOST_DEVICE
0244   int GetObjectSize() const { return SizeOf(GetMaxLevel()); }
0245
0246   VECCORE_ATT_HOST_DEVICE
0247   int SizeOf() const { return NavStatePath::SizeOfInstance(GetMaxLevel()); }
0248
0249   VECGEOM_FORCE_INLINE
0250   VECCORE_ATT_HOST_DEVICE
0251   NavStatePath &operator=(NavStatePath const &rhs);
0252
0253   // functions useful to "serialize" navigationstate
0254   // the Vector-of-Indices basically describes the path on the tree taken from top to bottom
0255   // an index corresponds to a daughter
0256   void GetPathAsListOfIndices(std::list<uint> &indices) const;
0257   void ResetPathFromListOfIndices(VPlacedVolume const *world, std::list<uint> const &indices);
0258
0259   VECCORE_ATT_HOST_DEVICE
0260   void CopyTo(NavStatePath *other) const
0261   {
0262     // Raw memcpy of the content to another existing state.
0263     //
0264     // in case NavStatePath was a virtual class: change to
0265     // std::memcpy(other->DataStart(), DataStart(), DataSize());
0266     bool alloc = other->fPath.fSelfAlloc;
0267     // std::memcpy(other, this, this->SizeOf());
0268     // we only need to copy to relevant depth
0269     // GetCurrentLevel indicates the 'next' level, i.e. currentLevel==0 is empty
0270     // fCurrentLevel = maxlevel+1 is full
0271     // SizeOfInstance expect [0,maxlevel] and add +1 to its params
0272     std::memcpy(static_cast<void *>(other), this, NavStatePath::SizeOfInstance(this->GetCurrentLevel()));
0273
0274     other->fPath.fSelfAlloc = alloc;
0275   }
0276
0277   // copies a fixed and predetermined number of bytes
0278   // might be useful for specialized navigators which know the depth + SizeOf in advance
0279   // N is number of bytes to be copied and can be obtained by a prior call to constexpr NavStatePath::SizeOf( ... );
0280   template <size_t N>
0281   void CopyToFixedSize(NavStatePath *other) const
0282   {
0283     bool alloc = other->fPath.fSelfAlloc;
0284     for (size_t i = 0; i < N; ++i) {
0285       ((char *)other)[i] = ((char *)this)[i];
0286     }
0287     other->fPath.fSelfAlloc = alloc;
0288   }
0289
0290   VECGEOM_FORCE_INLINE
0291   VECCORE_ATT_HOST_DEVICE
0292   ~NavStatePath();
0293
0294   // what else: operator new etc...
0295
0296   VECGEOM_FORCE_INLINE
0297   VECCORE_ATT_HOST_DEVICE
0298   unsigned char GetMaxLevel() const { return fPath.fN - 1; }
0299
0300   VECGEOM_FORCE_INLINE
0301   VECCORE_ATT_HOST_DEVICE
0302   unsigned char GetCurrentLevel() const { return fCurrentLevel; }
0303
0304   VECGEOM_FORCE_INLINE
0305   VECCORE_ATT_HOST_DEVICE
0306   VPlacedVolume const *GetLastExited() const
0307   { /*one beyond current*/
0308     return (fCurrentLevel < GetMaxLevel() - 1) ? ToPlacedVolume(fPath[fCurrentLevel]) : nullptr;
0309   }
0310
0311   VECGEOM_FORCE_INLINE
0312   VECCORE_ATT_HOST_DEVICE
0313   void SetLastExited() {}
0314
0315   // better to use pop and push
0316   VECGEOM_FORCE_INLINE
0317   VECCORE_ATT_HOST_DEVICE
0318   void Push(VPlacedVolume const *);
0319
0320   // Push a given daughter to the state
0321   VECGEOM_FORCE_INLINE
0322   VECCORE_ATT_HOST_DEVICE
0323   void Push(unsigned short);
0324
0325   // a push version operating on IndexTypes
0326   VECGEOM_FORCE_INLINE
0327   VECCORE_ATT_HOST_DEVICE
0328   void PushIndexType(NavStateIndex_t);
0329
0330   VECGEOM_FORCE_INLINE
0331   VECCORE_ATT_HOST_DEVICE
0332   VPlacedVolume const *Top() const;
0333
0334   VECGEOM_FORCE_INLINE
0335   VECCORE_ATT_HOST_DEVICE
0336   VPlacedVolume const *At(int level) const { return ToPlacedVolume(fPath[level]); }
0337
0338   VECGEOM_FORCE_INLINE
0339   VECCORE_ATT_HOST_DEVICE
0340   Value_t ValueAt(int level) const { return fPath[level]; }
0341
0342   // direct write access to the path
0343   // (no one should ever call this function unless you know what you are doing)
0344   // TODO: consider making this private + friend or so
0345   VECGEOM_FORCE_INLINE
0346   VECCORE_ATT_HOST_DEVICE
0347   void SetValueAt(int level, Value_t v) { fPath[level] = v; }
0348
0349   VECGEOM_FORCE_INLINE
0350   VECCORE_ATT_HOST_DEVICE
0351   void TopMatrix(Transformation3D &) const;
0352
0353 #ifdef VECGEOM_CACHED_TRANS
0354   /* @brief Update the cached top matrix */
0355   VECGEOM_FORCE_INLINE
0356   VECCORE_ATT_HOST_DEVICE
0357   void UpdateTopMatrix(Transformation3D *top_matrix = nullptr);
0358 #endif
0359
0360   // returning a "delta" transformation that can transform
0361   // coordinates given in reference frame of this->Top() to the reference frame of other->Top()
0362   // simply with otherlocalcoordinate = delta.Transform( thislocalcoordinate )
0363   VECCORE_ATT_HOST_DEVICE
0364   void DeltaTransformation(NavStatePath const &other, Transformation3D & /* delta */) const;
0365
0366   // VECGEOM_FORCE_INLINE
0367   VECCORE_ATT_HOST_DEVICE
0368   Vector3D<Precision> GlobalToLocal(Vector3D<Precision> const &) const;
0369
0370   VECCORE_ATT_HOST_DEVICE
0371   Vector3D<Precision> GlobalToLocal(Vector3D<Precision> const &, int tolevel) const;
0372
0373   VECCORE_ATT_HOST_DEVICE
0374   void TopMatrix(int tolevel, Transformation3D &) const;
0375
0376   VECGEOM_FORCE_INLINE
0377   VECCORE_ATT_HOST_DEVICE
0378   void Pop();
0379
0380   VECGEOM_FORCE_INLINE
0381   VECCORE_ATT_HOST_DEVICE
0382   int Distance(NavStatePath const &) const;
0383
0384   // returns a string representation of a (relative) sequence of operations/moves
0385   // that transforms this navigation state into the other navigation state
0386   // example:
0387   // state1 = /0/1/1/
0388   // state2 = /0/2/2/3
0389   // results in string
0390   // "/up/horiz/1/down/2/down/3" with 4 operations "up", "horiz", "down", "down"
0391   // the sequence of moves is the following
0392   // up: /0/1/1 --> /0/1/
0393   // horiz/1 : 0/1 --> /0/2 ( == /0/(1+1) )   "we are hopping from daughter 1 to 2 (which corresponds to a step of 1)"
0394   // down/2 : /0/2 --> /0/2/2   "going further down 2nd daughter"
0395   // down/3 : /0/2/2/3 --> /0/2/2/3  "going further down 2nd daughter"
0396   std::string RelativePath(NavStatePath const & /*other*/) const;
0397
0398   // clear all information
0399   VECGEOM_FORCE_INLINE
0400   VECCORE_ATT_HOST_DEVICE
0401   void Clear();
0402
0403   VECCORE_ATT_HOST_DEVICE
0404   void Print() const;
0405
0406   VECGEOM_FORCE_INLINE
0407   VECCORE_ATT_HOST_DEVICE
0408   void Dump() const;
0409
0410   VECGEOM_FORCE_INLINE
0411   VECCORE_ATT_HOST_DEVICE
0412   bool HasSamePathAsOther(NavStatePath const &other) const
0413   {
0414     if (other.fCurrentLevel != fCurrentLevel) return false;
0415     for (int i = fCurrentLevel - 1; i >= 0; --i) {
0416       if (fPath[i] != other.fPath[i]) return false;
0417     }
0418     return true;
0419   }
0420
0421   void printValueSequence(std::ostream & = std::cerr) const;
0422
0423   // calculates a checksum along the path
0424   // can be used (as a quick criterion) to see whether 2 states are same
0425   unsigned long getCheckSum() const
0426   {
0427     unsigned long s = 0;
0428     for (int i = 0; i < fCurrentLevel; ++i) {
0429       s += (unsigned long)(ValueAt(i) + 1); // + 1 as offset otherwise may not distinguish TOP level and OUTSIDE level
0430     }
0431     return s;
0432   }
0433
0434   /**
0435    * returns the number of FILLED LEVELS such that
0436    * state.GetNode( state.GetLevel() ) == state.Top()
0437    */
0438   VECGEOM_FORCE_INLINE
0439   VECCORE_ATT_HOST_DEVICE
0440   unsigned char GetLevel() const { return fCurrentLevel - 1; }
0441
0442   /**
0443     function returning whether the point (current navigation state) is outside the detector setup
0444   */
0445   VECGEOM_FORCE_INLINE
0446   VECCORE_ATT_HOST_DEVICE
0447   bool IsOutside() const { return !(fCurrentLevel > 0); }
0448
0449   VECGEOM_FORCE_INLINE
0450   VECCORE_ATT_HOST_DEVICE
0451   bool IsOnBoundary() const { return fOnBoundary; }
0452
0453 #ifdef VECGEOM_CACHED_TRANS
0454   VECGEOM_FORCE_INLINE
0455   VECCORE_ATT_HOST_DEVICE
0456   bool IsMatrixCached() const { return fCacheM; }
0457 #endif
0458
0459   VECGEOM_FORCE_INLINE
0460   VECCORE_ATT_HOST_DEVICE
0461   void SetBoundaryState(bool b) { fOnBoundary = b; }
0462
0463   VECGEOM_FORCE_INLINE
0464   VECCORE_ATT_HOST_DEVICE
0465   short GetCacheValue() const { return fCache; }
0466
0467   VECGEOM_FORCE_INLINE
0468   VECCORE_ATT_HOST_DEVICE
0469   void SetCacheValue(short v) { fCache = v; }
0470
0471 }; // end of class
0472
0473 VECCORE_ATT_HOST_DEVICE
0474 NavStatePath &NavStatePath::operator=(NavStatePath const &rhs)
0475 {
0476   if (this != &rhs) {
0477     fCurrentLevel = rhs.fCurrentLevel;
0478     fCache        = rhs.fCache;
0479     fOnBoundary   = rhs.fOnBoundary;
0480 #ifdef VECGEOM_CACHED_TRANS
0481     fCacheM   = rhs.fCacheM;
0482     fTopTrans = rhs.fTopTrans;
0483 #endif
0484     // Use memcpy.  Potential truncation if this is smaller than rhs.
0485     fPath = rhs.fPath;
0486   }
0487   return *this;
0488 }
0489
0490 // private implementation of standard constructor
0491 VECCORE_ATT_HOST_DEVICE
0492 NavStatePath::NavStatePath(size_t nvalues) : fCurrentLevel(0), fCache(-1), fOnBoundary(false), fPath(nvalues)
0493 {
0494   // clear the buffer
0495   std::memset(fPath.GetValues(), 0, nvalues * sizeof(NavStateIndex_t));
0496 }
0497
0498 VECCORE_ATT_HOST_DEVICE
0499 NavStatePath::~NavStatePath() {}
0500
0501 VECCORE_ATT_HOST_DEVICE
0502 void NavStatePath::Pop()
0503 {
0504   if (fCurrentLevel > 0) {
0505     fCurrentLevel--;
0506     // note that we are not invalidating the "popped volume" here
0507     // in order to be able to query the last "exited volume" later
0508     fCache = -1;
0509 #ifdef VECGEOM_CACHED_TRANS
0510     fCacheM = false;
0511 #endif
0512   }
0513 }
0514
0515 VECCORE_ATT_HOST_DEVICE
0516 void NavStatePath::Clear()
0517 {
0518   fCurrentLevel = 0;
0519   fOnBoundary   = false;
0520   fCache        = -1;
0521 #ifdef VECGEOM_CACHED_TRANS
0522   fCacheM = true;
0523 #endif
0524 }
0525
0526 VECCORE_ATT_HOST_DEVICE
0527 void NavStatePath::Push(VPlacedVolume const *v)
0528 {
0529 #ifdef DEBUG
0530   assert(fCurrentLevel < GetMaxLevel());
0531 #endif
0532   fPath[fCurrentLevel++] = ToIndex(v);
0533 #ifdef VECGEOM_CACHED_TRANS
0534   fCacheM = false;
0535 #endif
0536 }
0537
0538 // Allow pushing by child index
0539 VECCORE_ATT_HOST_DEVICE
0540 void NavStatePath::Push(unsigned short child)
0541 {
0542   if (fCurrentLevel > 0) {
0543     auto top = ToPlacedVolume(fPath[fCurrentLevel - 1]);
0544 #ifdef DEBUG
0545     assert(child < top->GetDaughters().size());
0546 #endif
0547     fPath[fCurrentLevel++] = ToIndex(top->GetDaughters().operator[](child));
0548 #ifdef VECGEOM_CACHED_TRANS
0549     fCacheM = false;
0550 #endif
0551   }
0552 }
0553
0554 VECCORE_ATT_HOST_DEVICE
0555 void NavStatePath::PushIndexType(NavStateIndex_t v)
0556 {
0557 #ifdef DEBUG
0558   assert(fCurrentLevel < GetMaxLevel());
0559 #endif
0560   fPath[fCurrentLevel++] = v;
0561 #ifdef VECGEOM_CACHED_TRANS
0562   fCacheM = false;
0563 #endif
0564 }
0565
0566 VECCORE_ATT_HOST_DEVICE
0567 VPlacedVolume const *NavStatePath::Top() const
0568 {
0569   return (fCurrentLevel > 0) ? ToPlacedVolume(fPath[fCurrentLevel - 1]) : nullptr;
0570 }
0571
0572 // calculates the global matrix to transform from global coordinates
0573 // to the frame of the top volume in the state
0574 // input: a reference to a transformation object ( which should be initialized to identity )
0575 VECGEOM_FORCE_INLINE
0576 VECCORE_ATT_HOST_DEVICE
0577 void NavStatePath::TopMatrix(Transformation3D &global_matrix) const
0578 {
0579   // this could be actually cached in case the path does not change ( particle stays inside a volume )
0580 #ifdef VECGEOM_CACHED_TRANS
0581   if (fCacheM) {
0582     global_matrix = fTopTrans;
0583     return;
0584   }
0585 #endif
0586   for (int i = 1; i < fCurrentLevel; ++i) {
0587     global_matrix.MultiplyFromRight(*(ToPlacedVolume(fPath[i])->GetTransformation()));
0588   }
0589 }
0590
0591 #ifdef VECGEOM_CACHED_TRANS
0592 // Update the cached top matrix
0593 VECGEOM_FORCE_INLINE
0594 VECCORE_ATT_HOST_DEVICE
0595 void NavStatePath::UpdateTopMatrix(Transformation3D *top_matrix)
0596 {
0597   // Update cached top transformation
0598   if (fCacheM) return;
0599   if (top_matrix) {
0600     fTopTrans = *top_matrix;
0601   } else {
0602     // need to recompute
0603     fTopTrans.Clear();
0604     TopMatrix(fTopTrans);
0605   }
0606   // Flag matrix as cached
0607   fCacheM = true;
0608 }
0609 #endif
0610
0611 VECGEOM_FORCE_INLINE
0612 VECCORE_ATT_HOST_DEVICE
0613 void NavStatePath::Dump() const
0614 {
0615   const unsigned int *ptr = (const unsigned int *)this;
0616 #pragma GCC diagnostic push
0617 #pragma GCC diagnostic ignored "-Wcast-qual"
0618   printf("NavState::Dump(): data: %p(%zu) : %p(%zu) : %p(%zu)\n", (void *)&fCurrentLevel, sizeof(fCurrentLevel),
0619          (void *)&fOnBoundary, sizeof(fOnBoundary), (void *)&fPath, sizeof(fPath));
0620   for (unsigned int i = 0; i < 20; ++i) {
0621     printf("%p: ", (void *)ptr);
0622     for (unsigned int j = 0; j < 8; ++j) {
0623       printf(" %08x ", *ptr);
0624       ptr++;
0625     }
0626     printf("\n");
0627   }
0628 #pragma GCC diagnostic pop
0629 }
0630
0631 /**
0632  * encodes the geometry path as a concatenated string of ( Value_t ) present in fPath
0633  */
0634 inline void NavStatePath::printValueSequence(std::ostream &stream) const
0635 {
0636   for (int i = 0; i < fCurrentLevel; ++i) {
0637     stream << "/" << fPath[i] << "(" << At(i)->GetLabel() << ")";
0638   }
0639 }
0640
0641 /**
0642  * calculates if other navigation state takes a different branch in geometry path or is on same branch
0643  * ( two states are on same branch if one can connect the states just by going upwards or downwards ( or do nothing ))
0644  */
0645 VECGEOM_FORCE_INLINE
0646 VECCORE_ATT_HOST_DEVICE
0647 int NavStatePath::Distance(NavStatePath const &other) const
0648 {
0649   int lastcommonlevel = -1;
0650   int maxlevel        = Min(GetCurrentLevel(), other.GetCurrentLevel());
0651
0652   //  algorithm: start on top and go down until paths split
0653   for (int i = 0; i < maxlevel; i++) {
0654     if (this->At(i) == other.At(i)) {
0655       lastcommonlevel = i;
0656     } else {
0657       break;
0658     }
0659   }
0660
0661   return (GetCurrentLevel() - lastcommonlevel) + (other.GetCurrentLevel() - lastcommonlevel) - 2;
0662 }
0663
0664 inline void NavStatePath::ConvertToGPUPointers()
0665 {
0666 #if !defined(VECCORE_CUDA) && defined(VECGEOM_ENABLE_CUDA) && !defined(VECGEOM_USE_INDEXEDNAVSTATES)
0667   for (int i = 0; i < fCurrentLevel; ++i) {
0668     auto *pvol = vecgeom::CudaManager::Instance().LookupPlaced(ToPlacedVolume(fPath[i])).GetPtr();
0669     fPath[i]   = ToIndex(pvol);
0670   }
0671 #endif
0672 }
0673
0674 inline void NavStatePath::ConvertToCPUPointers()
0675 {
0676 #if !defined(VECCORE_CUDA) && defined(VECGEOM_ENABLE_CUDA) && !defined(VECGEOM_USE_INDEXEDNAVSTATES)
0677   for (int i = 0; i < fCurrentLevel; ++i)
0678     fPath[i] = ToIndex(vecgeom::CudaManager::Instance().LookupPlacedCPUPtr((const void *)ToPlacedVolume(fPath[i])));
0679 #endif
0680 }
0681 } // namespace VECGEOM_IMPL_NAMESPACE
0682 } // namespace vecgeom
0683
0684 #if defined(GCC_DIAG_POP_NEEDED)
0685 #pragma GCC diagnostic pop
0686 #undef GCC_DIAG_POP_NEEDED
0687 #endif
0688
0689 #endif // VECGEOM_NAVIGATION_NAVSTATEPATH_H_