diff options
Diffstat (limited to 'intern/cycles/kernel/svm/svm_ao.h')
| -rw-r--r-- | intern/cycles/kernel/svm/svm_ao.h | 53 |
1 files changed, 38 insertions, 15 deletions
diff --git a/intern/cycles/kernel/svm/svm_ao.h b/intern/cycles/kernel/svm/svm_ao.h index 4cb986b897a..34ac2cb8fbf 100644 --- a/intern/cycles/kernel/svm/svm_ao.h +++ b/intern/cycles/kernel/svm/svm_ao.h @@ -14,20 +14,25 @@ * limitations under the License. */ +#include "kernel/bvh/bvh.h" + CCL_NAMESPACE_BEGIN #ifdef __SHADER_RAYTRACE__ -ccl_device_noinline float svm_ao(KernelGlobals *kg, - ShaderData *sd, - float3 N, - ccl_addr_space PathState *state, - float max_dist, - int num_samples, - int flags) +# 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, +# endif + ShaderData *sd, + float3 N, + float max_dist, + int num_samples, + int flags) { if (flags & NODE_AO_GLOBAL_RADIUS) { - max_dist = kernel_data.background.ao_distance; + max_dist = kernel_data.integrator.ao_bounces_distance; } /* Early out if no sampling needed. */ @@ -47,11 +52,14 @@ ccl_device_noinline float svm_ao(KernelGlobals *kg, float3 T, B; make_orthonormals(N, &T, &B); + /* TODO: support ray-tracing in shadow shader evaluation? */ + RNGState rng_state; + path_state_rng_load(INTEGRATOR_STATE_PASS, &rng_state); + int unoccluded = 0; for (int sample = 0; sample < num_samples; sample++) { float disk_u, disk_v; - path_branched_rng_2D( - kg, state->rng_hash, state, sample, num_samples, PRNG_BEVEL_U, &disk_u, &disk_v); + path_branched_rng_2D(kg, &rng_state, sample, num_samples, PRNG_BEVEL_U, &disk_u, &disk_v); float2 d = concentric_sample_disk(disk_u, disk_v); float3 D = make_float3(d.x, d.y, safe_sqrtf(1.0f - dot(d, d))); @@ -62,8 +70,8 @@ ccl_device_noinline float svm_ao(KernelGlobals *kg, ray.D = D.x * T + D.y * B + D.z * N; ray.t = max_dist; ray.time = sd->time; - ray.dP = sd->dP; - ray.dD = differential3_zero(); + ray.dP = differential_zero_compact(); + ray.dD = differential_zero_compact(); if (flags & NODE_AO_ONLY_LOCAL) { if (!scene_intersect_local(kg, &ray, NULL, sd->object, NULL, 0)) { @@ -81,8 +89,14 @@ ccl_device_noinline float svm_ao(KernelGlobals *kg, return ((float)unoccluded) / num_samples; } -ccl_device void svm_node_ao( - KernelGlobals *kg, ShaderData *sd, ccl_addr_space PathState *state, float *stack, uint4 node) +template<uint node_feature_mask> +# if defined(__KERNEL_OPTIX__) +ccl_device_inline +# else +ccl_device_noinline +# endif + void + svm_node_ao(INTEGRATOR_STATE_CONST_ARGS, ShaderData *sd, float *stack, uint4 node) { uint flags, dist_offset, normal_offset, out_ao_offset; svm_unpack_node_uchar4(node.y, &flags, &dist_offset, &normal_offset, &out_ao_offset); @@ -92,7 +106,16 @@ ccl_device void svm_node_ao( 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(kg, sd, normal, state, 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); |