diff options
| author | Sergey Sharybin <sergey.vfx@gmail.com> | 2016-08-23 17:32:09 +0300 |
|---|---|---|
| committer | Sergey Sharybin <sergey.vfx@gmail.com> | 2016-08-23 17:32:09 +0300 |
| commit | 36b80b185f2c01e72ba251b852b517b02c806c73 (patch) | |
| tree | d5fd0cae9606a0133a3d6fac642b418aae181215 | |
| parent | eb5cd578b8cb9839b4b1fb8b6e9f56bbaad7bfae (diff) | |
Cycles: Fix/workaround for missing AA on Pascal (GTX10x0) cards
| -rw-r--r-- | intern/cycles/kernel/kernel_random.h | 8 |
1 files changed, 7 insertions, 1 deletions
diff --git a/intern/cycles/kernel/kernel_random.h b/intern/cycles/kernel/kernel_random.h index 731dc0407c5..b534d9950c5 100644 --- a/intern/cycles/kernel/kernel_random.h +++ b/intern/cycles/kernel/kernel_random.h @@ -132,7 +132,13 @@ ccl_device_inline float path_rng_1D(KernelGlobals *kg, ccl_addr_space RNG *rng, #endif } -ccl_device_inline void path_rng_2D(KernelGlobals *kg, ccl_addr_space RNG *rng, int sample, int num_samples, int dimension, float *fx, float *fy) +/* Temporary workaround for Pascal cards, otherwise AA does not work properly. */ +#if defined(__KERNEL_GPU__) && __CUDA_ARCH__ >= 600 +__device__ __forceinline__ +#else +ccl_device_inline +#endif +void path_rng_2D(KernelGlobals *kg, ccl_addr_space RNG *rng, int sample, int num_samples, int dimension, float *fx, float *fy) { #ifdef __CMJ__ if(kernel_data.integrator.sampling_pattern == SAMPLING_PATTERN_CMJ) { |