mirrors/blender
1
0
Fork 0
mirror of https://projects.blender.org/blender/blender.git synced 2025-01-22 07:22:12 -05:00
Code Issues Projects Releases Packages Wiki Activity

Fix #130641: MetalRT: Motion Blur (render errors)

Browse source 
This PR fixes #130641. The bug was caused by a missing self-object constraint when performing SSS on motion blur scenes. scene_intersect_local tests were erroneously hitting other objects, and out of range primitive IDs were causing spurious downstream behavior.

Pull Request: https://projects.blender.org/blender/blender/pulls/131156
This commit is contained in:
Michael Jones 2024-12-03 20:24:36 +01:00 committed by Thomas Dinges
parent f2d81830bd
commit ab337e0fd8
2 changed files with 37 additions and 6 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

24
intern/cycles/kernel/device/metal/bvh.h
View file

@ -24,10 +24,16 @@ struct MetalRTIntersectionPayload {
struct MetalRTIntersectionLocalPayload_single_hit {
int self_prim;
#if defined(__METALRT_MOTION__)
int self_object;
#endif
};
struct MetalRTIntersectionLocalPayload {
int self_prim;
#if defined(__METALRT_MOTION__)
int self_object;
#endif
uint lcg_state;
uint hit_prim[LOCAL_MAX_HITS];
float hit_t[LOCAL_MAX_HITS];
@ -330,13 +336,11 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
MetalRTIntersectionLocalPayload_single_hit payload;
payload.self_prim = ray->self.prim - primitive_id_offset;
/* We only need custom intersection filtering (i.e. non_opaque) if we are performing a
* self-primitive intersection check. */
metalrt_intersect.force_opacity((ray->self.prim == PRIM_NONE) ?
metal::raytracing::forced_opacity::opaque :
metal::raytracing::forced_opacity::non_opaque);
# if defined(__METALRT_MOTION__)
/* We can't skip over the top-level BVH in the motion blur case, so still need to do
* the self-object check. */
payload.self_object = local_object;
metalrt_intersect.force_opacity(metal::raytracing::forced_opacity::non_opaque);
intersection = metalrt_intersect.intersect(r,
metal_ancillaries->accel_struct,
~0,
@ -344,6 +348,11 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
metal_ancillaries->ift_local_single_hit_mblur,
payload);
# else
/* We only need custom intersection filtering (i.e. non_opaque) if we are performing a
* self-primitive intersection check. */
metalrt_intersect.force_opacity((ray->self.prim == PRIM_NONE) ?
metal::raytracing::forced_opacity::opaque :
metal::raytracing::forced_opacity::non_opaque);
intersection = metalrt_intersect.intersect(
r,
metal_ancillaries->blas_accel_structs[local_object].blas,
@ -387,6 +396,9 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
metalrt_intersect.force_opacity(metal::raytracing::forced_opacity::non_opaque);
# if defined(__METALRT_MOTION__)
/* We can't skip over the top-level BVH in the motion blur case, so still need to do
* the self-object check. */
payload.self_object = local_object;
intersection = metalrt_intersect.intersect(r,
metal_ancillaries->accel_struct,
~0,

19
intern/cycles/kernel/device/metal/kernel.metal
View file

@ -54,11 +54,18 @@ __intersection__local_tri_single_hit(
extended_limits)]] PrimitiveIntersectionResult
__intersection__local_tri_single_hit_mblur(
ray_data MetalKernelContext::MetalRTIntersectionLocalPayload_single_hit &payload [[payload]],
# if defined(__METALRT_MOTION__)
uint object [[instance_id]],
# endif
uint primitive_id [[primitive_id]])
{
PrimitiveIntersectionResult result;
result.continue_search = true;
# if defined(__METALRT_MOTION__)
result.accept = (payload.self_prim != primitive_id) && (payload.self_object == object);
# else
result.accept = (payload.self_prim != primitive_id);
# endif
return result;
}
@ -156,9 +163,21 @@ __intersection__local_tri(ray_data MetalKernelContext::MetalRTIntersectionLocalP
__intersection__local_tri_mblur(
ray_data MetalKernelContext::MetalRTIntersectionLocalPayload &payload [[payload]],
uint primitive_id [[primitive_id]],
# if defined(__METALRT_MOTION__)
uint object [[instance_id]],
# endif
float2 barycentrics [[barycentric_coord]],
float ray_tmax [[distance]])
{
# if defined(__METALRT_MOTION__)
if (payload.self_object != object) {
PrimitiveIntersectionResult result;
result.continue_search = true;
result.accept = false;
return result;
}
# endif
return metalrt_local_hit<PrimitiveIntersectionResult, METALRT_HIT_TRIANGLE>(
payload, primitive_id, barycentrics, ray_tmax);
}