diff options
Diffstat (limited to 'intern/cycles/kernel/device/metal/compat.h')
| -rw-r--r-- | intern/cycles/kernel/device/metal/compat.h | 16 |
1 files changed, 10 insertions, 6 deletions
diff --git a/intern/cycles/kernel/device/metal/compat.h b/intern/cycles/kernel/device/metal/compat.h index b20cfca9a9c..130a9ebafae 100644 --- a/intern/cycles/kernel/device/metal/compat.h +++ b/intern/cycles/kernel/device/metal/compat.h @@ -29,11 +29,12 @@ using namespace metal::raytracing; /* Qualifiers */ -#if defined(__KERNEL_METAL_APPLE__) +/* Inline everything for Apple GPUs. This gives ~1.1x speedup and 10% spill + * reduction for integator_shade_surface. However it comes at the cost of + * longer compile times (~4.5 minutes on M1 Max) and is disabled for that + * reason, until there is a user option to manually enable it. */ -/* Inline everything for Apple GPUs. - * This gives ~1.1x speedup and 10% spill reduction for integator_shade_surface - * at the cost of longer compile times (~4.5 minutes on M1 Max). */ +#if 0 // defined(__KERNEL_METAL_APPLE__) # define ccl_device __attribute__((always_inline)) # define ccl_device_inline __attribute__((always_inline)) @@ -45,8 +46,11 @@ using namespace metal::raytracing; # define ccl_device # define ccl_device_inline ccl_device # define ccl_device_forceinline ccl_device -# define ccl_device_noinline ccl_device __attribute__((noinline)) - +# if defined(__KERNEL_METAL_APPLE__) +# define ccl_device_noinline ccl_device +# else +# define ccl_device_noinline ccl_device __attribute__((noinline)) +# endif #endif #define ccl_device_noinline_cpu ccl_device |