Cr8-xyz/blender
1
0
Fork 0
Code Issues Pull Requests Actions 1 Packages Projects Releases Wiki Activity

Cycles: Remove unused argument from the split kernel functions

Browse Source 
Should be no functional changes, just simplifies operation with kernels.
This commit is contained in:
Sergey Sharybin
2015-11-01 16:51:32 +05:00
parent 84e8b05e97
commit dc9e0b819b
19 changed files with 16 additions and 48 deletions
Download Patch File Download Diff File Expand all files Collapse all files

3
intern/cycles/kernel/split/kernel_background_buffer_update.h
View File

@@ -43,7 +43,7 @@
* Queue_index (QUEUE_ACTIVE_AND_REGENERATED_RAYS) ------| |--- work_array
* parallel_samples -------------------------------------| |--- PathState_coop
* end_sample -------------------------------------------| |--- throughput_coop
* kg (globals + data) ----------------------------------| |--- rng_coop
* kg (globals) -----------------------------------------| |--- rng_coop
* rng_state --------------------------------------------| |--- Ray
* PathRadiance_coop ------------------------------------| |
* sw ---------------------------------------------------| |
@@ -71,7 +71,6 @@
*/
ccl_device char kernel_background_buffer_update(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd,
ccl_global float *per_sample_output_buffers,
ccl_global uint *rng_state,

3
intern/cycles/kernel/split/kernel_direct_lighting.h
View File

@@ -33,7 +33,7 @@
* sd -----------------------------------------------| |--- LightRay_coop
* ray_state ----------------------------------------| |--- ray_state
* Queue_data (QUEUE_ACTIVE_AND_REGENERATED_RAYS) ---| |
* kg (globals + data) ------------------------------| |
* kg (globals) -------------------------------------| |
* queuesize ----------------------------------------| |
*
* note on sd_DL : sd_DL is neither input nor output to this kernel; sd_DL is filled and consumed in this kernel itself.
@@ -50,7 +50,6 @@
*/
ccl_device char kernel_direct_lighting(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd, /* Required for direct lighting */
ShaderData *sd_DL, /* Required for direct lighting */
ccl_global uint *rng_coop, /* Required for direct lighting */

3
intern/cycles/kernel/split/kernel_holdout_emission_blurring_pathtermination_ao.h
View File

@@ -40,7 +40,7 @@
* ray_state --------------------------------------------| |--- ray_state
* Queue_data (QUEUE_ACTIVE_AND_REGENERATED_RAYS) -------| |--- Queue_data (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)
* Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS) ---| |--- AOAlpha_coop
* kg (globals + data) ----------------------------------| |--- AOBSDF_coop
* kg (globals) -----------------------------------------| |--- AOBSDF_coop
* parallel_samples -------------------------------------| |--- AOLightRay_coop
* per_sample_output_buffers ----------------------------| |
* sw ---------------------------------------------------| |
@@ -72,7 +72,6 @@
*/
ccl_device void kernel_holdout_emission_blurring_pathtermination_ao(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd, /* Required throughout the kernel except probabilistic path termination and AO */
ccl_global float *per_sample_output_buffers,
ccl_global uint *rng_coop, /* Required for "kernel_write_data_passes" and AO */

3
intern/cycles/kernel/split/kernel_lamp_emission.h
View File

@@ -26,7 +26,7 @@
* Throughput_coop ------------------------------------|--- kernel_lamp_emission --|--- PathRadiance_coop
* Ray_coop -------------------------------------------| |--- Queue_data(QUEUE_ACTIVE_AND_REGENERATED_RAYS)
* PathState_coop -------------------------------------| |--- Queue_index(QUEUE_ACTIVE_AND_REGENERATED_RAYS)
* kg (globals + data) --------------------------------| |
* kg (globals) ---------------------------------------| |
* Intersection_coop ----------------------------------| |
* ray_state ------------------------------------------| |
* Queue_data (QUEUE_ACTIVE_AND_REGENERATED_RAYS) -----| |
@@ -41,7 +41,6 @@
*/
ccl_device void kernel_lamp_emission(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd, /* Required for lamp emission */
ccl_global float3 *throughput_coop, /* Required for lamp emission */
PathRadiance *PathRadiance_coop, /* Required for lamp emission */

3
intern/cycles/kernel/split/kernel_next_iteration_setup.h
View File

@@ -35,7 +35,7 @@
* Queue_data (QUEUE_ACTIVE_AND_REGENERATD_RAYS) --------| |--- Ray_coop
* Queue_index (QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS) ---| |--- use_queues_flag
* Ray_coop ---------------------------------------------| |
* kg (globals + data) ----------------------------------| |
* kg (globals) -----------------------------------------| |
* LightRay_dl_coop -------------------------------------|
* ISLamp_coop ------------------------------------------|
* BSDFEval_coop ----------------------------------------|
@@ -61,7 +61,6 @@
*/
ccl_device char kernel_next_iteration_setup(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd, /* Required for setting up ray for next iteration */
ccl_global uint *rng_coop, /* Required for setting up ray for next iteration */
ccl_global float3 *throughput_coop, /* Required for setting up ray for next iteration */

3
intern/cycles/kernel/split/kernel_scene_intersect.h
View File

@@ -32,7 +32,7 @@
* use_queues_flag --------------------------------| |
* parallel_samples -------------------------------| |
* QueueData(QUEUE_ACTIVE_AND_REGENERATED_RAYS) ---| |
* kg (data + globals) ----------------------------| |
* kg (globals) -----------------------------------| |
* rng_coop ---------------------------------------| |
* sw ---------------------------------------------| |
* sh ---------------------------------------------| |
@@ -64,7 +64,6 @@
ccl_device void kernel_scene_intersect(
KernelGlobals *kg,
ccl_constant KernelData *data,
ccl_global uint *rng_coop,
ccl_global Ray *Ray_coop, /* Required for scene_intersect */
ccl_global PathState *PathState_coop, /* Required for scene_intersect */

3
intern/cycles/kernel/split/kernel_shader_eval.h
View File

@@ -30,7 +30,7 @@
* Queue_data (QUEUE_ACTIVE_AND_REGENERATD_RAYS)-------| |
* Queue_index(QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS)---| |
* ray_state ------------------------------------------| |
* kg (globals + data) --------------------------------| |
* kg (globals) ---------------------------------------| |
* queuesize ------------------------------------------| |
*
* Note on Queues :
@@ -46,7 +46,6 @@
*/
ccl_device void kernel_shader_eval(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd, /* Output ShaderData structure to be filled */
ccl_global uint *rng_coop, /* Required for rbsdf calculation */
ccl_global Ray *Ray_coop, /* Required for setting up shader from ray */

3
intern/cycles/kernel/split/kernel_shadow_blocked.h
View File

@@ -31,7 +31,7 @@
QUEUE_SHADOW_RAY_CAST_DL_RAYS) -------| |
* Queue_index(QUEUE_SHADOW_RAY_CAST_AO_RAYS&
QUEUE_SHADOW_RAY_CAST_DL_RAYS) -------| |
* kg (globals + data) -----------------------------| |
* kg (globals) ------------------------------------| |
* queuesize ---------------------------------------| |
*
* Note on sd_shadow : sd_shadow is neither input nor output to this kernel. sd_shadow is filled and consumed in this kernel itself.
@@ -47,7 +47,6 @@
*/
ccl_device void kernel_shadow_blocked(
KernelGlobals *kg,
ccl_constant KernelData *data,
ShaderData *sd_shadow, /* Required for shadow blocked */
ccl_global PathState *PathState_coop, /* Required for shadow blocked */
ccl_global Ray *LightRay_dl_coop, /* Required for direct lighting's shadow blocked */

10
intern/cycles/kernel/split/kernel_sum_all_radiance.h
View File

@@ -24,7 +24,7 @@
* by all different samples and stores them in the RenderTile's output buffer.
*/
ccl_device void kernel_sum_all_radiance(
ccl_constant KernelData *data, /* To get pass_stride to offet into buffer */
KernelGlobals *kg, /* To get pass_stride to offet into buffer */
ccl_global float *buffer, /* Output buffer of RenderTile */
ccl_global float *per_sample_output_buffer, /* Radiance contributed by all samples */
int parallel_samples, int sw, int sh, int stride,
@@ -37,14 +37,14 @@ ccl_device void kernel_sum_all_radiance(
int y = get_global_id(1);
if(x < sw && y < sh) {
buffer += ((buffer_offset_x + x) + (buffer_offset_y + y) * buffer_stride) * (data->film.pass_stride);
per_sample_output_buffer += ((x + y * stride) * parallel_samples) * (data->film.pass_stride);
buffer += ((buffer_offset_x + x) + (buffer_offset_y + y) * buffer_stride) * (kg->data->film.pass_stride);
per_sample_output_buffer += ((x + y * stride) * parallel_samples) * (kg->data->film.pass_stride);
int sample_stride = (data->film.pass_stride);
int sample_stride = (kg->data->film.pass_stride);
int sample_iterator = 0;
int pass_stride_iterator = 0;
int num_floats = data->film.pass_stride;
int num_floats = kg->data->film.pass_stride;
for(sample_iterator = 0; sample_iterator < parallel_samples; sample_iterator++) {
for(pass_stride_iterator = 0; pass_stride_iterator < num_floats; pass_stride_iterator++) {