[Bf-blender-cvs] [d1f8f8b4a4b] cycles-x: Fix Cycles X use of uninitialized memory for AO and bevel nodes
Brecht Van Lommel
noreply at git.blender.org
Thu Sep 9 21:22:07 CEST 2021
Commit: d1f8f8b4a4bdeead3d6b36167b7804cd0825bd4d
Author: Brecht Van Lommel
Date: Thu Sep 9 21:14:18 2021 +0200
Branches: cycles-x
https://developer.blender.org/rBd1f8f8b4a4bdeead3d6b36167b7804cd0825bd4d
Fix Cycles X use of uninitialized memory for AO and bevel nodes
Always write to the SVM stack also in cases where no actual AO or bevel
ray-tracing can be performed.
===================================================================
M intern/cycles/kernel/svm/svm.h
M intern/cycles/kernel/svm/svm_ao.h
M intern/cycles/kernel/svm/svm_bevel.h
===================================================================
diff --git a/intern/cycles/kernel/svm/svm.h b/intern/cycles/kernel/svm/svm.h
index 35d7cc96c59..4aee1ef11b3 100644
--- a/intern/cycles/kernel/svm/svm.h
+++ b/intern/cycles/kernel/svm/svm.h
@@ -545,14 +545,10 @@ ccl_device void svm_eval_nodes(INTEGRATOR_STATE_CONST_ARGS,
break;
#ifdef __SHADER_RAYTRACE__
case NODE_BEVEL:
- if (KERNEL_NODES_FEATURE(RAYTRACE)) {
- svm_node_bevel(INTEGRATOR_STATE_PASS, sd, stack, node);
- }
+ svm_node_bevel<node_feature_mask>(INTEGRATOR_STATE_PASS, sd, stack, node);
break;
case NODE_AMBIENT_OCCLUSION:
- if (KERNEL_NODES_FEATURE(RAYTRACE)) {
- svm_node_ao(INTEGRATOR_STATE_PASS, sd, stack, node);
- }
+ svm_node_ao<node_feature_mask>(INTEGRATOR_STATE_PASS, sd, stack, node);
break;
#endif
diff --git a/intern/cycles/kernel/svm/svm_ao.h b/intern/cycles/kernel/svm/svm_ao.h
index 74617f9bb9e..34ac2cb8fbf 100644
--- a/intern/cycles/kernel/svm/svm_ao.h
+++ b/intern/cycles/kernel/svm/svm_ao.h
@@ -20,12 +20,16 @@ CCL_NAMESPACE_BEGIN
#ifdef __SHADER_RAYTRACE__
+# ifdef __KERNEL_OPTIX__
+extern "C" __device__ float __direct_callable__svm_node_ao(INTEGRATOR_STATE_CONST_ARGS,
+# else
ccl_device float svm_ao(INTEGRATOR_STATE_CONST_ARGS,
- ShaderData *sd,
- float3 N,
- float max_dist,
- int num_samples,
- int flags)
+# endif
+ ShaderData *sd,
+ float3 N,
+ float max_dist,
+ int num_samples,
+ int flags)
{
if (flags & NODE_AO_GLOBAL_RADIUS) {
max_dist = kernel_data.integrator.ao_bounces_distance;
@@ -85,6 +89,7 @@ ccl_device float svm_ao(INTEGRATOR_STATE_CONST_ARGS,
return ((float)unoccluded) / num_samples;
}
+template<uint node_feature_mask>
# if defined(__KERNEL_OPTIX__)
ccl_device_inline
# else
@@ -93,16 +98,6 @@ ccl_device_noinline
void
svm_node_ao(INTEGRATOR_STATE_CONST_ARGS, ShaderData *sd, float *stack, uint4 node)
{
-# if defined(__KERNEL_OPTIX__)
- optixDirectCall<void>(0, INTEGRATOR_STATE_PASS, sd, stack, node);
-}
-
-extern "C" __device__ void __direct_callable__svm_node_ao(INTEGRATOR_STATE_CONST_ARGS,
- ShaderData *sd,
- float *stack,
- uint4 node)
-{
-# endif
uint flags, dist_offset, normal_offset, out_ao_offset;
svm_unpack_node_uchar4(node.y, &flags, &dist_offset, &normal_offset, &out_ao_offset);
@@ -111,7 +106,16 @@ extern "C" __device__ void __direct_callable__svm_node_ao(INTEGRATOR_STATE_CONST
float dist = stack_load_float_default(stack, dist_offset, node.w);
float3 normal = stack_valid(normal_offset) ? stack_load_float3(stack, normal_offset) : sd->N;
- float ao = svm_ao(INTEGRATOR_STATE_PASS, sd, normal, dist, samples, flags);
+
+ float ao = 1.0f;
+
+ if (KERNEL_NODES_FEATURE(RAYTRACE)) {
+# ifdef __KERNEL_OPTIX__
+ ao = optixDirectCall<float>(0, INTEGRATOR_STATE_PASS, sd, normal, dist, samples, flags);
+# else
+ ao = svm_ao(INTEGRATOR_STATE_PASS, sd, normal, dist, samples, flags);
+# endif
+ }
if (stack_valid(out_ao_offset)) {
stack_store_float(stack, out_ao_offset, ao);
diff --git a/intern/cycles/kernel/svm/svm_bevel.h b/intern/cycles/kernel/svm/svm_bevel.h
index ab2b3e110d3..aab089d19ea 100644
--- a/intern/cycles/kernel/svm/svm_bevel.h
+++ b/intern/cycles/kernel/svm/svm_bevel.h
@@ -95,10 +95,14 @@ ccl_device void svm_bevel_cubic_sample(const float radius, float xi, float *r, f
* http://library.imageworks.com/pdfs/imageworks-library-BSSRDF-sampling.pdf
*/
+# ifdef __KERNEL_OPTIX__
+extern "C" __device__ float3 __direct_callable__svm_node_bevel(INTEGRATOR_STATE_CONST_ARGS,
+# else
ccl_device float3 svm_bevel(INTEGRATOR_STATE_CONST_ARGS,
- ShaderData *sd,
- float radius,
- int num_samples)
+# endif
+ ShaderData *sd,
+ float radius,
+ int num_samples)
{
/* Early out if no sampling needed. */
if (radius <= 0.0f || num_samples < 1 || sd->object == OBJECT_NONE) {
@@ -276,6 +280,7 @@ ccl_device float3 svm_bevel(INTEGRATOR_STATE_CONST_ARGS,
return is_zero(N) ? sd->N : (sd->flag & SD_BACKFACING) ? -N : N;
}
+template<uint node_feature_mask>
# if defined(__KERNEL_OPTIX__)
ccl_device_inline
# else
@@ -284,26 +289,25 @@ ccl_device_noinline
void
svm_node_bevel(INTEGRATOR_STATE_CONST_ARGS, ShaderData *sd, float *stack, uint4 node)
{
-# if defined(__KERNEL_OPTIX__)
- optixDirectCall<void>(1, INTEGRATOR_STATE_PASS, sd, stack, node);
-}
-
-extern "C" __device__ void __direct_callable__svm_node_bevel(INTEGRATOR_STATE_CONST_ARGS,
- ShaderData *sd,
- float *stack,
- uint4 node)
-{
-# endif
uint num_samples, radius_offset, normal_offset, out_offset;
svm_unpack_node_uchar4(node.y, &num_samples, &radius_offset, &normal_offset, &out_offset);
float radius = stack_load_float(stack, radius_offset);
- float3 bevel_N = svm_bevel(INTEGRATOR_STATE_PASS, sd, radius, num_samples);
- if (stack_valid(normal_offset)) {
- /* Preserve input normal. */
- float3 ref_N = stack_load_float3(stack, normal_offset);
- bevel_N = normalize(ref_N + (bevel_N - sd->N));
+ float3 bevel_N = sd->N;
+
+ if (KERNEL_NODES_FEATURE(RAYTRACE)) {
+# ifdef __KERNEL_OPTIX__
+ bevel_N = optixDirectCall<float3>(1, INTEGRATOR_STATE_PASS, sd, radius, num_samples);
+# else
+ bevel_N = svm_bevel(INTEGRATOR_STATE_PASS, sd, radius, num_samples);
+# endif
+
+ if (stack_valid(normal_offset)) {
+ /* Preserve input normal. */
+ float3 ref_N = stack_load_float3(stack, normal_offset);
+ bevel_N = normalize(ref_N + (bevel_N - sd->N));
+ }
}
stack_store_float3(stack, out_offset, bevel_N);
More information about the Bf-blender-cvs
mailing list