diff options
11 files changed, 11 insertions, 11 deletions
diff --git a/intern/cycles/kernel/split/kernel_background_buffer_update.h b/intern/cycles/kernel/split/kernel_background_buffer_update.h index 95de1a4b2a9..32c7e6f8c0a 100644 --- a/intern/cycles/kernel/split/kernel_background_buffer_update.h +++ b/intern/cycles/kernel/split/kernel_background_buffer_update.h @@ -70,7 +70,7 @@ * QUEUE_ACTIVE_AND_REGENERATED_RAYS will be filled with RAY_ACTIVE and RAY_REGENERATED rays * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be empty */ -ccl_device void kernel_background_buffer_update( +__kernel void kernel_background_buffer_update( ccl_global char *globals, ccl_constant KernelData *data, ccl_global char *shader_data, diff --git a/intern/cycles/kernel/split/kernel_data_init.h b/intern/cycles/kernel/split/kernel_data_init.h index b7a4d847d03..006f2c8e4df 100644 --- a/intern/cycles/kernel/split/kernel_data_init.h +++ b/intern/cycles/kernel/split/kernel_data_init.h @@ -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; */ -ccl_device void kernel_data_init( +__kernel void kernel_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/split/kernel_direct_lighting.h b/intern/cycles/kernel/split/kernel_direct_lighting.h index 6b83d892057..91c3ef11682 100644 --- a/intern/cycles/kernel/split/kernel_direct_lighting.h +++ b/intern/cycles/kernel/split/kernel_direct_lighting.h @@ -49,7 +49,7 @@ * QUEUE_SHADOW_RAY_CAST_DL_RAYS queue will be filled with rays for which a shadow_blocked function must be executed, after this * kernel call. Before this kernel call the QUEUE_SHADOW_RAY_CAST_DL_RAYS will be empty. */ -ccl_device void kernel_direct_lighting( +__kernel void kernel_direct_lighting( ccl_global char *globals, ccl_constant KernelData *data, ccl_global char *shader_data, /* Required for direct lighting */ diff --git a/intern/cycles/kernel/split/kernel_holdout_emission_blurring_pathtermination_ao.h b/intern/cycles/kernel/split/kernel_holdout_emission_blurring_pathtermination_ao.h index 393ea4bcadc..174070ad5bb 100644 --- a/intern/cycles/kernel/split/kernel_holdout_emission_blurring_pathtermination_ao.h +++ b/intern/cycles/kernel/split/kernel_holdout_emission_blurring_pathtermination_ao.h @@ -72,7 +72,7 @@ * QUEUE_SHADOW_RAY_CAST_AO_RAYS will be filled with rays marked with flag RAY_SHADOW_RAY_CAST_AO */ -ccl_device void kernel_holdout_emission_blurring_pathtermination_ao( +__kernel void kernel_holdout_emission_blurring_pathtermination_ao( ccl_global char *globals, ccl_constant KernelData *data, ccl_global char *shader_data, /* Required throughout the kernel except probabilistic path termination and AO */ diff --git a/intern/cycles/kernel/split/kernel_lamp_emission.h b/intern/cycles/kernel/split/kernel_lamp_emission.h index f400a99e229..b804bfc8630 100644 --- a/intern/cycles/kernel/split/kernel_lamp_emission.h +++ b/intern/cycles/kernel/split/kernel_lamp_emission.h @@ -40,7 +40,7 @@ * * note : shader_data is neither input nor output. Its just filled and consumed in the same, kernel_lamp_emission, kernel. */ -ccl_device void kernel_lamp_emission( +__kernel void kernel_lamp_emission( ccl_global char *globals, ccl_constant KernelData *data, ccl_global char *shader_data, /* Required for lamp emission */ diff --git a/intern/cycles/kernel/split/kernel_next_iteration_setup.h b/intern/cycles/kernel/split/kernel_next_iteration_setup.h index 343dbb06e99..6ce56e45733 100644 --- a/intern/cycles/kernel/split/kernel_next_iteration_setup.h +++ b/intern/cycles/kernel/split/kernel_next_iteration_setup.h @@ -61,7 +61,7 @@ * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be filled with RAY_TO_REGENERATE and more RAY_UPDATE_BUFFER rays */ -ccl_device void kernel_next_iteration_setup( +__kernel void kernel_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/split/kernel_queue_enqueue.h b/intern/cycles/kernel/split/kernel_queue_enqueue.h index 9bcf8f540b4..5a9838f44d7 100644 --- a/intern/cycles/kernel/split/kernel_queue_enqueue.h +++ b/intern/cycles/kernel/split/kernel_queue_enqueue.h @@ -52,7 +52,7 @@ * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be filled with RAY_TO_REGENERATE, RAY_UPDATE_BUFFER, RAY_HIT_BACKGROUND rays. */ -ccl_device void kernel_queue_enqueue( +__kernel void kernel_queue_enqueue( ccl_global int *Queue_data, /* Queue memory */ ccl_global int *Queue_index, /* Tracks the number of elements in each queue */ ccl_global char *ray_state, /* Denotes the state of each ray */ diff --git a/intern/cycles/kernel/split/kernel_scene_intersect.h b/intern/cycles/kernel/split/kernel_scene_intersect.h index 01e0b1fd19e..6cb3cdcc182 100644 --- a/intern/cycles/kernel/split/kernel_scene_intersect.h +++ b/intern/cycles/kernel/split/kernel_scene_intersect.h @@ -63,7 +63,7 @@ * QUEUE_HITBF_BUFF_UPDATE_TOREGEN_RAYS - no change */ -ccl_device void kernel_scene_intersect( +__kernel void kernel_scene_intersect( ccl_global char *globals, ccl_constant KernelData *data, ccl_global uint *rng_coop, diff --git a/intern/cycles/kernel/split/kernel_shader_eval.h b/intern/cycles/kernel/split/kernel_shader_eval.h index 0a8d77f52b0..924c7fab2a3 100644 --- a/intern/cycles/kernel/split/kernel_shader_eval.h +++ b/intern/cycles/kernel/split/kernel_shader_eval.h @@ -46,7 +46,7 @@ * QUEUE_HITBG_BUFF_UPDATE_TOREGEN_RAYS will be filled with RAY_TO_REGENERATE rays */ -ccl_device void kernel_shader_eval( +__kernel void kernel_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/split/kernel_shadow_blocked.h b/intern/cycles/kernel/split/kernel_shadow_blocked.h index 71fab19518c..52bd9eb3bbc 100644 --- a/intern/cycles/kernel/split/kernel_shadow_blocked.h +++ b/intern/cycles/kernel/split/kernel_shadow_blocked.h @@ -47,7 +47,7 @@ * QUEUE_SHADOW_RAY_CAST_AO_RAYS and QUEUE_SHADOW_RAY_CAST_DL_RAYS will be empty at kernel exit. */ -ccl_device void kernel_shadow_blocked( +__kernel void kernel_shadow_blocked( ccl_global char *globals, ccl_constant KernelData *data, ccl_global char *shader_shadow, /* Required for shadow blocked */ diff --git a/intern/cycles/kernel/split/kernel_sum_all_radiance.h b/intern/cycles/kernel/split/kernel_sum_all_radiance.h index eeb7da76e73..faa4162b46f 100644 --- a/intern/cycles/kernel/split/kernel_sum_all_radiance.h +++ b/intern/cycles/kernel/split/kernel_sum_all_radiance.h @@ -25,7 +25,7 @@ * by all different samples and stores them in the RenderTile's output buffer. */ -ccl_device void kernel_sum_all_radiance( +__kernel void kernel_sum_all_radiance( ccl_constant KernelData *data, /* 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 */ |