Cycles: Cleanup, synchronize name across file name, program and kernel names

5 files changed, 47 insertions, 47 deletions

diff --git a/intern/cycles/device/device_opencl.cpp b/intern/cycles/device/device_opencl.cpp
index 8be36fc4675..68c4af81317 100644
--- a/intern/cycles/device/device_opencl.cpp
+++ b/intern/cycles/device/device_opencl.cpp
@@ -1514,11 +1514,11 @@ public:
 	cl_kernel ckPathTraceKernel_lamp_emission;
 	cl_kernel ckPathTraceKernel_queue_enqueue;
 	cl_kernel ckPathTraceKernel_background_buffer_update;
-	cl_kernel ckPathTraceKernel_shader_lighting;
+	cl_kernel ckPathTraceKernel_shader_eval;
 	cl_kernel ckPathTraceKernel_holdout_emission_blurring_pathtermination_ao;
 	cl_kernel ckPathTraceKernel_direct_lighting;
-	cl_kernel ckPathTraceKernel_shadow_blocked_direct_lighting;
-	cl_kernel ckPathTraceKernel_setup_next_iteration;
+	cl_kernel ckPathTraceKernel_shadow_blocked;
+	cl_kernel ckPathTraceKernel_next_iteration_setup;
 	cl_kernel ckPathTraceKernel_sum_all_radiance;
 
 	/* cl_program declaration. */
@@ -1684,11 +1684,11 @@ public:
 		ckPathTraceKernel_scene_intersect = NULL;
 		ckPathTraceKernel_lamp_emission = NULL;
 		ckPathTraceKernel_background_buffer_update = NULL;
-		ckPathTraceKernel_shader_lighting = NULL;
+		ckPathTraceKernel_shader_eval = NULL;
 		ckPathTraceKernel_holdout_emission_blurring_pathtermination_ao = NULL;
 		ckPathTraceKernel_direct_lighting = NULL;
-		ckPathTraceKernel_shadow_blocked_direct_lighting = NULL;
-		ckPathTraceKernel_setup_next_iteration = NULL;
+		ckPathTraceKernel_shadow_blocked = NULL;
+		ckPathTraceKernel_next_iteration_setup = NULL;
 		ckPathTraceKernel_sum_all_radiance = NULL;
 		ckPathTraceKernel_queue_enqueue = NULL;
 
@@ -2017,6 +2017,7 @@ public:
 		                      custom_kernel_build_options, \
 		                      &GLUE(name, _program))) \
 		{ \
+			fprintf(stderr, "Faled to compile %s\n", #name); \
 			return false; \
 		} \
 	} while(false)
@@ -2035,29 +2036,28 @@ public:
 
 #undef LOAD_KERNEL
 
