diff options
| author | Sergey Sharybin <sergey.vfx@gmail.com> | 2015-05-15 10:30:04 +0300 |
|---|---|---|
| committer | Sergey Sharybin <sergey.vfx@gmail.com> | 2015-05-15 11:22:47 +0300 |
| commit | c9e8888f877c7c4332d4d1a1dbb372a3e0b7f28a (patch) | |
| tree | 2cba46389c6755226cf53ac5d22ea6776eb31b46 /intern | |
| parent | 960d7df56f11cd8161bc7ba4478d5c6988c8232f (diff) | |
Cycles: Disable bake OpenCL kernel for NVidia devices prior to sm_30
Driver fails to compile kernel in reasonable time for those devices here,
so for easier testing of the OpenCL split kernel work disabling bake kernel
for now.
Diffstat (limited to 'intern')
| -rw-r--r-- | intern/cycles/kernel/kernel.cl | 13 |
1 files changed, 12 insertions, 1 deletions
diff --git a/intern/cycles/kernel/kernel.cl b/intern/cycles/kernel/kernel.cl index 19e394936ee..cbc0592fe1f 100644 --- a/intern/cycles/kernel/kernel.cl +++ b/intern/cycles/kernel/kernel.cl @@ -102,8 +102,19 @@ __kernel void kernel_ocl_bake( int x = sx + get_global_id(0); - if(x < sx + sw) + if(x < sx + sw) { +#if defined(__KERNEL_OPENCL_NVIDIA__) && __COMPUTE_CAPABILITY__ < 300 + /* NVidia compiler is spending infinite amount of time trying + * to deal with kernel_bake_evaluate() on architectures prior + * to sm_30. + * For now we disable baking kernel for those devices, so at + * least rendering with split kernel could be compiled. + */ + output[x] = make_float4(0.0f, 0.0f, 0.0f, 0.0f); +#else kernel_bake_evaluate(kg, input, output, (ShaderEvalType)type, x, offset, sample); +#endif + } } __kernel void kernel_ocl_convert_to_byte( |