diff options
author | Mai Lavelle <mai.lavelle@gmail.com> | 2017-03-11 13:23:11 +0300 |
---|---|---|
committer | Mai Lavelle <mai.lavelle@gmail.com> | 2017-03-11 13:39:28 +0300 |
commit | 96868a39419f1c9a8962c56e02480fabbf1e5156 (patch) | |
tree | 64128db60a690d0223dad4e6e4073230ffd174ac /intern/cycles/kernel | |
parent | 5afe4c787f0ed3ac30f7609c7f07c5092a20eac9 (diff) |
Fix T50888: Numeric overflow in split kernel state buffer size calculation
Overflow led to the state buffer being too small and the split kernel to
get stuck doing nothing forever.
Diffstat (limited to 'intern/cycles/kernel')
-rw-r--r-- | intern/cycles/kernel/kernels/cuda/kernel_split.cu | 2 | ||||
-rw-r--r-- | intern/cycles/kernel/kernels/opencl/kernel_state_buffer_size.cl | 2 | ||||
-rw-r--r-- | intern/cycles/kernel/split/kernel_split_data.h | 4 |
3 files changed, 4 insertions, 4 deletions
diff --git a/intern/cycles/kernel/kernels/cuda/kernel_split.cu b/intern/cycles/kernel/kernels/cuda/kernel_split.cu index 6c508c2cd79..fbdf79697d5 100644 --- a/intern/cycles/kernel/kernels/cuda/kernel_split.cu +++ b/intern/cycles/kernel/kernels/cuda/kernel_split.cu @@ -46,7 +46,7 @@ /* kernels */ extern "C" __global__ void CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS) -kernel_cuda_state_buffer_size(uint num_threads, uint *size) +kernel_cuda_state_buffer_size(uint num_threads, uint64_t *size) { *size = split_data_buffer_size(NULL, num_threads); } diff --git a/intern/cycles/kernel/kernels/opencl/kernel_state_buffer_size.cl b/intern/cycles/kernel/kernels/opencl/kernel_state_buffer_size.cl index 0a1843ff8bd..4c9bf63ef51 100644 --- a/intern/cycles/kernel/kernels/opencl/kernel_state_buffer_size.cl +++ b/intern/cycles/kernel/kernels/opencl/kernel_state_buffer_size.cl @@ -21,7 +21,7 @@ __kernel void kernel_ocl_path_trace_state_buffer_size( KernelGlobals *kg, ccl_constant KernelData *data, uint num_threads, - ccl_global uint *size) + ccl_global uint64_t *size) { kg->data = data; *size = split_data_buffer_size(kg, num_threads); diff --git a/intern/cycles/kernel/split/kernel_split_data.h b/intern/cycles/kernel/split/kernel_split_data.h index 81dcdbaedde..d319514c190 100644 --- a/intern/cycles/kernel/split/kernel_split_data.h +++ b/intern/cycles/kernel/split/kernel_split_data.h @@ -22,11 +22,11 @@ CCL_NAMESPACE_BEGIN -ccl_device_inline size_t split_data_buffer_size(KernelGlobals *kg, size_t num_elements) +ccl_device_inline uint64_t split_data_buffer_size(KernelGlobals *kg, size_t num_elements) { (void)kg; /* Unused on CPU. */ - size_t size = 0; + uint64_t size = 0; #define SPLIT_DATA_ENTRY(type, name, num) + align_up(num_elements * num * sizeof(type), 16) size = size SPLIT_DATA_ENTRIES; #undef SPLIT_DATA_ENTRY |