diff options
| author | Brecht Van Lommel <brechtvanlommel@gmail.com> | 2014-05-27 15:20:07 +0400 |
|---|---|---|
| committer | Brecht Van Lommel <brechtvanlommel@gmail.com> | 2014-05-27 17:11:32 +0400 |
| commit | 69c7522b2463245ef16ebcf2806645c78e83b4df (patch) | |
| tree | 96721c9d2bc755d68cf9eeded77255f4c58018f3 /intern/cycles/kernel/kernel.cu | |
| parent | bc9e66f0830627e807f720dca4b9d5d8d39e732a (diff) | |
Fix T40379: world MIS causing too much CUDA memory usage.
The kernel for baking the world texture was the same as the one used for
baking. Now that's separate which allows the kernel to reserve much less
memory.
Diffstat (limited to 'intern/cycles/kernel/kernel.cu')
| -rw-r--r-- | intern/cycles/kernel/kernel.cu | 9 |
1 files changed, 9 insertions, 0 deletions
diff --git a/intern/cycles/kernel/kernel.cu b/intern/cycles/kernel/kernel.cu index b9b41f755ba..bb20819f6fc 100644 --- a/intern/cycles/kernel/kernel.cu +++ b/intern/cycles/kernel/kernel.cu @@ -153,5 +153,14 @@ kernel_cuda_shader(uint4 *input, float4 *output, int type, int sx) kernel_shader_evaluate(NULL, input, output, (ShaderEvalType)type, x); } +extern "C" __global__ void +CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS) +kernel_cuda_bake(uint4 *input, float4 *output, int type, int sx) +{ + int x = sx + blockDim.x*blockIdx.x + threadIdx.x; + + kernel_bake_evaluate(NULL, input, output, (ShaderEvalType)type, x); +} + #endif |