diff options
author | Hans-Kristian Arntzen <post@arntzen-software.no> | 2019-02-20 16:58:44 +0300 |
---|---|---|
committer | GitHub <noreply@github.com> | 2019-02-20 16:58:44 +0300 |
commit | 58f264c99de40b4047a2cf35ab8ca707b4ec6b80 (patch) | |
tree | 9937621481c7fafa262b148ec47a61e2f028a647 | |
parent | f10d37a64a8b53f188f43b094fb975aaccccb21d (diff) | |
parent | 4ef51331b26a96fe5435bd665bf1fbc7aaac3e59 (diff) |
Merge pull request #865 from KhronosGroup/fix-8632019-02-20
Always value-cast FP16 constants instead of using literals.
-rw-r--r-- | reference/opt/shaders-msl/comp/bitcast-16bit-1.invalid.comp | 2 | ||||
-rw-r--r-- | reference/opt/shaders-msl/frag/16bit-constants.frag | 2 | ||||
-rw-r--r-- | reference/opt/shaders/comp/bitcast-16bit-1.invalid.comp | 2 | ||||
-rw-r--r-- | reference/opt/shaders/desktop-only/frag/fp16.invalid.desktop.frag | 20 | ||||
-rw-r--r-- | reference/opt/shaders/frag/16bit-constants.frag | 2 | ||||
-rw-r--r-- | reference/shaders-msl/comp/bitcast-16bit-1.invalid.comp | 2 | ||||
-rw-r--r-- | reference/shaders-msl/frag/16bit-constants.frag | 2 | ||||
-rw-r--r-- | reference/shaders-msl/frag/fp16.desktop.invalid.frag | 20 | ||||
-rw-r--r-- | reference/shaders/comp/bitcast-16bit-1.invalid.comp | 2 | ||||
-rw-r--r-- | reference/shaders/desktop-only/frag/fp16.invalid.desktop.frag | 20 | ||||
-rw-r--r-- | reference/shaders/frag/16bit-constants.frag | 2 | ||||
-rw-r--r-- | spirv_glsl.cpp | 58 | ||||
-rw-r--r-- | spirv_glsl.hpp | 1 | ||||
-rw-r--r-- | spirv_hlsl.cpp | 1 | ||||
-rw-r--r-- | spirv_msl.cpp | 1 |
15 files changed, 57 insertions, 80 deletions
diff --git a/reference/opt/shaders-msl/comp/bitcast-16bit-1.invalid.comp b/reference/opt/shaders-msl/comp/bitcast-16bit-1.invalid.comp index f91e5fea..ad9733a8 100644 --- a/reference/opt/shaders-msl/comp/bitcast-16bit-1.invalid.comp +++ b/reference/opt/shaders-msl/comp/bitcast-16bit-1.invalid.comp @@ -15,7 +15,7 @@ struct SSBO1 kernel void main0(device SSBO0& _25 [[buffer(0)]], device SSBO1& _39 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]]) { - _39.outputs[gl_GlobalInvocationID.x].x = int(as_type<uint>(as_type<half2>(_25.inputs[gl_GlobalInvocationID.x].xy) + half2(1.0h))); + _39.outputs[gl_GlobalInvocationID.x].x = int(as_type<uint>(as_type<half2>(_25.inputs[gl_GlobalInvocationID.x].xy) + half2(half(1.0)))); _39.outputs[gl_GlobalInvocationID.x].y = as_type<int>(_25.inputs[gl_GlobalInvocationID.x].zw); _39.outputs[gl_GlobalInvocationID.x].z = int(as_type<uint>(ushort2(_25.inputs[gl_GlobalInvocationID.x].xy))); } diff --git a/reference/opt/shaders-msl/frag/16bit-constants.frag b/reference/opt/shaders-msl/frag/16bit-constants.frag index 2aae1e8a..56c7ea5d 100644 --- a/reference/opt/shaders-msl/frag/16bit-constants.frag +++ b/reference/opt/shaders-msl/frag/16bit-constants.frag @@ -13,7 +13,7 @@ struct main0_out fragment main0_out main0() { main0_out out = {}; - out.foo = 1.0h; + out.foo = half(1.0); out.bar = 2; out.baz = 3u; return out; diff --git a/reference/opt/shaders/comp/bitcast-16bit-1.invalid.comp b/reference/opt/shaders/comp/bitcast-16bit-1.invalid.comp index a4014e89..501f9795 100644 --- a/reference/opt/shaders/comp/bitcast-16bit-1.invalid.comp +++ b/reference/opt/shaders/comp/bitcast-16bit-1.invalid.comp @@ -27,7 +27,7 @@ void main() { uint ident = gl_GlobalInvocationID.x; f16vec2 a = int16BitsToFloat16(_25.inputs[ident].xy); - _39.outputs[ident].x = int(packFloat2x16(a + f16vec2(1.0hf))); + _39.outputs[ident].x = int(packFloat2x16(a + f16vec2(float16_t(1.0)))); _39.outputs[ident].y = packInt2x16(_25.inputs[ident].zw); _39.outputs[ident].z = int(packUint2x16(u16vec2(_25.inputs[ident].xy))); } diff --git a/reference/opt/shaders/desktop-only/frag/fp16.invalid.desktop.frag b/reference/opt/shaders/desktop-only/frag/fp16.invalid.desktop.frag index e1fc045c..faf79b2b 100644 --- a/reference/opt/shaders/desktop-only/frag/fp16.invalid.desktop.frag +++ b/reference/opt/shaders/desktop-only/frag/fp16.invalid.desktop.frag @@ -30,19 +30,19 @@ f16mat3 test_mat3(f16vec3 a, f16vec3 b, f16vec3 c, f16vec3 d, f16vec3 e, f16vec3 void test_constants() { - float16_t a = 1.0hf; - float16_t b = 1.5hf; - float16_t c = -1.5hf; - float16_t d = (0.0hf / 0.0hf); - float16_t e = (1.0hf / 0.0hf); - float16_t f = (-1.0hf / 0.0hf); - float16_t g = 1014.0hf; - float16_t h = 9.5367431640625e-07hf; + float16_t a = float16_t(1.0); + float16_t b = float16_t(1.5); + float16_t c = float16_t(-1.5); + float16_t d = float16_t(0.0 / 0.0); + float16_t e = float16_t(1.0 / 0.0); + float16_t f = float16_t(-1.0 / 0.0); + float16_t g = float16_t(1014.0); + float16_t h = float16_t(9.5367431640625e-07); } float16_t test_result() { - return 1.0hf; + return float16_t(1.0); } void test_conversions() @@ -50,7 +50,7 @@ void test_conversions() float16_t one = test_result(); int a = int(one); uint b = uint(one); - bool c = one != 0.0hf; + bool c = one != float16_t(0.0); float d = float(one); double e = double(one); float16_t a2 = float16_t(a); diff --git a/reference/opt/shaders/frag/16bit-constants.frag b/reference/opt/shaders/frag/16bit-constants.frag index f9ce356f..57d82561 100644 --- a/reference/opt/shaders/frag/16bit-constants.frag +++ b/reference/opt/shaders/frag/16bit-constants.frag @@ -18,7 +18,7 @@ layout(location = 2) out uint16_t baz; void main() { - foo = 1.0hf; + foo = float16_t(1.0); bar = 2s; baz = 3us; } diff --git a/reference/shaders-msl/comp/bitcast-16bit-1.invalid.comp b/reference/shaders-msl/comp/bitcast-16bit-1.invalid.comp index 8315f792..170e4920 100644 --- a/reference/shaders-msl/comp/bitcast-16bit-1.invalid.comp +++ b/reference/shaders-msl/comp/bitcast-16bit-1.invalid.comp @@ -17,7 +17,7 @@ kernel void main0(device SSBO0& _25 [[buffer(0)]], device SSBO1& _39 [[buffer(1) { uint ident = gl_GlobalInvocationID.x; half2 a = as_type<half2>(_25.inputs[ident].xy); - _39.outputs[ident].x = int(as_type<uint>(a + half2(1.0h))); + _39.outputs[ident].x = int(as_type<uint>(a + half2(half(1.0)))); _39.outputs[ident].y = as_type<int>(_25.inputs[ident].zw); _39.outputs[ident].z = int(as_type<uint>(ushort2(_25.inputs[ident].xy))); } diff --git a/reference/shaders-msl/frag/16bit-constants.frag b/reference/shaders-msl/frag/16bit-constants.frag index 2aae1e8a..56c7ea5d 100644 --- a/reference/shaders-msl/frag/16bit-constants.frag +++ b/reference/shaders-msl/frag/16bit-constants.frag @@ -13,7 +13,7 @@ struct main0_out fragment main0_out main0() { main0_out out = {}; - out.foo = 1.0h; + out.foo = half(1.0); out.bar = 2; out.baz = 3u; return out; diff --git a/reference/shaders-msl/frag/fp16.desktop.invalid.frag b/reference/shaders-msl/frag/fp16.desktop.invalid.frag index b914e4b8..001944fc 100644 --- a/reference/shaders-msl/frag/fp16.desktop.invalid.frag +++ b/reference/shaders-msl/frag/fp16.desktop.invalid.frag @@ -52,19 +52,19 @@ half3x3 test_mat3(thread const half3& a, thread const half3& b, thread const hal void test_constants() { - half a = 1.0h; - half b = 1.5h; - half c = -1.5h; - half d = (0.0h / 0.0h); - half e = (1.0h / 0.0h); - half f = (-1.0h / 0.0h); - half g = 1014.0h; - half h = 9.5367431640625e-07h; + half a = half(1.0); + half b = half(1.5); + half c = half(-1.5); + half d = half(0.0 / 0.0); + half e = half(1.0 / 0.0); + half f = half(-1.0 / 0.0); + half g = half(1014.0); + half h = half(9.5367431640625e-07); } half test_result() { - return 1.0h; + return half(1.0); } void test_conversions() @@ -72,7 +72,7 @@ void test_conversions() half one = test_result(); int a = int(one); uint b = uint(one); - bool c = one != 0.0h; + bool c = one != half(0.0); float d = float(one); half a2 = half(a); half b2 = half(b); diff --git a/reference/shaders/comp/bitcast-16bit-1.invalid.comp b/reference/shaders/comp/bitcast-16bit-1.invalid.comp index a4014e89..501f9795 100644 --- a/reference/shaders/comp/bitcast-16bit-1.invalid.comp +++ b/reference/shaders/comp/bitcast-16bit-1.invalid.comp @@ -27,7 +27,7 @@ void main() { uint ident = gl_GlobalInvocationID.x; f16vec2 a = int16BitsToFloat16(_25.inputs[ident].xy); - _39.outputs[ident].x = int(packFloat2x16(a + f16vec2(1.0hf))); + _39.outputs[ident].x = int(packFloat2x16(a + f16vec2(float16_t(1.0)))); _39.outputs[ident].y = packInt2x16(_25.inputs[ident].zw); _39.outputs[ident].z = int(packUint2x16(u16vec2(_25.inputs[ident].xy))); } diff --git a/reference/shaders/desktop-only/frag/fp16.invalid.desktop.frag b/reference/shaders/desktop-only/frag/fp16.invalid.desktop.frag index e1fc045c..faf79b2b 100644 --- a/reference/shaders/desktop-only/frag/fp16.invalid.desktop.frag +++ b/reference/shaders/desktop-only/frag/fp16.invalid.desktop.frag @@ -30,19 +30,19 @@ f16mat3 test_mat3(f16vec3 a, f16vec3 b, f16vec3 c, f16vec3 d, f16vec3 e, f16vec3 void test_constants() { - float16_t a = 1.0hf; - float16_t b = 1.5hf; - float16_t c = -1.5hf; - float16_t d = (0.0hf / 0.0hf); - float16_t e = (1.0hf / 0.0hf); - float16_t f = (-1.0hf / 0.0hf); - float16_t g = 1014.0hf; - float16_t h = 9.5367431640625e-07hf; + float16_t a = float16_t(1.0); + float16_t b = float16_t(1.5); + float16_t c = float16_t(-1.5); + float16_t d = float16_t(0.0 / 0.0); + float16_t e = float16_t(1.0 / 0.0); + float16_t f = float16_t(-1.0 / 0.0); + float16_t g = float16_t(1014.0); + float16_t h = float16_t(9.5367431640625e-07); } float16_t test_result() { - return 1.0hf; + return float16_t(1.0); } void test_conversions() @@ -50,7 +50,7 @@ void test_conversions() float16_t one = test_result(); int a = int(one); uint b = uint(one); - bool c = one != 0.0hf; + bool c = one != float16_t(0.0); float d = float(one); double e = double(one); float16_t a2 = float16_t(a); diff --git a/reference/shaders/frag/16bit-constants.frag b/reference/shaders/frag/16bit-constants.frag index f9ce356f..57d82561 100644 --- a/reference/shaders/frag/16bit-constants.frag +++ b/reference/shaders/frag/16bit-constants.frag @@ -18,7 +18,7 @@ layout(location = 2) out uint16_t baz; void main() { - foo = 1.0hf; + foo = float16_t(1.0); bar = 2s; baz = 3us; } diff --git a/spirv_glsl.cpp b/spirv_glsl.cpp index 74724e79..affaf697 100644 --- a/spirv_glsl.cpp +++ b/spirv_glsl.cpp @@ -2961,51 +2961,31 @@ string CompilerGLSL::convert_half_to_string(const SPIRConstant &c, uint32_t col, string res; float float_value = c.scalar_f16(col, row); + // There is no literal "hf" in GL_NV_gpu_shader5, so to avoid lots + // of complicated workarounds, just value-cast to the half type always. if (std::isnan(float_value) || std::isinf(float_value)) { - if (backend.half_literal_suffix) - { - // There is no uintBitsToFloat for 16-bit, so have to rely on legacy fallback here. - if (float_value == numeric_limits<float>::infinity()) - res = join("(1.0", backend.half_literal_suffix, " / 0.0", backend.half_literal_suffix, ")"); - else if (float_value == -numeric_limits<float>::infinity()) - res = join("(-1.0", backend.half_literal_suffix, " / 0.0", backend.half_literal_suffix, ")"); - else if (std::isnan(float_value)) - res = join("(0.0", backend.half_literal_suffix, " / 0.0", backend.half_literal_suffix, ")"); - else - SPIRV_CROSS_THROW("Cannot represent non-finite floating point constant."); - } + SPIRType type; + type.basetype = SPIRType::Half; + type.vecsize = 1; + type.columns = 1; + + if (float_value == numeric_limits<float>::infinity()) + res = join(type_to_glsl(type), "(1.0 / 0.0)"); + else if (float_value == -numeric_limits<float>::infinity()) + res = join(type_to_glsl(type), "(-1.0 / 0.0)"); + else if (std::isnan(float_value)) + res = join(type_to_glsl(type), "(0.0 / 0.0)"); else - { - SPIRType type; - type.basetype = SPIRType::Half; - type.vecsize = 1; - type.columns = 1; - - if (float_value == numeric_limits<float>::infinity()) - res = join(type_to_glsl(type), "(1.0 / 0.0)"); - else if (float_value == -numeric_limits<float>::infinity()) - res = join(type_to_glsl(type), "(-1.0 / 0.0)"); - else if (std::isnan(float_value)) - res = join(type_to_glsl(type), "(0.0 / 0.0)"); - else - SPIRV_CROSS_THROW("Cannot represent non-finite floating point constant."); - } + SPIRV_CROSS_THROW("Cannot represent non-finite floating point constant."); } else { - if (backend.half_literal_suffix) - res = convert_to_string(float_value) + backend.half_literal_suffix; - else - { - // In HLSL (FXC), it's important to cast the literals to half precision right away. - // There is no literal for it. - SPIRType type; - type.basetype = SPIRType::Half; - type.vecsize = 1; - type.columns = 1; - res = join(type_to_glsl(type), "(", convert_to_string(float_value), ")"); - } + SPIRType type; + type.basetype = SPIRType::Half; + type.vecsize = 1; + type.columns = 1; + res = join(type_to_glsl(type), "(", convert_to_string(float_value), ")"); } return res; diff --git a/spirv_glsl.hpp b/spirv_glsl.hpp index f94043b9..7477cf3b 100644 --- a/spirv_glsl.hpp +++ b/spirv_glsl.hpp @@ -380,7 +380,6 @@ protected: const char *basic_uint8_type = "uint8_t"; const char *basic_int16_type = "int16_t"; const char *basic_uint16_type = "uint16_t"; - const char *half_literal_suffix = "hf"; const char *int16_t_literal_suffix = "s"; const char *uint16_t_literal_suffix = "us"; bool swizzle_is_function = false; diff --git a/spirv_hlsl.cpp b/spirv_hlsl.cpp index c3191577..1b9a8270 100644 --- a/spirv_hlsl.cpp +++ b/spirv_hlsl.cpp @@ -4618,7 +4618,6 @@ string CompilerHLSL::compile() options.vulkan_semantics = true; backend.float_literal_suffix = true; backend.double_literal_suffix = false; - backend.half_literal_suffix = nullptr; backend.long_long_literal_suffix = true; backend.uint32_t_literal_suffix = true; backend.int16_t_literal_suffix = nullptr; diff --git a/spirv_msl.cpp b/spirv_msl.cpp index 1d369469..2682ae0a 100644 --- a/spirv_msl.cpp +++ b/spirv_msl.cpp @@ -564,7 +564,6 @@ string CompilerMSL::compile() options.version = 450; backend.null_pointer_literal = "nullptr"; backend.float_literal_suffix = false; - backend.half_literal_suffix = "h"; backend.uint32_t_literal_suffix = true; backend.int16_t_literal_suffix = nullptr; backend.uint16_t_literal_suffix = "u"; |