Page MenuHome
Differential D12888 Diff 43495 intern/cycles/kernel/svm/svm_bevel.h

Changeset View

Standalone View

View Options

intern/cycles/kernel/svm/svm_bevel.h

Show First 20 Lines • Show All 93 Lines • ▼ Show 20 Lines
/* Bevel shader averaging normals from nearby surfaces. /* Bevel shader averaging normals from nearby surfaces.
* *
* Sampling strategy from: BSSRDF Importance Sampling, SIGGRAPH 2013 * Sampling strategy from: BSSRDF Importance Sampling, SIGGRAPH 2013
* http://library.imageworks.com/pdfs/imageworks-library-BSSRDF-sampling.pdf * http://library.imageworks.com/pdfs/imageworks-library-BSSRDF-sampling.pdf
*/ */
# ifdef __KERNEL_OPTIX__ # ifdef __KERNEL_OPTIX__
extern "C" __device__ float3 __direct_callable__svm_node_bevel(INTEGRATOR_STATE_CONST_ARGS, extern "C" __device__ float3 __direct_callable__svm_node_bevel(KernelGlobals kg,
ConstIntegratorState state,
# else # else
ccl_device float3 svm_bevel(INTEGRATOR_STATE_CONST_ARGS, ccl_device float3 svm_bevel(KernelGlobals kg,
ConstIntegratorState state,
# endif # endif
ccl_private ShaderData *sd, ccl_private ShaderData *sd,
float radius, float radius,
int num_samples) int num_samples)
{ {
/* Early out if no sampling needed. */ /* Early out if no sampling needed. */
if (radius <= 0.0f || num_samples < 1 || sd->object == OBJECT_NONE) { if (radius <= 0.0f || num_samples < 1 || sd->object == OBJECT_NONE) {
return sd->N; return sd->N;
} }
/* Can't raytrace from shaders like displacement, before BVH exists. */ /* Can't raytrace from shaders like displacement, before BVH exists. */
if (kernel_data.bvh.bvh_layout == BVH_LAYOUT_NONE) { if (kernel_data.bvh.bvh_layout == BVH_LAYOUT_NONE) {
return sd->N; return sd->N;
} }
/* Don't bevel for blurry indirect rays. */ /* Don't bevel for blurry indirect rays. */
if (INTEGRATOR_STATE(path, min_ray_pdf) < 8.0f) { if (INTEGRATOR_STATE(state, path, min_ray_pdf) < 8.0f) {
return sd->N; return sd->N;
} }
/* Setup for multi intersection. */ /* Setup for multi intersection. */
LocalIntersection isect; LocalIntersection isect;
uint lcg_state = lcg_state_init(INTEGRATOR_STATE(path, rng_hash), uint lcg_state = lcg_state_init(INTEGRATOR_STATE(state, path, rng_hash),
INTEGRATOR_STATE(path, rng_offset), INTEGRATOR_STATE(state, path, rng_offset),
INTEGRATOR_STATE(path, sample), INTEGRATOR_STATE(state, path, sample),
0x64c6a40e); 0x64c6a40e);
/* Sample normals from surrounding points on surface. */ /* Sample normals from surrounding points on surface. */
float3 sum_N = make_float3(0.0f, 0.0f, 0.0f); float3 sum_N = make_float3(0.0f, 0.0f, 0.0f);
/* TODO: support ray-tracing in shadow shader evaluation? */ /* TODO: support ray-tracing in shadow shader evaluation? */
RNGState rng_state; RNGState rng_state;
path_state_rng_load(INTEGRATOR_STATE_PASS, &rng_state); path_state_rng_load(state, &rng_state);
for (int sample = 0; sample < num_samples; sample++) { for (int sample = 0; sample < num_samples; sample++) {
float disk_u, disk_v; float disk_u, disk_v;
path_branched_rng_2D(kg, &rng_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);
/* Pick random axis in local frame and point on disk. */ /* Pick random axis in local frame and point on disk. */
float3 disk_N, disk_T, disk_B; float3 disk_N, disk_T, disk_B;
float pick_pdf_N, pick_pdf_T, pick_pdf_B; float pick_pdf_N, pick_pdf_T, pick_pdf_B;
▲ Show 20 Lines • Show All 136 Lines • ▼ Show 20 Lines
template<uint node_feature_mask> template<uint node_feature_mask>
# if defined(__KERNEL_OPTIX__) # if defined(__KERNEL_OPTIX__)
ccl_device_inline ccl_device_inline
# else # else
ccl_device_noinline ccl_device_noinline
# endif # endif
void void
svm_node_bevel(INTEGRATOR_STATE_CONST_ARGS, svm_node_bevel(KernelGlobals kg,
ConstIntegratorState state,
ccl_private ShaderData *sd, ccl_private ShaderData *sd,
ccl_private float *stack, ccl_private float *stack,
uint4 node) uint4 node)
{ {
uint num_samples, radius_offset, normal_offset, out_offset; uint num_samples, radius_offset, normal_offset, out_offset;
svm_unpack_node_uchar4(node.y, &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); float radius = stack_load_float(stack, radius_offset);
float3 bevel_N = sd->N; float3 bevel_N = sd->N;
if (KERNEL_NODES_FEATURE(RAYTRACE)) { IF_KERNEL_NODES_FEATURE(RAYTRACE)
{
# ifdef __KERNEL_OPTIX__ # ifdef __KERNEL_OPTIX__
bevel_N = optixDirectCall<float3>(1, INTEGRATOR_STATE_PASS, sd, radius, num_samples); bevel_N = optixDirectCall<float3>(1, kg, state, sd, radius, num_samples);
# else # else
bevel_N = svm_bevel(INTEGRATOR_STATE_PASS, sd, radius, num_samples); bevel_N = svm_bevel(kg, state, sd, radius, num_samples);
# endif # endif
if (stack_valid(normal_offset)) { if (stack_valid(normal_offset)) {
/* Preserve input normal. */ /* Preserve input normal. */
float3 ref_N = stack_load_float3(stack, normal_offset); float3 ref_N = stack_load_float3(stack, normal_offset);
bevel_N = normalize(ref_N + (bevel_N - sd->N)); bevel_N = normalize(ref_N + (bevel_N - sd->N));
} }
} }
stack_store_float3(stack, out_offset, bevel_N); stack_store_float3(stack, out_offset, bevel_N);
} }
#endif /* __SHADER_RAYTRACE__ */ #endif /* __SHADER_RAYTRACE__ */
CCL_NAMESPACE_END CCL_NAMESPACE_END