blender
/
blender
136
406
Fork 589
Code Issues 6.2k Pull Requests 751 Projects 19 Wiki Activity

Fix T99201: Cycles render difference with 3D hair curves between OptiX and Emrbee 

It should consistently use the Cycles pirmitive ID for self intersection detection,
not the one from the OptiX or Embree acceleration structure.

Differential Revision: https://developer.blender.org/D15632
This commit is contained in:
Brecht Van Lommel 2022-08-05 14:35:39 +02:00
parent 45f483681f
commit fa514564b0
Notes: blender-bot 2023-05-22 12:40:41 +02:00 
Referenced by issue #99201, Cycles hair rendering tiling artifacts when using OptiX+CPU render
3 changed files with 54 additions and 23 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

34
intern/cycles/kernel/device/cpu/bvh.h
View File

@ -114,27 +114,37 @@ ccl_device_inline bool kernel_embree_is_self_intersection(const KernelGlobals kg
const RTCHit *hit,
const Ray *ray)
{
bool status = false;
int object, prim;
if (hit->instID[0] != RTC_INVALID_GEOMETRY_ID) {
const int oID = hit->instID[0] / 2;
if ((ray->self.object == oID) || (ray->self.light_object == oID)) {
object = hit->instID[0] / 2;
if ((ray->self.object == object) || (ray->self.light_object == object)) {
RTCScene inst_scene = (RTCScene)rtcGetGeometryUserData(
rtcGetGeometry(kernel_data.device_bvh, hit->instID[0]));
const int pID = hit->primID +
(intptr_t)rtcGetGeometryUserData(rtcGetGeometry(inst_scene, hit->geomID));
status = intersection_skip_self_shadow(ray->self, oID, pID);
prim = hit->primID +
(intptr_t)rtcGetGeometryUserData(rtcGetGeometry(inst_scene, hit->geomID));
}
else {
return false;
}
}
else {
const int oID = hit->geomID / 2;
if ((ray->self.object == oID) || (ray->self.light_object == oID)) {
const int pID = hit->primID + (intptr_t)rtcGetGeometryUserData(
rtcGetGeometry(kernel_data.device_bvh, hit->geomID));
status = intersection_skip_self_shadow(ray->self, oID, pID);
object = hit->geomID / 2;
if ((ray->self.object == object) || (ray->self.light_object == object)) {
prim = hit->primID +
(intptr_t)rtcGetGeometryUserData(rtcGetGeometry(kernel_data.device_bvh, hit->geomID));
}
else {
return false;
}
}
return status;
const bool is_hair = hit->geomID & 1;
if (is_hair) {
prim = kernel_data_fetch(curve_segments, prim).prim;
}
return intersection_skip_self_shadow(ray->self, object, prim);
}
ccl_device_inline void kernel_embree_convert_hit(KernelGlobals kg,

18
intern/cycles/kernel/device/metal/kernel.metal
View File

@ -180,11 +180,6 @@ bool metalrt_shadow_all_hit(constant KernelParamsMetal &launch_params_metal,
}
# endif
if (intersection_skip_self_shadow(payload.self, object, prim)) {
/* continue search */
return true;
}
const float u = barycentrics.x;
const float v = barycentrics.y;
int type = 0;
@ -205,6 +200,11 @@ bool metalrt_shadow_all_hit(constant KernelParamsMetal &launch_params_metal,
}
# endif
if (intersection_skip_self_shadow(payload.self, object, prim)) {
/* continue search */
return true;
}
# ifndef __TRANSPARENT_SHADOWS__
/* No transparent shadows support compiled in, make opaque. */
payload.result = true;
@ -346,6 +346,14 @@ inline TReturnType metalrt_visibility_test(
}
#endif
if (intersection_type == METALRT_HIT_TRIANGLE) {
}
# ifdef __HAIR__
else {
prim = kernel_data_fetch(curve_segments, prim).prim;
}
# endif
/* Shadow ray early termination. */
if (visibility & PATH_RAY_SHADOW_OPAQUE) {
if (intersection_skip_self_shadow(payload.self, object, prim)) {

25
intern/cycles/kernel/device/optix/bvh.h
View File

@ -143,14 +143,10 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
}
# endif
ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
if (intersection_skip_self_shadow(ray->self, object, prim)) {
return optixIgnoreIntersection();
}
float u = 0.0f, v = 0.0f;
int type = 0;
if (optixIsTriangleHit()) {
/* Triangle. */
const float2 barycentrics = optixGetTriangleBarycentrics();
u = barycentrics.x;
v = barycentrics.y;
@ -158,6 +154,7 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
}
# ifdef __HAIR__
else if ((optixGetHitKind() & (~PRIMITIVE_MOTION)) != PRIMITIVE_POINT) {
/* Curve. */
u = __uint_as_float(optixGetAttribute_0());
v = __uint_as_float(optixGetAttribute_1());
@ -174,11 +171,17 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
}
# endif
else {
/* Point. */
type = kernel_data_fetch(objects, object).primitive_type;
u = 0.0f;
v = 0.0f;
}
ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
if (intersection_skip_self_shadow(ray->self, object, prim)) {
return optixIgnoreIntersection();
}
# ifndef __TRANSPARENT_SHADOWS__
/* No transparent shadows support compiled in, make opaque. */
optixSetPayload_5(true);
@ -307,7 +310,17 @@ extern "C" __global__ void __anyhit__kernel_optix_visibility_test()
}
#endif
const int prim = optixGetPrimitiveIndex();
int prim = optixGetPrimitiveIndex();
if (optixIsTriangleHit()) {
/* Triangle. */
}
#ifdef __HAIR__
else if ((optixGetHitKind() & (~PRIMITIVE_MOTION)) != PRIMITIVE_POINT) {
/* Curve. */
prim = kernel_data_fetch(curve_segments, prim).prim;
}
#endif
ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
if (visibility & PATH_RAY_SHADOW_OPAQUE) {