diff options
| author | Bill Hollings <bill.hollings@brenwill.com> | 2021-09-23 01:58:31 +0300 |
|---|---|---|
| committer | Bill Hollings <bill.hollings@brenwill.com> | 2021-09-23 01:58:31 +0300 |
| commit | 40141ffddfe6a00fba0cce7e1b5f3759b8d520b8 (patch) | |
| tree | 9f97661170f9ccdb3de82110a0df7d3738505f61 /reference/shaders-msl-no-opt | |
| parent | b81334a513e357cc15d6e93b9395b7733056f993 (diff) | |
MSL: Selectively enable fast-math in MSL code to match Vulkan CTS results.
Based on CTS testing, math optimizations between MSL and Vulkan are inconsistent.
In some cases, enabling MSL's fast-math compilation option matches Vulkan's math
results. In other cases, disabling it does. Broadly enabling or disabling fast-math
across all shaders results in some CTS test failures either way.
To fix this, selectively enable/disable fast-math optimizations in the MSL code,
using metal::fast and metal::precise function namespaces, where supported, and
the [[clang::optnone]] function attribute otherwise.
Adjust SPIRV-Cross unit test reference shaders to accommodate these changes.
Diffstat (limited to 'reference/shaders-msl-no-opt')
| -rw-r--r-- | reference/shaders-msl-no-opt/comp/glsl.std450.comp | 12 | ||||
| -rw-r--r-- | reference/shaders-msl-no-opt/frag/fp16.desktop.invalid.frag | 10 |
2 files changed, 11 insertions, 11 deletions
diff --git a/reference/shaders-msl-no-opt/comp/glsl.std450.comp b/reference/shaders-msl-no-opt/comp/glsl.std450.comp index 3c505abe..1b3295f4 100644 --- a/reference/shaders-msl-no-opt/comp/glsl.std450.comp +++ b/reference/shaders-msl-no-opt/comp/glsl.std450.comp @@ -170,7 +170,7 @@ float2x2 spvInverse2x2(float2x2 m) } template<typename T> -inline T spvReflect(T i, T n) +[[clang::optnone]] T spvReflect(T i, T n) { return i - T(2) * i * n * n; } @@ -217,13 +217,13 @@ kernel void main0(device SSBO& _19 [[buffer(0)]]) _19.res = asin(((device float*)&_19.f32)[0u]); _19.res = acos(((device float*)&_19.f32)[0u]); _19.res = atan(((device float*)&_19.f32)[0u]); - _19.res = sinh(((device float*)&_19.f32)[0u]); - _19.res = cosh(((device float*)&_19.f32)[0u]); - _19.res = tanh(((device float*)&_19.f32)[0u]); + _19.res = fast::sinh(((device float*)&_19.f32)[0u]); + _19.res = fast::cosh(((device float*)&_19.f32)[0u]); + _19.res = precise::tanh(((device float*)&_19.f32)[0u]); _19.res = asinh(((device float*)&_19.f32)[0u]); _19.res = acosh(((device float*)&_19.f32)[0u]); _19.res = atanh(((device float*)&_19.f32)[0u]); - _19.res = atan2(((device float*)&_19.f32)[0u], ((device float*)&_19.f32)[1u]); + _19.res = precise::atan2(((device float*)&_19.f32)[0u], ((device float*)&_19.f32)[1u]); _19.res = pow(((device float*)&_19.f32)[0u], ((device float*)&_19.f32)[1u]); _19.res = exp(((device float*)&_19.f32)[0u]); _19.res = log(((device float*)&_19.f32)[0u]); @@ -239,7 +239,7 @@ kernel void main0(device SSBO& _19 [[buffer(0)]]) _19.res = spvRefract(((device float*)&_19.f32)[0u], ((device float*)&_19.f32)[1u], ((device float*)&_19.f32)[2u]); _19.res = length(_19.f32.xy); _19.res = distance(_19.f32.xy, _19.f32.zw); - float2 v2 = normalize(_19.f32.xy); + float2 v2 = fast::normalize(_19.f32.xy); v2 = faceforward(_19.f32.xy, _19.f32.yz, _19.f32.zw); v2 = reflect(_19.f32.xy, _19.f32.zw); v2 = refract(_19.f32.xy, _19.f32.yz, ((device float*)&_19.f32)[3u]); diff --git a/reference/shaders-msl-no-opt/frag/fp16.desktop.invalid.frag b/reference/shaders-msl-no-opt/frag/fp16.desktop.invalid.frag index 3bf42962..de53d681 100644 --- a/reference/shaders-msl-no-opt/frag/fp16.desktop.invalid.frag +++ b/reference/shaders-msl-no-opt/frag/fp16.desktop.invalid.frag @@ -94,11 +94,11 @@ void test_builtins(thread half4& v4, thread half3& v3, thread half& v1) res = cos(v4); res = tan(v4); res = asin(v4); - res = atan2(v4, v3.xyzz); + res = precise::atan2(v4, v3.xyzz); res = atan(v4); - res = sinh(v4); - res = cosh(v4); - res = tanh(v4); + res = fast::sinh(v4); + res = fast::cosh(v4); + res = precise::tanh(v4); res = asinh(v4); res = acosh(v4); res = atanh(v4); @@ -143,7 +143,7 @@ void test_builtins(thread half4& v4, thread half3& v3, thread half& v1) t0 = distance(v4, v4); t0 = dot(v4, v4); half3 res3 = cross(v3, v3); - res = normalize(v4); + res = fast::normalize(v4); res = faceforward(v4, v4, v4); res = reflect(v4, v4); res = refract(v4, v4, v1); |