[Bf-blender-cvs] [4a2cde3f0e8] master: Cycles: Enable SSS and volumes for CUDA and Nvidia OpenCL split kernel

[Mai Lavelle]

Commit: 4a2cde3f0e88e84c94617a9a3175c67f8801eecf
Author: Mai Lavelle
Date:   Fri Mar 10 02:09:41 2017 -0500
Branches: master
https://developer.blender.org/rB4a2cde3f0e88e84c94617a9a3175c67f8801eecf

Cycles: Enable SSS and volumes for CUDA and Nvidia OpenCL split kernel

===================================================================

M	intern/cycles/kernel/kernel_types.h
M	intern/cycles/kernel/kernels/cuda/kernel_split.cu

===================================================================

diff --git a/intern/cycles/kernel/kernel_types.h b/intern/cycles/kernel/kernel_types.h
index cac710c5ff3..f2ba3586c22 100644
--- a/intern/cycles/kernel/kernel_types.h
+++ b/intern/cycles/kernel/kernel_types.h
@@ -90,13 +90,13 @@ CCL_NAMESPACE_BEGIN
 #ifdef __KERNEL_CUDA__
 #  define __KERNEL_SHADING__
 #  define __KERNEL_ADV_SHADING__
+#  define __VOLUME__
+#  define __VOLUME_SCATTER__
+#  define __SUBSURFACE__
+#  define __SHADOW_RECORD_ALL__
 #  ifndef __SPLIT_KERNEL__
 #    define __BRANCHED_PATH__
-#    define __VOLUME__
-#    define __VOLUME_SCATTER__
-#    define __SUBSURFACE__
 #    define __CMJ__
-#    define __SHADOW_RECORD_ALL__
 #  endif
 #endif  /* __KERNEL_CUDA__ */
 
@@ -107,6 +107,10 @@ CCL_NAMESPACE_BEGIN
 #  ifdef __KERNEL_OPENCL_NVIDIA__
 #    define __KERNEL_SHADING__
 #    define __KERNEL_ADV_SHADING__
+#    define __SUBSURFACE__
+#    define __VOLUME__
+#    define __VOLUME_SCATTER__
+#    define __SHADOW_RECORD_ALL__
 #    ifdef __KERNEL_EXPERIMENTAL__
 #      define __CMJ__
 #    endif
diff --git a/intern/cycles/kernel/kernels/cuda/kernel_split.cu b/intern/cycles/kernel/kernels/cuda/kernel_split.cu
index 759475b175f..6c508c2cd79 100644
--- a/intern/cycles/kernel/kernels/cuda/kernel_split.cu
+++ b/intern/cycles/kernel/kernels/cuda/kernel_split.cu
@@ -28,13 +28,18 @@
 #include "../../split/kernel_path_init.h"
 #include "../../split/kernel_scene_intersect.h"
 #include "../../split/kernel_lamp_emission.h"
+#include "../../split/kernel_do_volume.h"
 #include "../../split/kernel_queue_enqueue.h"
-#include "../../split/kernel_background_buffer_update.h"
+#include "../../split/kernel_indirect_background.h"
 #include "../../split/kernel_shader_eval.h"
 #include "../../split/kernel_holdout_emission_blurring_pathtermination_ao.h"
+#include "../../split/kernel_subsurface_scatter.h"
 #include "../../split/kernel_direct_lighting.h"
-#include "../../split/kernel_shadow_blocked.h"
+#include "../../split/kernel_shadow_blocked_ao.h"
+#include "../../split/kernel_shadow_blocked_dl.h"
 #include "../../split/kernel_next_iteration_setup.h"
+#include "../../split/kernel_indirect_subsurface.h"
+#include "../../split/kernel_buffer_update.h"
 
 #include "../../kernel_film.h"
 
@@ -91,13 +96,18 @@ kernel_cuda_path_trace_data_init(
 DEFINE_SPLIT_KERNEL_FUNCTION(path_init)
 DEFINE_SPLIT_KERNEL_FUNCTION(scene_intersect)
 DEFINE_SPLIT_KERNEL_FUNCTION(lamp_emission)
+DEFINE_SPLIT_KERNEL_FUNCTION(do_volume)
 DEFINE_SPLIT_KERNEL_FUNCTION(queue_enqueue)
-DEFINE_SPLIT_KERNEL_FUNCTION(background_buffer_update)
+DEFINE_SPLIT_KERNEL_FUNCTION(indirect_background)
 DEFINE_SPLIT_KERNEL_FUNCTION(shader_eval)
 DEFINE_SPLIT_KERNEL_FUNCTION(holdout_emission_blurring_pathtermination_ao)
+DEFINE_SPLIT_KERNEL_FUNCTION(subsurface_scatter)
 DEFINE_SPLIT_KERNEL_FUNCTION(direct_lighting)
-DEFINE_SPLIT_KERNEL_FUNCTION(shadow_blocked)
+DEFINE_SPLIT_KERNEL_FUNCTION(shadow_blocked_ao)
+DEFINE_SPLIT_KERNEL_FUNCTION(shadow_blocked_dl)
 DEFINE_SPLIT_KERNEL_FUNCTION(next_iteration_setup)
+DEFINE_SPLIT_KERNEL_FUNCTION(indirect_subsurface)
+DEFINE_SPLIT_KERNEL_FUNCTION(buffer_update)
 
 extern "C" __global__ void
 CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)