diff options
author | Patrick Mours <pmours@nvidia.com> | 2019-11-22 19:30:22 +0300 |
---|---|---|
committer | Patrick Mours <pmours@nvidia.com> | 2019-11-22 19:30:22 +0300 |
commit | 59aef0ad5d6440e855f53db791b28076f50df2b2 (patch) | |
tree | 87879e85d7b56c4e502c26d89e917c590b3a27db /intern/cycles/kernel/kernels | |
parent | aadbb794cd6ee757c0b62c4735fec5ec696b02a9 (diff) |
Fix T71255: Particle hair not showing in viewport with OptiX after scaling
The OptiX intersection program for curves uses "optixGetObjectRayDirection"
to get the ray direction in object space (which was inverse transformed
with the current transformation matrix). OptiX does no additional operations
on it, so if there is a scaling transform, the direction is not normalized.
But the curve intersection routine expects that. In addition, the distances
used in "optixGetRayTmax()" and "optixReportIntersection()" are in world
space, so need to adjust them accordingly.
Diffstat (limited to 'intern/cycles/kernel/kernels')
-rw-r--r-- | intern/cycles/kernel/kernels/optix/kernel_optix.cu | 15 |
1 files changed, 11 insertions, 4 deletions
diff --git a/intern/cycles/kernel/kernels/optix/kernel_optix.cu b/intern/cycles/kernel/kernels/optix/kernel_optix.cu index c7223a49d79..e03504316ad 100644 --- a/intern/cycles/kernel/kernels/optix/kernel_optix.cu +++ b/intern/cycles/kernel/kernels/optix/kernel_optix.cu @@ -44,7 +44,7 @@ template<bool always = false> ccl_device_forceinline uint get_object_id() #endif // Choose between always returning object ID or only for instances if (always) - // Can just remove the high bit since instace always contains object ID + // Can just remove the high bit since instance always contains object ID return object & 0x7FFFFF; // Set to OBJECT_NONE if this is not an instanced object else if (object & 0x800000) @@ -263,8 +263,12 @@ extern "C" __global__ void __intersection__curve() const uint type = kernel_tex_fetch(__prim_type, prim); const uint visibility = optixGetPayload_4(); - const float3 P = optixGetObjectRayOrigin(); - const float3 dir = optixGetObjectRayDirection(); + float3 P = optixGetObjectRayOrigin(); + float3 dir = optixGetObjectRayDirection(); + + // The direction is not normalized by default, but the curve intersection routine expects that + float len; + dir = normalize_len(dir, &len); # ifdef __OBJECT_MOTION__ const float time = optixGetRayTime(); @@ -274,11 +278,14 @@ extern "C" __global__ void __intersection__curve() Intersection isect; isect.t = optixGetRayTmax(); + // Transform maximum distance into object space + if (isect.t != FLT_MAX) + isect.t *= len; if (!(kernel_data.curve.curveflags & CURVE_KN_INTERPOLATE) ? curve_intersect(NULL, &isect, P, dir, visibility, object, prim, time, type) : cardinal_curve_intersect(NULL, &isect, P, dir, visibility, object, prim, time, type)) { - optixReportIntersection(isect.t, + optixReportIntersection(isect.t / len, type & PRIMITIVE_ALL, __float_as_int(isect.u), // Attribute_0 __float_as_int(isect.v)); // Attribute_1 |