VecGeom/navigation/NavStatePool.h

0001 /*
0002  * NavStatePool.h
0003  *
0004  *  Created on: 14.11.2014
0005  *      Author: swenzel
0006  */
0007
0008 #ifndef NAVSTATEPOOL_H_
0009 #define NAVSTATEPOOL_H_
0010
0011 #include "VecGeom/base/Config.h"
0012 #include "VecGeom/base/Cuda.h"
0013 #include "VecGeom/base/Global.h"
0014 #include "VecGeom/navigation/NavigationState.h"
0015 #ifdef VECGEOM_ENABLE_CUDA
0016 #include "VecGeom/management/CudaManager.h"
0017 #endif
0018 #ifdef VECGEOM_CUDA_INTERFACE
0019 #include "VecGeom/backend/cuda/Interface.h"
0020 #endif
0021
0022 // a fixed (runtime) size  "array" or contiguous
0023 // memory pool of navigation states
0024 // testing some ideas to copy to gpu
0025 // it is supposed to be long-lived ( it has some initialization time overhead because it allocates the
0026 // GPU pointer at startup
0027
0028 #include <iostream>
0029 #include <fstream>
0030
0031 namespace vecgeom {
0032
0033 VECGEOM_DEVICE_FORWARD_DECLARE(template <typename Type> class SOA3D;);
0034 VECGEOM_HOST_FORWARD_DECLARE(class NavStatePool;);
0035 VECGEOM_DEVICE_FORWARD_DECLARE(class NavStatePool;);
0036 inline namespace VECGEOM_IMPL_NAMESPACE {
0037
0038 VECCORE_ATT_HOST_DEVICE
0039 VECCORE_FORCE_INLINE
0040 NavigationState const *GetNavigationState(int i, const char *buffer, int depth)
0041 {
0042   return reinterpret_cast<NavigationState const *>(buffer + NavigationState::SizeOfInstanceAlignAware(depth) * i);
0043 }
0044
0045 VECCORE_ATT_HOST_DEVICE
0046 VECCORE_FORCE_INLINE
0047 NavigationState *GetNavigationState(int i, char *buffer, int depth)
0048 {
0049   return reinterpret_cast<NavigationState*>(buffer + NavigationState::SizeOfInstanceAlignAware(depth) * i);
0050 }
0051
0052 class NavStatePoolView {
0053 public:
0054
0055   VECCORE_ATT_HOST_DEVICE
0056   NavStatePoolView(char *buffer, int depth, int capacity) : fCapacity(capacity), fDepth(depth), fBuffer(buffer) {}
0057
0058   VECCORE_ATT_HOST_DEVICE
0059   NavigationState *operator[](int i)
0060   {
0061     assert(i < fCapacity);
0062     return GetNavigationState(i, fBuffer, fDepth);
0063   }
0064
0065   VECCORE_ATT_HOST_DEVICE
0066   NavigationState const *operator[](int i) const
0067   {
0068     assert(i < fCapacity);
0069     return GetNavigationState(i, fBuffer, fDepth);
0070   }
0071
0072   VECCORE_ATT_HOST_DEVICE
0073   int Capacity() const { return fCapacity; }
0074
0075   VECCORE_ATT_HOST_DEVICE
0076   int Depth() const { return fDepth; }
0077
0078   VECCORE_ATT_HOST_DEVICE
0079   int IsValid() const { return fBuffer && fCapacity > 0 && fDepth > 0; }
0080
0081 private:         // members
0082   int   fCapacity; // Allocated size of the container.
0083   int   fDepth;    // depth of the navigation objects to cover
0084   char *fBuffer;   // the memory buffer in which we place states
0085 }; // end class
0086
0087 class NavStatePool {
0088
0089 public:
0090   NavStatePool(int size, int depth)
0091       : fCapacity(size), fDepth(depth), fBuffer(new char[NavigationState::SizeOfInstanceAlignAware(depth) * size]),
0092         fGPUPointer(NULL)
0093   {
0094
0095 #if !defined(VECCORE_CUDA) && defined(VECGEOM_ENABLE_CUDA)
0096     vecgeom::CudaMalloc(&fGPUPointer, NavigationState::SizeOfInstanceAlignAware(depth) * size);
0097 #endif
0098     // now create the states
0099     for (int i = 0; i < (int)fCapacity; ++i) {
0100       NavigationState::MakeInstanceAt(depth, fBuffer + NavigationState::SizeOfInstanceAlignAware(depth) * i);
0101     }
0102   }
0103
0104   ~NavStatePool() { delete[] fBuffer; }
0105 #if !defined(VECCORE_CUDA) && defined(VECGEOM_ENABLE_CUDA)
0106   void CopyToGpu();
0107   void CopyFromGpu();
0108 #endif
0109
0110   // quick and dirty serialization and deserialization
0111   void ToFile(std::string filename) const
0112   {
0113 #ifdef VECGEOM_USE_INDEXEDNAVSTATES
0114     std::ofstream outfile(filename, std::ios::binary);
0115     outfile.write(reinterpret_cast<const char *>(&fCapacity), sizeof(fCapacity));
0116     outfile.write(reinterpret_cast<const char *>(&fDepth), sizeof(fDepth));
0117     outfile.write(reinterpret_cast<char *>(fBuffer), fCapacity * NavigationState::SizeOfInstanceAlignAware(fDepth));
0118 #else
0119     std::cerr << "serializing pointer based navstates not supported \n";
0120 #endif
0121   }
0122
0123   static void ReadDepthAndCapacityFromFile(std::string filename, int &cap, int &dep)
0124   {
0125     std::ifstream fin(filename, std::ios::binary);
0126     fin.read(reinterpret_cast<char *>(&cap), sizeof(cap));
0127     fin.read(reinterpret_cast<char *>(&dep), sizeof(dep));
0128   }
0129
0130   // return number of elements read or -1 if failure
0131   int FromFile(std::string filename)
0132   {
0133 #ifdef VECGEOM_USE_INDEXEDNAVSTATES
0134     // assumes existing NavStatePool object
0135     decltype(fCapacity) cap;
0136     decltype(fDepth) dep;
0137     std::ifstream fin(filename, std::ios::binary);
0138     if (!fin) return -1;
0139     fin.read(reinterpret_cast<char *>(&cap), sizeof(cap));
0140     if (!fin) return -2;
0141     fin.read(reinterpret_cast<char *>(&dep), sizeof(dep));
0142     if (!fin) return -2;
0143     if (cap != fCapacity || dep != fDepth) std::cerr << " warning: reading from navstate with different size\n";
0144     fin.read(reinterpret_cast<char *>(fBuffer), fCapacity * NavigationState::SizeOfInstanceAlignAware(fDepth));
0145     if (!fin) return -3;
0146 #else
0147     std::cerr << "serializing pointer based navstates not supported \n";
0148 #endif
0149     return fCapacity;
0150   }
0151
0152   VECCORE_ATT_HOST_DEVICE
0153   NavigationState *operator[](int i)
0154   {
0155      return GetNavigationState(i, fBuffer, fDepth);
0156   }
0157
0158   VECCORE_ATT_HOST_DEVICE
0159   NavigationState const *operator[](int i) const
0160   {
0161      return GetNavigationState(i, fBuffer, fDepth);
0162   }
0163
0164   // convert/init this to a plain NavigationState** array
0165   // so that array[0] points to the first state in the NavStatePool, etc
0166   // this method also allocates memory; array should be a nullptr initially
0167   // this is a convenience function
0168   VECCORE_ATT_HOST_DEVICE
0169   void ToPlainPointerArray(NavigationState const **&array) const
0170   {
0171     array = new NavigationState const *[fCapacity];
0172     for (int i = 0; i < fCapacity; ++i) {
0173       array[i] = (*this)[i];
0174     }
0175   }
0176
0177   // dito for the non-const version
0178   VECCORE_ATT_HOST_DEVICE
0179   void ToPlainPointerArray(NavigationState **&array)
0180   {
0181     array = new NavigationState *[fCapacity];
0182     for (int i = 0; i < fCapacity; ++i) {
0183       array[i] = (*this)[i];
0184     }
0185   }
0186
0187   void Print() const
0188   {
0189     for (int i = 0; i < fCapacity; ++i)
0190       (*this)[i]->Print();
0191   }
0192
0193   void *GetGPUPointer() const { return fGPUPointer; }
0194
0195   int capacity() const { return fCapacity; }
0196
0197 private: // protected methods
0198 #ifdef VECGEOM_ENABLE_CUDA
0199   // This constructor used to build NavStatePool at the GPU.  BufferGPU
0200   VECCORE_ATT_DEVICE
0201   NavStatePool(int size, int depth, char *fBufferGPU)
0202       : fCapacity(size), fDepth(depth), fBuffer(fBufferGPU), fGPUPointer(NULL)
0203   {
0204   }
0205 #endif
0206
0207 private:         // members
0208   int fCapacity; // the number of states in the pool
0209   int fDepth;    // depth of the navigation objects to cover
0210   char *fBuffer; // the memory buffer in which we place states
0211
0212   // assume it keeps a GPU pointer directly
0213   // the target of the copy operation
0214   void *fGPUPointer;
0215
0216 }; // end class
0217
0218 // an implementation of the CopyOperation could be as follows
0219 #if !defined(VECCORE_CUDA) && defined(VECGEOM_ENABLE_CUDA)
0220 inline void NavStatePool::CopyToGpu()
0221 {
0222
0223   // modify content temporarily to convert CPU pointers to GPU pointers
0224   NavigationState *state;
0225   for (int i = 0; i < fCapacity; ++i) {
0226     state = operator[](i);
0227     state->ConvertToGPUPointers();
0228   }
0229
0230   // we also have to fix the fPath pointers
0231
0232   // copy
0233   vecgeom::CopyToGpu((void *)fBuffer, fGPUPointer, fCapacity * NavigationState::SizeOfInstanceAlignAware(fDepth));
0234   // CudaAssertError( cudaMemcpy(fGPUPointer, (void*)fBuffer, fCapacity*NavigationState::SizeOf(fDepth),
0235   // cudaMemcpyHostToDevice) );
0236
0237   // modify back pointers
0238   for (int i = 0; i < fCapacity; ++i) {
0239     state = operator[](i);
0240     state->ConvertToCPUPointers();
0241   }
0242
0243   // now some kernel can be launched on GPU side
0244 } // end CopyFunction
0245
0246 inline void NavStatePool::CopyFromGpu()
0247 {
0248   // this does not work
0249   // modify content temporarily to convert CPU pointers to GPU pointers
0250
0251   // std::cerr << "Starting to COPY" << std::endl;
0252   // std::cerr << "GPU pointer " << fGPUPointer << std::endl;
0253   vecgeom::CopyFromGpu(fGPUPointer, (void *)fBuffer, fCapacity * NavigationState::SizeOfInstanceAlignAware(fDepth));
0254
0255   NavigationState *state;
0256   for (int i = 0; i < fCapacity; ++i) {
0257     state = operator[](i);
0258     state->ConvertToCPUPointers();
0259   }
0260 } // end CopyFunction
0261 #endif
0262 } // namespace VECGEOM_IMPL_NAMESPACE
0263 } // namespace vecgeom
0264
0265 #endif /* NAVSTATEPOOL_H_ */