eic-opticks/sysrap/SOPTIX_Accel.h

0001 #pragma once
0002 /**
0003 SOPTIX_Accel.h : builds acceleration structure GAS or IAS from the buildInputs
0004 ===============================================================================
0005
0006 Used by::
0007
0008     SBT::createGAS
0009     SOPTIX_MeshGroup.h for GAS
0010     SOPTIX_Scene.h for IAS
0011
0012 Q: Is the buildInputs reference going to stay valid ?
0013 A: Each SOPTIX_MeshGroup holds the buildInputs vector for GAS
0014 A: SOPTIX_Scene.h (used for triangulated rendering) holds buildInputs vector for IAS, with single entry
0015
0016 **/
0017
0018 #include "SOPTIX_Desc.h"
0019 #include "SOPTIX_BuildInput.h"
0020 #include "SOPTIX_Options.h"
0021
0022
0023 struct SOPTIX_Accel
0024 {
0025     static int Initialize();
0026     int irc ;
0027     CUdeviceptr buffer ;
0028     OptixTraversableHandle handle ;
0029
0030     OptixBuildFlags buildFlags = {} ;
0031     OptixAccelBufferSizes accelBufferSizes = {} ;
0032     size_t compacted_size ;
0033     bool   compacted ;
0034
0035     std::vector<const SOPTIX_BuildInput*> bis ;
0036     std::vector<OptixBuildInput> buildInputs ;
0037
0038     std::string desc() const ;
0039     static SOPTIX_Accel* Create(OptixDeviceContext& context, const SOPTIX_BuildInput* _bi );
0040     static SOPTIX_Accel* Create(OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis );
0041 private:
0042     SOPTIX_Accel( OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis );
0043     void init(    OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis );
0044
0045
0046 };
0047
0048
0049 inline int SOPTIX_Accel::Initialize()
0050 {
0051     if(SOPTIX_Options::Level()>0) std::cout << "-SOPTIX_Accel::Initialize\n" ;
0052     return 0 ;
0053 }
0054
0055
0056 inline std::string SOPTIX_Accel::desc() const
0057 {
0058     std::stringstream ss ;
0059     ss << "[SOPTIX_Accel::desc\n"
0060        << " buildInputs.size " << buildInputs.size() << "\n"
0061        << " compacted " << ( compacted ? "YES" : "NO " ) << "\n"
0062        << " compacted_size " << compacted_size << "\n"
0063        << SOPTIX_Desc::AccelBufferSizes(accelBufferSizes) << "\n"
0064        << "]SOPTIX_Accel::desc\n"
0065        ;
0066     std::string str = ss.str();
0067     return str ;
0068 }
0069
0070
0071 inline SOPTIX_Accel* SOPTIX_Accel::Create( OptixDeviceContext& context, const SOPTIX_BuildInput* bi )
0072 {
0073     std::vector<const SOPTIX_BuildInput*> _bis ;
0074     _bis.push_back(bi);
0075     return new SOPTIX_Accel( context, _bis  );
0076 }
0077
0078 inline SOPTIX_Accel* SOPTIX_Accel::Create( OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis )
0079 {
0080     return new SOPTIX_Accel( context, _bis  );
0081 }
0082
0083 /**
0084 SOPTIX_Accel::SOPTIX_Accel
0085 ---------------------------
0086
0087 1. collect SOPTIX_BuildInput:bi into bis and bi->buildInput into buildInputs.
0088    The names of all the SOPTIX_BuildInput in the bis vector are required to match.
0089
0090 2. create the acceleration structure from the buildInputs
0091
0092 **/
0093
0094
0095 inline SOPTIX_Accel::SOPTIX_Accel( OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis )
0096     :
0097     irc(Initialize()),
0098     buffer(0),
0099     handle(0),
0100     compacted_size(0),
0101     compacted(false)
0102 {
0103     init(context, _bis);
0104 }
0105
0106 inline void SOPTIX_Accel::init( OptixDeviceContext& context, const std::vector<const SOPTIX_BuildInput*>& _bis )
0107 {
0108     if(SOPTIX_Options::Level()>0) std::cout << "[SOPTIX_Accel::init\n" ;
0109
0110     const char* name0 = nullptr ;
0111
0112     for(unsigned i=0 ; i < _bis.size() ; i++)
0113     {
0114         const SOPTIX_BuildInput* bi = _bis[i] ;
0115         bis.push_back(bi);
0116         buildInputs.push_back(bi->buildInput) ;
0117
0118         if(i == 0)
0119         {
0120             name0 = bi->name ;
0121         }
0122         else
0123         {
0124             assert( strcmp(bi->name, name0) == 0 );
0125         }
0126     }
0127
0128     OptixAccelBuildOptions accel_options = {};
0129
0130     unsigned _buildFlags = 0 ;
0131     _buildFlags |= OPTIX_BUILD_FLAG_PREFER_FAST_TRACE  ;
0132     _buildFlags |= OPTIX_BUILD_FLAG_ALLOW_COMPACTION ;
0133     //_buildFlags |= OPTIX_BUILD_FLAG_ALLOW_RANDOM_VERTEX_ACCESS ; // see optixGetTriangleVertexData
0134     //_buildFlags |= OPTIX_BUILD_FLAG_ALLOW_RANDOM_INSTANCE_ACCESS ; // see optixGetInstanceTraversableFromIAS
0135
0136     buildFlags = (OptixBuildFlags)_buildFlags ;
0137
0138     OptixBuildOperation operation = OPTIX_BUILD_OPERATION_BUILD ;
0139     OptixMotionOptions motionOptions = {} ;
0140
0141     accel_options.buildFlags = buildFlags ;
0142     accel_options.operation  = operation ;
0143     accel_options.motionOptions = motionOptions ;
0144
0145     OPTIX_CHECK( optixAccelComputeMemoryUsage( context,
0146                                                &accel_options,
0147                                                buildInputs.data(),
0148                                                buildInputs.size(),
0149                                                &accelBufferSizes
0150                                              ) );
0151
0152
0153
0154     CUstream stream = 0 ;
0155     const OptixAccelBuildOptions* accelOptions = &accel_options ;
0156
0157     CUdeviceptr tempBuffer ;
0158     CUDA_CHECK( cudaMalloc(
0159                 reinterpret_cast<void**>( &tempBuffer ),
0160                 accelBufferSizes.tempSizeInBytes
0161                 ) );
0162
0163     size_t      compactedSizeOffset = roundUp<size_t>( accelBufferSizes.outputSizeInBytes, 8ull );
0164
0165     // expand the outputBuffer size by 8 bytes plus any 8-byte alignment padding
0166     // to give somewhere to write the emitted property
0167
0168     CUdeviceptr outputBuffer ;
0169     CUDA_CHECK( cudaMalloc(
0170                 reinterpret_cast<void**>( &outputBuffer),
0171                 compactedSizeOffset + 8
0172                 ) );
0173
0174
0175     OptixAccelEmitDesc emitProperty = {};
0176     emitProperty.type    = OPTIX_PROPERTY_TYPE_COMPACTED_SIZE;
0177     emitProperty.result  = ( CUdeviceptr )( (char*)outputBuffer + compactedSizeOffset );
0178     unsigned numEmittedProperties = 1 ;
0179
0180     if(SOPTIX_Options::Level()>0) std::cout << "-[SOPTIX_Accel::init.optixAccelBuild\n" ;
0181
0182     OPTIX_CHECK( optixAccelBuild( context,
0183                                   stream,
0184                                   accelOptions,
0185                                   buildInputs.data(),
0186                                   buildInputs.size(),
0187                                   tempBuffer,
0188                                   accelBufferSizes.tempSizeInBytes,
0189                                   outputBuffer,
0190                                   accelBufferSizes.outputSizeInBytes,
0191                                   &handle,
0192                                   &emitProperty,
0193                                   numEmittedProperties
0194                                   ) );
0195
0196     if(SOPTIX_Options::Level()>0) std::cout << "-]SOPTIX_Accel::init.optixAccelBuild\n" ;
0197
0198     CUDA_CHECK( cudaFree( (void*)tempBuffer ) );
0199
0200     CUDA_CHECK( cudaMemcpy( &compacted_size, (void*)emitProperty.result, sizeof(size_t), cudaMemcpyDeviceToHost ) );
0201
0202     if( compacted_size < accelBufferSizes.outputSizeInBytes )
0203     {
0204         compacted = true ;
0205         CUDA_CHECK( cudaMalloc( reinterpret_cast<void**>( &buffer ), compacted_size ) );
0206
0207         // use handle as input and output
0208         OPTIX_CHECK( optixAccelCompact( context,
0209                                         stream,
0210                                         handle,
0211                                         buffer,
0212                                         compacted_size,
0213                                         &handle ) );
0214
0215         CUDA_CHECK( cudaFree( (void*)outputBuffer ) );
0216     }
0217     else
0218     {
0219         compacted = false ;
0220         buffer = outputBuffer ;
0221     }
0222     if(SOPTIX_Options::Level()>0) std::cout << "]SOPTIX_Accel::init\n" ;
0223 }
0224