Using OptiX 7.1 with CUDA 10.2, driver 461.09 with GTX 1050Ti
When I run the optixCallablePrograms
from the samples that came with SDK in debug mode, it crashes
Caught exception: CUDA error on synchronize with error ‘an illegal memory access was encountered’ (D:\ProgramData\NVIDIA Corporation\OptiX SDK 7.1.0\SDK\optixCallablePrograms\optixCallablePrograms.cpp:703)
But works fine in release mode.
When I run my sample, it get
[ 2][COMPILE FEEDBACK]: COMPILE ERROR: “__direct_callable__scatterMetallic” not found in programDescriptions[0].callables.moduleDC
Optix call (optixProgramGroupCreate( optixContext, &pgDesc, 1, &pgOptions, log, &sizeofLog, &callablePGs[MetallicMaterial])) failed with code 7001 (line 525)
The module
is the same as the one used in raygen, miss, hitgroup.
void SampleRenderer::createCallablePrograms() {
callablePGs.resize(MaterialTypeCount);
{
OptixProgramGroupOptions pgOptions{};
OptixProgramGroupDesc pgDesc{};
pgDesc.kind = OPTIX_PROGRAM_GROUP_KIND_CALLABLES;
pgDesc.callables.moduleDC = module;
pgDesc.callables.entryFunctionNameDC = "__direct_callable__scatterMetallic";
char log[2048];
size_t sizeofLog = sizeof(log);
OPTIX_CHECK(optixProgramGroupCreate(
optixContext,
&pgDesc,
1,
&pgOptions,
log,
&sizeofLog,
&callablePGs[MetallicMaterial]));
if (sizeofLog > 1) PRINT(log);
}
{
OptixProgramGroupOptions pgOptions{};
OptixProgramGroupDesc pgDesc{};
pgDesc.kind = OPTIX_PROGRAM_GROUP_KIND_CALLABLES;
pgDesc.callables.moduleDC = module;
pgDesc.callables.entryFunctionNameDC = "__direct_callable__scatterLambertian";
char log[2048];
size_t sizeofLog = sizeof(log);
OPTIX_CHECK(optixProgramGroupCreate(
optixContext,
&pgDesc,
1,
&pgOptions,
log,
&sizeofLog,
&callablePGs[LambertianMaterial]));
if (sizeofLog > 1) PRINT(log);
}
}
These are the callable and ordinary kernels/shaders from the .cu file:
static __forceinline__ __device__ bool scatterMetallic(const Ray& inputRay, vec3f& attenuation, Ray& scatteredRay, gdt::LCG<16>& prng, RadiancePRD* prd) {
vec3f reflected = reflect(inputRay.direction, prd->hitNormal);
scatteredRay.origin = prd->hitPoint;
scatteredRay.direction = reflected + prd->material.data.metallic.fuzz * randomInUnitSphere(prng);
attenuation = prd->material.data.metallic.albedo;
//Is the angle between scattered ray and the surface normal acute?
return(0.0f < dot(scatteredRay.direction, prd->hitNormal));
}
static __forceinline__ __device__ bool scatterLambertian(const Ray& inputRay, vec3f& attenuation, Ray& scatteredRay, gdt::LCG<16>& prng, RadiancePRD* prd) {
return true;
}
extern "C" __device__ bool __direct_callable__scatterMetallic(const Ray& inputRay, vec3f& attenuation, Ray& scatteredRay, gdt::LCG<16>& prng, RadiancePRD* prd) {
vec3f reflected = reflect(inputRay.direction, prd->hitNormal);
scatteredRay.origin = prd->hitPoint;
scatteredRay.direction = reflected + prd->material.data.metallic.fuzz * randomInUnitSphere(prng);
attenuation = prd->material.data.metallic.albedo;
//Is the angle between scattered ray and the surface normal acute?
return(0.0f < dot(scatteredRay.direction, prd->hitNormal));
}
extern "C" __device__ bool __direct_callable__scatterLambertian(const Ray & inputRay, vec3f & attenuation, Ray & scatteredRay, gdt::LCG<16> & prng, RadiancePRD * prd) {
return true;
}
For completeness, this is how I invoke (tried both ways-but it doesn’t come this far; fails at compilation for some reason):
__direct_callable__scatterMetallic(Ray{ rayOrigin , rayDirection }, attenuation, scattered, prng, &perRayData)
optixDirectCall<bool, const Ray&, vec3f&, Ray&, gdt::LCG<16>&, RadiancePRD*>(0u, Ray{ rayOrigin , rayDirection }, attenuation, scattered, prng, &perRayData)