diff options
| author | Brecht Van Lommel <brechtvanlommel@pandora.be> | 2011-09-05 16:24:28 +0400 |
|---|---|---|
| committer | Brecht Van Lommel <brechtvanlommel@pandora.be> | 2011-09-05 16:24:28 +0400 |
| commit | 7d9d9fa976589683a7ae880848937b05ade8e74f (patch) | |
| tree | 7081e0b30e185465dcb8614a63a77108f4087666 /intern/cycles/device/device_opencl.cpp | |
| parent | f3ee10ce5ca3050174b84032b6df6edbe548c5f1 (diff) | |
Cycles: use workgroup size from opencl, attempt to fix issue with apple opencl.
Diffstat (limited to 'intern/cycles/device/device_opencl.cpp')
| -rw-r--r-- | intern/cycles/device/device_opencl.cpp | 19 |
1 files changed, 17 insertions, 2 deletions
diff --git a/intern/cycles/device/device_opencl.cpp b/intern/cycles/device/device_opencl.cpp index 956fdfb08c2..c142701c873 100644 --- a/intern/cycles/device/device_opencl.cpp +++ b/intern/cycles/device/device_opencl.cpp @@ -26,6 +26,7 @@ #include "device_intern.h" #include "util_map.h" +#include "util_math.h" #include "util_opencl.h" #include "util_opengl.h" #include "util_path.h" @@ -412,7 +413,14 @@ public: opencl_assert(ciErr); - size_t local_size[2] = {8, 8}; + size_t workgroup_size; + + clGetKernelWorkGroupInfo(ckPathTraceKernel, cdDevice, + CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &workgroup_size, NULL); + + workgroup_size = max(sqrt((double)workgroup_size), 1.0); + + size_t local_size[2] = {workgroup_size, workgroup_size}; size_t global_size[2] = {global_size_round_up(local_size[0], d_w), global_size_round_up(local_size[1], d_h)}; /* run kernel */ @@ -480,7 +488,14 @@ public: opencl_assert(ciErr); - size_t local_size[2] = {8, 8}; + size_t workgroup_size; + + clGetKernelWorkGroupInfo(ckFilmConvertKernel, cdDevice, + CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &workgroup_size, NULL); + + workgroup_size = max(sqrt((double)workgroup_size), 1.0); + + size_t local_size[2] = {workgroup_size, workgroup_size}; size_t global_size[2] = {global_size_round_up(local_size[0], d_w), global_size_round_up(local_size[1], d_h)}; /* run kernel */ |