-#define FIND_KERNEL(kernel, program, function) \
+#define FIND_KERNEL(name) \
 	do { \
-		GLUE(ckPathTraceKernel_, kernel) = \
-			clCreateKernel(GLUE(program, _program), \
-			               "kernel_ocl_path_trace_"  function, &ciErr); \
+		GLUE(ckPathTraceKernel_, name) = \
+			clCreateKernel(GLUE(name, _program), \
+			               "kernel_ocl_path_trace_"  #name, &ciErr); \
 		if(opencl_error(ciErr)) { \
+			fprintf(stderr,"Missing kernel kernel_ocl_path_trace_%s\n", #name); \
 			return false; \
 		} \
 	} while(false)
 
-		FIND_KERNEL(data_init, data_init, "data_initialization");
-		FIND_KERNEL(scene_intersect, scene_intersect, "scene_intersect");
-		FIND_KERNEL(lamp_emission, lamp_emission, "lamp_emission");
-		FIND_KERNEL(queue_enqueue, queue_enqueue, "queue_enqueue");
-		FIND_KERNEL(background_buffer_update, background_buffer_update, "background_buffer_update");
-		FIND_KERNEL(shader_lighting, shader_eval, "shader_evaluation");
-		FIND_KERNEL(holdout_emission_blurring_pathtermination_ao,
-		            holdout_emission_blurring_pathtermination_ao,
-		            "holdout_emission_blurring_pathtermination_ao");
-		FIND_KERNEL(direct_lighting, direct_lighting, "direct_lighting");
-		FIND_KERNEL(shadow_blocked_direct_lighting, shadow_blocked, "shadow_blocked_direct_lighting");
-		FIND_KERNEL(setup_next_iteration, next_iteration_setup, "setup_next_iteration");
-		FIND_KERNEL(sum_all_radiance, sum_all_radiance, "sum_all_radiance");
+		FIND_KERNEL(data_init);
+		FIND_KERNEL(scene_intersect);
+		FIND_KERNEL(lamp_emission);
+		FIND_KERNEL(queue_enqueue);
+		FIND_KERNEL(background_buffer_update);
+		FIND_KERNEL(shader_eval);
+		FIND_KERNEL(holdout_emission_blurring_pathtermination_ao);
+		FIND_KERNEL(direct_lighting);
+		FIND_KERNEL(shadow_blocked);
+		FIND_KERNEL(next_iteration_setup);
+		FIND_KERNEL(sum_all_radiance);
 #undef FIND_KERNEL
 #undef GLUE
 
@@ -2076,11 +2076,11 @@ public:
 		release_kernel_safe(ckPathTraceKernel_lamp_emission);
 		release_kernel_safe(ckPathTraceKernel_queue_enqueue);
 		release_kernel_safe(ckPathTraceKernel_background_buffer_update);
-		release_kernel_safe(ckPathTraceKernel_shader_lighting);
+		release_kernel_safe(ckPathTraceKernel_shader_eval);
 		release_kernel_safe(ckPathTraceKernel_holdout_emission_blurring_pathtermination_ao);
 		release_kernel_safe(ckPathTraceKernel_direct_lighting);
-		release_kernel_safe(ckPathTraceKernel_shadow_blocked_direct_lighting);
-		release_kernel_safe(ckPathTraceKernel_setup_next_iteration);
+		release_kernel_safe(ckPathTraceKernel_shadow_blocked);
+		release_kernel_safe(ckPathTraceKernel_next_iteration_setup);
 		release_kernel_safe(ckPathTraceKernel_sum_all_radiance);
 
 		/* Release global memory */
@@ -2595,7 +2595,7 @@ public:
 #endif
 		                 num_parallel_samples);
 
-		kernel_set_args(ckPathTraceKernel_shader_lighting,
+		kernel_set_args(ckPathTraceKernel_shader_eval,
 		                0,
 		                kgbuffer,
 		                d_data,
@@ -2655,7 +2655,7 @@ public:
 		                Queue_index,
 		                dQueue_size);
 
-		kernel_set_args(ckPathTraceKernel_shadow_blocked_direct_lighting,
+		kernel_set_args(ckPathTraceKernel_shadow_blocked,
 		                0,
 		                kgbuffer,
 		                d_data,
@@ -2671,7 +2671,7 @@ public:
 		                dQueue_size,
 		                total_num_rays);
 
-		kernel_set_args(ckPathTraceKernel_setup_next_iteration,
+		kernel_set_args(ckPathTraceKernel_next_iteration_setup,
 		                0,
 		                kgbuffer,
 		                d_data,
@@ -2741,11 +2741,11 @@ public:
 				ENQUEUE_SPLIT_KERNEL(lamp_emission, global_size, local_size);
 				ENQUEUE_SPLIT_KERNEL(queue_enqueue, global_size, local_size);
 				ENQUEUE_SPLIT_KERNEL(background_buffer_update, global_size, local_size);
-				ENQUEUE_SPLIT_KERNEL(shader_lighting, global_size, local_size);
+				ENQUEUE_SPLIT_KERNEL(shader_eval, global_size, local_size);
 				ENQUEUE_SPLIT_KERNEL(holdout_emission_blurring_pathtermination_ao, global_size, local_size);
 				ENQUEUE_SPLIT_KERNEL(direct_lighting, global_size, local_size);
-				ENQUEUE_SPLIT_KERNEL(shadow_blocked_direct_lighting, global_size_shadow_blocked, local_size);
-				ENQUEUE_SPLIT_KERNEL(setup_next_iteration, global_size, local_size);
+				ENQUEUE_SPLIT_KERNEL(shadow_blocked, global_size_shadow_blocked, local_size);
+				ENQUEUE_SPLIT_KERNEL(next_iteration_setup, global_size, local_size);
 			}
 
 			/* Read ray-state into Host memory to decide if we should exit
diff --git a/intern/cycles/kernel/kernel_data_init.cl b/intern/cycles/kernel/kernel_data_init.cl
index 4c5f08639dd..62b5c4e6a29 100644
--- a/intern/cycles/kernel/kernel_data_init.cl
+++ b/intern/cycles/kernel/kernel_data_init.cl
@@ -51,7 +51,7 @@
  * All slots in queues are initialized to queue empty slot;
  * The number of elements in the queues is initialized to 0;
  */
-__kernel void kernel_ocl_path_trace_data_initialization(
+__kernel void kernel_ocl_path_trace_data_init(
 	ccl_global char *globals,
 	ccl_global char *shader_data_sd,                  /* Arguments related to ShaderData */
 	ccl_global char *shader_data_sd_DL_shadow,     /* Arguments related to ShaderData */
diff --git a/intern/cycles/kernel/kernel_next_iteration_setup.cl b/intern/cycles/kernel/kernel_next_iteration_setup.cl
index 49562ca6ed5..3c0e4e9240d 100644
--- a/intern/cycles/kernel/kernel_next_iteration_setup.cl
+++ b/intern/cycles/kernel/kernel_next_iteration_setup.cl
@@ -27,7 +27,7 @@
  *
  * The input and output are as follows,
  *
- * rng_coop ---------------------------------------------|--- kernel_ocl_path_trace_setup_next_iteration -|--- Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
+ * rng_coop ---------------------------------------------|--- kernel_ocl_path_trace_next_iteration_setup -|--- Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
  * throughput_coop --------------------------------------|                                                |--- Queue_data (QUEUE_HITBF_BUFF_UPDATE_TOREGEN_RAYS)
  * PathRadiance_coop ------------------------------------|                                                |--- throughput_coop
  * PathState_coop ---------------------------------------|                                                |--- PathRadiance_coop
@@ -61,7 +61,7 @@
  * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be filled with RAY_TO_REGENERATE and more RAY_UPDATE_BUFFER rays
  */
 
-__kernel void kernel_ocl_path_trace_setup_next_iteration(
+__kernel void kernel_ocl_path_trace_next_iteration_setup(
 	ccl_global char *globals,
 	ccl_constant KernelData *data,
 	ccl_global char *shader_data,               /* Required for setting up ray for next iteration */
diff --git a/intern/cycles/kernel/kernel_shader_eval.cl b/intern/cycles/kernel/kernel_shader_eval.cl
index 78cf19a3df8..b3983081be6 100644
--- a/intern/cycles/kernel/kernel_shader_eval.cl
+++ b/intern/cycles/kernel/kernel_shader_eval.cl
@@ -17,22 +17,22 @@
 #include "kernel_split.h"
 
 /*
- * Note on kernel_ocl_path_trace_shader_evaluation kernel
+ * Note on kernel_ocl_path_trace_shader_eval kernel
  * This kernel is the 5th kernel in the ray tracing logic. This is
  * the 4rd kernel in path iteration. This kernel sets up the ShaderData
  * structure from the values computed by the previous kernels. It also identifies
  * the rays of state RAY_TO_REGENERATE and enqueues them in QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS queue.
  *
  * The input and output of the kernel is as follows,
- * rng_coop -------------------------------------------|--- kernel_ocl_path_trace_shader_evaluation --|--- shader_data
- * Ray_coop -------------------------------------------|                                              |--- Queue_data (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
- * PathState_coop -------------------------------------|                                              |--- Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
- * Intersection_coop ----------------------------------|                                              |
- * Queue_data (QUEUE_ACTIVE_AND_REGENERATD_RAYS)-------|                                              |
- * Queue_index(QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)---|                                              |
- * ray_state ------------------------------------------|                                              |
- * kg (globals + data) --------------------------------|                                              |
- * queuesize ------------------------------------------|                                              |
+ * rng_coop -------------------------------------------|--- kernel_ocl_path_trace_shader_eval --|--- shader_data
+ * Ray_coop -------------------------------------------|                                        |--- Queue_data (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
+ * PathState_coop -------------------------------------|                                        |--- Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
+ * Intersection_coop ----------------------------------|                                        |
+ * Queue_data (QUEUE_ACTIVE_AND_REGENERATD_RAYS)-------|                                        |
+ * Queue_index(QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)---|                                        |
+ * ray_state ------------------------------------------|                                        |
+ * kg (globals + data) --------------------------------|                                        |
+ * queuesize ------------------------------------------|                                        |
  *
  * Note on Queues :
  * This kernel reads from the QUEUE_ACTIVE_AND_REGENERATED_RAYS queue and processes
@@ -46,7 +46,7 @@
  * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be filled with RAY_TO_REGENERATE rays
  */
 
-__kernel void kernel_ocl_path_trace_shader_evaluation(
+__kernel void kernel_ocl_path_trace_shader_eval(
 	ccl_global char *globals,
 	ccl_constant KernelData *data,
 	ccl_global char *shader_data,               /* Output ShaderData structure to be filled */
diff --git a/intern/cycles/kernel/kernel_shadow_blocked.cl b/intern/cycles/kernel/kernel_shadow_blocked.cl
index 72a2d0affb0..2fc4be6d528 100644
--- a/intern/cycles/kernel/kernel_shadow_blocked.cl
+++ b/intern/cycles/kernel/kernel_shadow_blocked.cl
@@ -47,7 +47,7 @@
  * QUEUE_SHADOW_RAY_CAST_AO_RAYS and QUEUE_SHADOW_RAY_CAST_DL_RAYS will be empty at kernel exit.
  */
 
-__kernel void kernel_ocl_path_trace_shadow_blocked_direct_lighting(
+__kernel void kernel_ocl_path_trace_shadow_blocked(
 	ccl_global char *globals,
 	ccl_constant KernelData *data,
 	ccl_global char *shader_shadow,             /* Required for shadow blocked */