[Bf-blender-cvs] [1c90f8209d4] master: Cycles: fix rendering with Nishita Sky Texture on Intel Arc GPUs

[Xavier Hallade]

Commit: 1c90f8209d4191792267d4f7df821c0911afe2d7
Author: Xavier Hallade
Date:   Mon Jan 23 16:22:24 2023 +0100
Branches: master
https://developer.blender.org/rB1c90f8209d4191792267d4f7df821c0911afe2d7

Cycles: fix rendering with Nishita Sky Texture on Intel Arc GPUs

Speckles and missing lights were experienced in scenes with Nishita Sky
Texture and a Sun Size smaller than 1.5°, such as in Lone Monk and Attic
scenes.
Increasing the precision of cosf fixes it.

===================================================================

M	intern/cycles/kernel/device/oneapi/compat.h

===================================================================

diff --git a/intern/cycles/kernel/device/oneapi/compat.h b/intern/cycles/kernel/device/oneapi/compat.h
index 0691c01b3b5..a43b68e5b49 100644
--- a/intern/cycles/kernel/device/oneapi/compat.h
+++ b/intern/cycles/kernel/device/oneapi/compat.h
@@ -195,7 +195,15 @@ using sycl::half;
 #define fmodf(x, y) sycl::fmod((x), (y))
 #define lgammaf(x) sycl::lgamma((x))
 
-#define cosf(x) sycl::native::cos(((float)(x)))
+/* sycl::native::cos precision is not sufficient and -ffast-math lets
+ * the current DPC++ compiler overload sycl::cos with it.
+ * We work around this issue by directly calling the spirv implementation which
+ * provides greater precision. */
+#if defined(__SYCL_DEVICE_ONLY__) && defined(__SPIR__)
+#  define cosf(x) __spirv_ocl_cos(((float)(x)))
+#else
+#  define cosf(x) sycl::cos(((float)(x)))
+#endif
 #define sinf(x) sycl::native::sin(((float)(x)))
 #define powf(x, y) sycl::native::powr(((float)(x)), ((float)(y)))
 #define tanf(x) sycl::native::tan(((float)(x)))