[Bf-blender-cvs] [cb5b0e0] cycles_kernel_split: Cycles kernel split: Move address space specification to DebugData typedef
Sergey Sharybin
noreply at git.blender.org
Thu May 7 11:57:49 CEST 2015
Commit: cb5b0e037b4fd872778720fff9b6712b65ac9c34
Author: Sergey Sharybin
Date: Thu May 7 14:56:13 2015 +0500
Branches: cycles_kernel_split
https://developer.blender.org/rBcb5b0e037b4fd872778720fff9b6712b65ac9c34
Cycles kernel split: Move address space specification to DebugData typedef
Debug pass does not really work, but it isn't caused by this change and would
need to have separate look.
===================================================================
M intern/cycles/kernel/kernel_Background_BufferUpdate.cl
M intern/cycles/kernel/kernel_DataInit.cl
M intern/cycles/kernel/kernel_SceneIntersect.cl
M intern/cycles/kernel/kernel_debug.h
M intern/cycles/kernel/kernel_types.h
===================================================================
diff --git a/intern/cycles/kernel/kernel_Background_BufferUpdate.cl b/intern/cycles/kernel/kernel_Background_BufferUpdate.cl
index 9634e60..f5621f2 100644
--- a/intern/cycles/kernel/kernel_Background_BufferUpdate.cl
+++ b/intern/cycles/kernel/kernel_Background_BufferUpdate.cl
@@ -123,7 +123,7 @@ __kernel void kernel_ocl_path_trace_Background_BufferUpdate_SPLIT_KERNEL(
unsigned int num_samples,
#endif
#ifdef __KERNEL_DEBUG__
- ccl_global DebugData *debugdata_coop,
+ DebugData *debugdata_coop,
#endif
int parallel_samples /* Number of samples to be processed in parallel */
)
@@ -160,7 +160,7 @@ __kernel void kernel_ocl_path_trace_Background_BufferUpdate_SPLIT_KERNEL(
ShaderData *sd = (ShaderData *)shader_data;
#ifdef __KERNEL_DEBUG__
- ccl_global DebugData *debug_data = &debugdata_coop[ray_index];
+ DebugData *debug_data = &debugdata_coop[ray_index];
#endif
ccl_global PathState *state = &PathState_coop[ray_index];
PathRadiance *L = L = &PathRadiance_coop[ray_index];
diff --git a/intern/cycles/kernel/kernel_DataInit.cl b/intern/cycles/kernel/kernel_DataInit.cl
index 8019f3d..15897dd 100644
--- a/intern/cycles/kernel/kernel_DataInit.cl
+++ b/intern/cycles/kernel/kernel_DataInit.cl
@@ -190,7 +190,7 @@ __kernel void kernel_ocl_path_trace_data_initialization_SPLIT_KERNEL(
unsigned int num_samples, /* Total number of samples per pixel */
#endif
#ifdef __KERNEL_DEBUG__
- ccl_global DebugData *debugdata_coop,
+ DebugData *debugdata_coop,
#endif
int parallel_samples /* Number of samples to be processed in parallel */
)
diff --git a/intern/cycles/kernel/kernel_SceneIntersect.cl b/intern/cycles/kernel/kernel_SceneIntersect.cl
index 34bb359..111fa25 100644
--- a/intern/cycles/kernel/kernel_SceneIntersect.cl
+++ b/intern/cycles/kernel/kernel_SceneIntersect.cl
@@ -77,7 +77,7 @@ __kernel void kernel_ocl_path_trace_SceneIntersect_SPLIT_KERNEL(
int queuesize, /* Size (capacity) of queues */
ccl_global char *use_queues_flag, /* used to decide if this kernel should use queues to fetch ray index */
#ifdef __KERNEL_DEBUG__
- ccl_global DebugData *debugdata_coop,
+ DebugData *debugdata_coop,
#endif
int parallel_samples /* Number of samples to be processed in parallel */
)
@@ -119,7 +119,7 @@ __kernel void kernel_ocl_path_trace_SceneIntersect_SPLIT_KERNEL(
KernelGlobals *kg = (KernelGlobals *)globals;
#ifdef __KERNEL_DEBUG__
- ccl_global DebugData *debug_data = &debugdata_coop[ray_index];
+ DebugData *debug_data = &debugdata_coop[ray_index];
#endif
Intersection *isect = &Intersection_coop[ray_index];
PathState state = PathState_coop[ray_index];
diff --git a/intern/cycles/kernel/kernel_debug.h b/intern/cycles/kernel/kernel_debug.h
index 9c1de97..94ede39 100644
--- a/intern/cycles/kernel/kernel_debug.h
+++ b/intern/cycles/kernel/kernel_debug.h
@@ -16,7 +16,7 @@
CCL_NAMESPACE_BEGIN
-ccl_device_inline void debug_data_init(ccl_addr_space DebugData *debug_data)
+ccl_device_inline void debug_data_init(DebugData *debug_data)
{
debug_data->num_bvh_traversal_steps = 0;
}
@@ -24,7 +24,7 @@ ccl_device_inline void debug_data_init(ccl_addr_space DebugData *debug_data)
ccl_device_inline void kernel_write_debug_passes(KernelGlobals *kg,
ccl_global float *buffer,
ccl_addr_space PathState *state,
- ccl_addr_space DebugData *debug_data,
+ DebugData *debug_data,
int sample)
{
int flag = kernel_data.film.pass_flag;
diff --git a/intern/cycles/kernel/kernel_types.h b/intern/cycles/kernel/kernel_types.h
index 434f792..ab20127 100644
--- a/intern/cycles/kernel/kernel_types.h
+++ b/intern/cycles/kernel/kernel_types.h
@@ -962,7 +962,7 @@ typedef struct KernelData {
} KernelData;
#ifdef __KERNEL_DEBUG__
-typedef struct DebugData {
+typedef ccl_addr_space struct DebugData {
// Total number of BVH node traversal steps and primitives intersections
// for the camera rays.
int num_bvh_traversal_steps;
More information about the Bf-blender-cvs
mailing list