griefith/test2
1
0
Fork 0
Code Issues Pull Requests Actions 1 Packages Projects Releases Wiki Activity

Fix Cycles rendering with OptiX after instance limit increase when building with old SDK

Browse Source 
Commit d259e7dcfb increased the instance limit, but only provided
a fall back for the host code for older OptiX SDKs, not for kernel code. This caused a mismatch when
an old SDK was used (as is currently the case on buildbot) and subsequent rendering artifacts. This
fixes that by moving the bit that is checked to a common location that works with both old an new
SDK versions.
This commit is contained in:
Patrick Mours
2021-01-08 13:38:26 +01:00
parent e3ae7d1f2f
commit c66f00dc26
2 changed files with 19 additions and 21 deletions
Download Patch File Download Diff File Expand all files Collapse all files

27
intern/cycles/device/device_optix.cpp
View File

@@ -1514,16 +1514,19 @@ class OptiXDevice : public CUDADevice {
}
else {
unsigned int num_instances = 0;
unsigned int max_num_instances = 0xFFFFFFFF;
bvh_optix->as_data.free();
bvh_optix->traversable_handle = 0;
bvh_optix->motion_transform_data.free();
# if OPTIX_ABI_VERSION < 23
if (bvh->objects.size() > 0x7FFFFF) {
# else
if (bvh->objects.size() > 0x7FFFFFF) {
# endif
optixDeviceContextGetProperty(context,
OPTIX_DEVICE_PROPERTY_LIMIT_MAX_INSTANCE_ID,
&max_num_instances,
sizeof(max_num_instances));
// Do not count first bit, which is used to distinguish instanced and non-instanced objects
max_num_instances >>= 1;
if (bvh->objects.size() > max_num_instances) {
progress.set_error(
"Failed to build OptiX acceleration structure because there are too many instances");
return;
@@ -1582,8 +1585,8 @@ class OptiXDevice : public CUDADevice {
instance.transform[5] = 1.0f;
instance.transform[10] = 1.0f;
// Set user instance ID to object index
instance.instanceId = ob->get_device_index();
// Set user instance ID to object index (but leave low bit blank)
instance.instanceId = ob->get_device_index() << 1;
// Have to have at least one bit in the mask, or else instance would always be culled
instance.visibilityMask = 1;
@@ -1689,13 +1692,9 @@ class OptiXDevice : public CUDADevice {
else {
// Disable instance transform if geometry already has it applied to vertex data
instance.flags = OPTIX_INSTANCE_FLAG_DISABLE_TRANSFORM;
// Non-instanced objects read ID from prim_object, so
// distinguish them from instanced objects with high bit set
# if OPTIX_ABI_VERSION < 23
instance.instanceId |= 0x800000;
# else
instance.instanceId |= 0x8000000;
# endif
// Non-instanced objects read ID from 'prim_object', so distinguish
// them from instanced objects with the low bit set
instance.instanceId |= 1;
}
}
}

13
intern/cycles/kernel/kernels/optix/kernel_optix.cu
View File

@@ -45,13 +45,12 @@ template<bool always = false> ccl_device_forceinline uint get_object_id()
uint object = optixGetInstanceId();
#endif
// Choose between always returning object ID or only for instances
if (always)
// Can just remove the high bit since instance always contains object ID
return object & 0x7FFFFFF; // OPTIX_ABI_VERSION >= 23 ? 0x7FFFFFF : 0x7FFFFF
// Set to OBJECT_NONE if this is not an instanced object
else if (object & 0x8000000) // OPTIX_ABI_VERSION >= 23 ? 0x8000000 : 0x800000
object = OBJECT_NONE;
return object;
if (always || (object & 1) == 0)
// Can just remove the low bit since instance always contains object ID
return object >> 1;
else
// Set to OBJECT_NONE if this is not an instanced object
return OBJECT_NONE;
}
extern "C" __global__ void __raygen__kernel_optix_path_trace()