eic-opticks/sysrap/SOPTIX_SBT.h

0001 #pragma once
0002 /**
0003 SOPTIX_SBT.h : create sbt from pipeline and scene by uploding the prog and hitgroup records
0004 ============================================================================================
0005
0006 OptixShaderBindingTable binds together geometry, "shader" programs, records with program data
0007
0008 Used for example from::
0009
0010    sysrap/tests/SOPTIX_Scene_test.cc
0011    sysrap/tests/SGLFW_SOPTIX_Scene_test.cc
0012
0013 Good general explanatiom of SBT Shader Binding Table
0014
0015 * https://www.willusher.io/graphics/2019/11/20/the-sbt-three-ways
0016
0017 **/
0018
0019 #include "ssys.h"
0020 #include "SOPTIX_Binding.h"
0021
0022 struct SOPTIX_SBT
0023 {
0024     SOPTIX_Pipeline& pip ;
0025     SOPTIX_Scene& scn ;
0026
0027     std::vector<SOPTIX_HitgroupRecord> hitgroup_records;
0028     OptixShaderBindingTable sbt = {};
0029
0030     SOPTIX_SBT( SOPTIX_Pipeline& pip, SOPTIX_Scene& scn );
0031
0032     void init();
0033     void initRaygen();
0034     void initMiss();
0035     void initHitgroup();
0036     static constexpr const char* __initHitgroup_DUMP = "SOPTIX_SBT__initHitgroup_DUMP" ;
0037
0038     std::string desc() const ;
0039     std::string descPartBI() const ;
0040
0041     static std::string Desc(const OptixShaderBindingTable& sbt) ;
0042
0043 };
0044
0045
0046 inline SOPTIX_SBT::SOPTIX_SBT( SOPTIX_Pipeline& _pip, SOPTIX_Scene& _scn )
0047     :
0048     pip(_pip),
0049     scn(_scn)
0050 {
0051     init();
0052 }
0053
0054 inline void SOPTIX_SBT::init()
0055 {
0056     initRaygen();
0057     initMiss();
0058     initHitgroup();
0059 }
0060
0061 inline void SOPTIX_SBT::initRaygen()
0062 {
0063     CUdeviceptr  raygen_record ;
0064
0065     const size_t raygen_record_size = sizeof( SOPTIX_RaygenRecord );
0066     CUDA_CHECK( cudaMalloc( reinterpret_cast<void**>( &raygen_record ), raygen_record_size ) );
0067
0068     SOPTIX_RaygenRecord rg_sbt;
0069     OPTIX_CHECK( optixSbtRecordPackHeader( pip.raygen_pg, &rg_sbt ) );
0070     //SOPTIX_RaygenData& data = rg_sbt.data
0071
0072     CUDA_CHECK( cudaMemcpy(
0073                 reinterpret_cast<void*>(raygen_record),
0074                 &rg_sbt,
0075                 raygen_record_size,
0076                 cudaMemcpyHostToDevice
0077                 ) );
0078
0079     sbt.raygenRecord = raygen_record ;
0080 }
0081
0082 inline void SOPTIX_SBT::initMiss()
0083 {
0084     CUdeviceptr miss_record ;
0085
0086     const size_t miss_record_size = sizeof( SOPTIX_MissRecord );
0087     CUDA_CHECK( cudaMalloc( reinterpret_cast<void**>( &miss_record ), miss_record_size ) );
0088
0089     SOPTIX_MissRecord ms_sbt;
0090     OPTIX_CHECK( optixSbtRecordPackHeader( pip.miss_pg, &ms_sbt ) );
0091     SOPTIX_MissData& data = ms_sbt.data ;
0092     //data.bg_color = make_float3( 0.3f, 0.1f, 0.f ) ;
0093     //data.bg_color = make_float3( 1.0f, 0.0f, 0.f ) ;  // after diddling becomes nasty pink
0094     data.bg_color = make_float3( 0.0f, 0.0f, 0.f ) ;
0095
0096     CUDA_CHECK( cudaMemcpy(
0097                 reinterpret_cast<void*>( miss_record ),
0098                 &ms_sbt,
0099                 miss_record_size,
0100                 cudaMemcpyHostToDevice
0101                 ) );
0102
0103     sbt.missRecordBase = miss_record;
0104     sbt.missRecordStrideInBytes = static_cast<uint32_t>( miss_record_size );
0105     sbt.missRecordCount = 1 ;
0106 }
0107
0108
0109 /**
0110 SOPTIX_SBT::initHitgroup
0111 -------------------------
0112
0113 HMM: with analytic geometry have "boundary" that
0114 comes from the CSGNode. To do that with triangles
0115 need to plant the boundary indices into HitgroupData.
0116 That means need hitgroup records for each sub-SMesh
0117 (thats equivalent to each CSGPrim)
0118
0119 TODO: check this labelling
0120
0121 Need nested loop like CSGOptiX/SBT.cc SBT::createHitgroup::
0122
0123      GAS
0124         BuildInput       (actually 1:1 with GAS for analytic)
0125            sub-SMesh
0126
0127 So need access to scene data to form the SBT
0128
0129 NB this uses CPU/GPU types defined in SOPTIX_Binding.h
0130
0131 **/
0132
0133 inline void SOPTIX_SBT::initHitgroup()
0134 {
0135     bool DUMP = ssys::getenvbool(__initHitgroup_DUMP) ;
0136     if(DUMP) std::cout
0137          << "SOPTIX_SBT::initHitgroup "
0138          << __initHitgroup_DUMP << " : " << ( DUMP ? "YES" : "NO " )
0139          << descPartBI()
0140          << std::endl
0141          ;
0142
0143     size_t num_mg = scn.meshgroup.size();  // SOPTIX_Scene
0144     for(size_t i=0 ; i < num_mg ; i++)
0145     {
0146         SOPTIX_MeshGroup* xmg = scn.meshgroup[i] ;
0147         size_t num_bi = xmg->num_buildInputs()  ;
0148         const SCUDA_MeshGroup* cmg = xmg->cmg ;
0149         size_t num_part = cmg->num_part()  ;
0150         assert( num_part == num_bi );
0151         size_t edge = 20 ;
0152
0153         for(size_t j=0 ; j < num_bi ; j++)
0154         {
0155             const SOPTIX_BuildInput* bi = xmg->bis[j] ;
0156             assert( bi->is_BuildInputTriangleArray() );
0157             unsigned numSbtRecords = bi->numSbtRecords();
0158
0159             bool at_edge = ( j < edge || j > (num_bi - edge)) ;
0160             if(at_edge && DUMP ) std::cout
0161                 << "SOPTIX_SBT::initHitgroup"
0162                 << " i " << i
0163                 << " num_mg " << num_mg
0164                 << " j " << j
0165                 << " num_bi " << num_bi
0166                 << " numSbtRecords " << numSbtRecords
0167                 << " edge " << edge
0168                 << "\n"
0169                 ;
0170             assert( numSbtRecords == 1 );
0171
0172             SOPTIX_HitgroupRecord hg_sbt;
0173             OPTIX_CHECK( optixSbtRecordPackHeader( pip.hitgroup_pg, &hg_sbt ) );
0174             SOPTIX_HitgroupData& data = hg_sbt.data ;
0175             SOPTIX_TriMesh& trimesh = data.mesh ;
0176
0177             trimesh.vertex = reinterpret_cast<float3*>( cmg->vtx.pointer(j) );
0178             trimesh.normal = reinterpret_cast<float3*>( cmg->nrm.pointer(j) );
0179             trimesh.indice = reinterpret_cast<uint3*>(  cmg->idx.pointer(j) );
0180
0181             hitgroup_records.push_back(hg_sbt);
0182         }
0183     }
0184
0185     CUdeviceptr hitgroup_record_base ;
0186     const size_t hitgroup_record_size = sizeof( SOPTIX_HitgroupRecord );
0187     const size_t hitgroup_record_bytes = hitgroup_record_size*hitgroup_records.size() ;
0188
0189     CUDA_CHECK( cudaMalloc(
0190                 reinterpret_cast<void**>( &hitgroup_record_base ),
0191                 hitgroup_record_bytes ) );
0192
0193     CUDA_CHECK( cudaMemcpy(
0194                 reinterpret_cast<void*>( hitgroup_record_base ),
0195                 hitgroup_records.data(),
0196                 hitgroup_record_bytes,
0197                 cudaMemcpyHostToDevice
0198                 ) );
0199
0200     sbt.hitgroupRecordBase = hitgroup_record_base ;
0201     sbt.hitgroupRecordStrideInBytes = static_cast<uint32_t>( hitgroup_record_size );
0202     sbt.hitgroupRecordCount         = static_cast<uint32_t>( hitgroup_records.size() );
0203 }
0204
0205
0206 inline std::string SOPTIX_SBT::desc() const
0207 {
0208     std::stringstream ss ;
0209     ss << "[SOPTIX_SBT::desc\n" ;
0210     ss << " hitgroup_records.size " << hitgroup_records.size() << "\n" ;
0211     ss << Desc(sbt) ;
0212     ss << "]SOPTIX_SBT::desc\n" ;
0213     std::string str = ss.str();
0214     return str ;
0215 }
0216
0217 inline std::string SOPTIX_SBT::descPartBI() const
0218 {
0219     std::stringstream ss ;
0220     ss << "[SOPTIX_SBT::descPartBI\n" ;
0221     size_t num_mg = scn.meshgroup.size();
0222     for(size_t i=0 ; i < num_mg ; i++)
0223     {
0224         SOPTIX_MeshGroup* xmg = scn.meshgroup[i] ;
0225         size_t num_bi = xmg->num_buildInputs()  ;
0226         const SCUDA_MeshGroup* cmg = xmg->cmg ;
0227         size_t num_part = cmg->num_part()  ;
0228         assert( num_part == num_bi );
0229
0230         ss
0231             << std::setw(4) << i
0232             << " SCUDA_MeshGroup::num_part " << std::setw(6) << num_part
0233             << " SOPTIX_MeshGroup::num_bi  " << std::setw(6) << num_bi
0234             << "\n"
0235             ;
0236     }
0237     ss << "]SOPTIX_SBT::descPartBI\n" ;
0238     std::string str = ss.str();
0239     return str ;
0240 }
0241
0242
0243 inline std::string SOPTIX_SBT::Desc(const OptixShaderBindingTable& sbt)  // static
0244 {
0245     std::stringstream ss ;
0246     ss
0247         << " sbt.raygenRecord                " << sbt.raygenRecord << "\n"
0248         << " sbt.missRecordBase              " << sbt.missRecordBase  << "\n"
0249         << " sbt.missRecordStrideInBytes     " << sbt.missRecordStrideInBytes << "\n"
0250         << " sbt.missRecordCount             " << sbt.missRecordCount << "\n"
0251         << " sbt.hitgroupRecordBase          " << sbt.hitgroupRecordBase << "\n"
0252         << " sbt.hitgroupRecordStrideInBytes " << sbt.hitgroupRecordStrideInBytes << "\n"
0253         << " sbt.hitgroupRecordCount         " << sbt.hitgroupRecordCount << "\n"
0254        ;
0255     std::string str = ss.str();
0256     return str ;
0257 }
0258
0259