diff options
author | Antonio Vazquez <blendergit@gmail.com> | 2022-06-21 20:09:54 +0300 |
---|---|---|
committer | Antonio Vazquez <blendergit@gmail.com> | 2022-06-21 20:09:54 +0300 |
commit | dd6d0aefcc56288bfe154fe06df56163d6961372 (patch) | |
tree | 7b74c26bcd48d0fbd844fd8b286c9a399813b203 /intern/cycles/kernel/device/cuda | |
parent | c27a395d8cd8b58b3155cfdc26ae35198fbcdb14 (diff) | |
parent | 714001683888770f9d870ba73edd65f4c2757bda (diff) |
Merge branch 'master' into asset-greasepencilasset-greasepencil
Diffstat (limited to 'intern/cycles/kernel/device/cuda')
-rw-r--r-- | intern/cycles/kernel/device/cuda/globals.h | 26 |
1 files changed, 16 insertions, 10 deletions
diff --git a/intern/cycles/kernel/device/cuda/globals.h b/intern/cycles/kernel/device/cuda/globals.h index e77fcd2b424..f5f7bcf58ee 100644 --- a/intern/cycles/kernel/device/cuda/globals.h +++ b/intern/cycles/kernel/device/cuda/globals.h @@ -20,18 +20,24 @@ struct KernelGlobalsGPU { }; typedef ccl_global const KernelGlobalsGPU *ccl_restrict KernelGlobals; -/* Global scene data and textures */ -__constant__ KernelData __data; -#define KERNEL_TEX(type, name) const __constant__ __device__ type *name; -#include "kernel/textures.h" +struct KernelParamsCUDA { + /* Global scene data and textures */ + KernelData data; +#define KERNEL_DATA_ARRAY(type, name) const type *name; +#include "kernel/data_arrays.h" + + /* Integrator state */ + IntegratorStateGPU integrator_state; +}; -/* Integrator state */ -__constant__ IntegratorStateGPU __integrator_state; +#ifdef __KERNEL_GPU__ +__constant__ KernelParamsCUDA kernel_params; +#endif /* Abstraction macros */ -#define kernel_data __data -#define kernel_tex_fetch(t, index) t[(index)] -#define kernel_tex_array(t) (t) -#define kernel_integrator_state __integrator_state +#define kernel_data kernel_params.data +#define kernel_data_fetch(name, index) kernel_params.name[(index)] +#define kernel_data_array(name) (kernel_params.name) +#define kernel_integrator_state kernel_params.integrator_state CCL_NAMESPACE_END |