From b7e560dea0551e8aa0f041ef7842559e324c0467 Mon Sep 17 00:00:00 2001 From: Stephen White Date: Tue, 5 Apr 2022 18:51:15 +0000 Subject: [PATCH] GLSL: fix "uniform" qualifier on texture parameters. Textures as function parameters should not have the "uniform" qualifier. Fixed by handling StorageClass::kUniformConstant the same as StorageClass::kUniform, and removing the unconditional "uniform" qualifier output. (Global texture variables have StorageClass::kUniformConstant set, while function parameters don't.) Change-Id: I9d380550ac4554917527ff330171a76a90a290e8 Bug: tint:1492 Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/85820 Reviewed-by: Antonio Maiorano Commit-Queue: Stephen White Kokoro: Kokoro --- src/tint/writer/glsl/generator_impl.cc | 5 +- .../writer/glsl/generator_impl_type_test.cc | 2 +- .../builtins/textureLoad/texture_param.wgsl | 25 ++++++ .../texture_param.wgsl.expected.glsl | 63 +++++++++++++ .../texture_param.wgsl.expected.hlsl | 36 ++++++++ .../texture_param.wgsl.expected.msl | 37 ++++++++ .../texture_param.wgsl.expected.spvasm | 90 +++++++++++++++++++ .../texture_param.wgsl.expected.wgsl | 25 ++++++ 8 files changed, 280 insertions(+), 3 deletions(-) create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl.expected.glsl create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl.expected.hlsl create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl.expected.msl create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl.expected.spvasm create mode 100644 test/tint/builtins/textureLoad/texture_param.wgsl.expected.wgsl diff --git a/src/tint/writer/glsl/generator_impl.cc b/src/tint/writer/glsl/generator_impl.cc index c4722f6beb..05fd7613b3 100644 --- a/src/tint/writer/glsl/generator_impl.cc +++ b/src/tint/writer/glsl/generator_impl.cc @@ -2488,7 +2488,8 @@ bool GeneratorImpl::EmitType(std::ostream& out, out << "out "; break; } - case ast::StorageClass::kUniform: { + case ast::StorageClass::kUniform: + case ast::StorageClass::kUniformConstant: { out << "uniform "; break; } @@ -2552,7 +2553,7 @@ bool GeneratorImpl::EmitType(std::ostream& out, auto* depth_ms = tex->As(); auto* sampled = tex->As(); - out << "uniform highp "; + out << "highp "; if (storage && storage->access() != ast::Access::kRead) { out << "writeonly "; diff --git a/src/tint/writer/glsl/generator_impl_type_test.cc b/src/tint/writer/glsl/generator_impl_type_test.cc index 1d29857fec..c6d7e86e74 100644 --- a/src/tint/writer/glsl/generator_impl_type_test.cc +++ b/src/tint/writer/glsl/generator_impl_type_test.cc @@ -472,7 +472,7 @@ TEST_F(GlslGeneratorImplTest_Type, EmitMultisampledTexture) { ASSERT_TRUE(gen.EmitType(out, s, ast::StorageClass::kNone, ast::Access::kReadWrite, "")) << gen.error(); - EXPECT_EQ(out.str(), "uniform highp sampler2DMS"); + EXPECT_EQ(out.str(), "highp sampler2DMS"); } struct GlslStorageTextureData { diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl b/test/tint/builtins/textureLoad/texture_param.wgsl new file mode 100644 index 0000000000..81f359ea41 --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl @@ -0,0 +1,25 @@ +@group(1) @binding(0) var arg_0: texture_2d; + +fn textureLoad2d(texture: texture_2d, coords: vec2, level: i32) -> vec4 { + return textureLoad(texture, coords, level); +} + +fn doTextureLoad() { + var res: vec4 = textureLoad2d(arg_0, vec2(), 0); +} + +@stage(vertex) +fn vertex_main() -> @builtin(position) vec4 { + doTextureLoad(); + return vec4(); +} + +@stage(fragment) +fn fragment_main() { + doTextureLoad(); +} + +@stage(compute) @workgroup_size(1) +fn compute_main() { + doTextureLoad(); +} diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl.expected.glsl b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.glsl new file mode 100644 index 0000000000..862fcee33f --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.glsl @@ -0,0 +1,63 @@ +#version 310 es + +ivec4 textureLoad2d(highp isampler2D tint_symbol_1, ivec2 coords, int level) { + return texelFetch(tint_symbol_1, coords, level); +} + +uniform highp isampler2D arg_0_1; +void doTextureLoad() { + ivec4 res = textureLoad2d(arg_0_1, ivec2(0, 0), 0); +} + +vec4 vertex_main() { + doTextureLoad(); + return vec4(0.0f, 0.0f, 0.0f, 0.0f); +} + +void main() { + vec4 inner_result = vertex_main(); + gl_Position = inner_result; + gl_Position.y = -(gl_Position.y); + gl_Position.z = ((2.0f * gl_Position.z) - gl_Position.w); + return; +} +#version 310 es +precision mediump float; + +ivec4 textureLoad2d(highp isampler2D tint_symbol_1, ivec2 coords, int level) { + return texelFetch(tint_symbol_1, coords, level); +} + +uniform highp isampler2D arg_0_1; +void doTextureLoad() { + ivec4 res = textureLoad2d(arg_0_1, ivec2(0, 0), 0); +} + +void fragment_main() { + doTextureLoad(); +} + +void main() { + fragment_main(); + return; +} +#version 310 es + +ivec4 textureLoad2d(highp isampler2D tint_symbol_1, ivec2 coords, int level) { + return texelFetch(tint_symbol_1, coords, level); +} + +uniform highp isampler2D arg_0_1; +void doTextureLoad() { + ivec4 res = textureLoad2d(arg_0_1, ivec2(0, 0), 0); +} + +void compute_main() { + doTextureLoad(); +} + +layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in; +void main() { + compute_main(); + return; +} diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl.expected.hlsl b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.hlsl new file mode 100644 index 0000000000..1544f5ec7c --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.hlsl @@ -0,0 +1,36 @@ +Texture2D arg_0 : register(t0, space1); + +int4 textureLoad2d(Texture2D tint_symbol, int2 coords, int level) { + return tint_symbol.Load(int3(coords, level)); +} + +void doTextureLoad() { + int4 res = textureLoad2d(arg_0, int2(0, 0), 0); +} + +struct tint_symbol_1 { + float4 value : SV_Position; +}; + +float4 vertex_main_inner() { + doTextureLoad(); + return float4(0.0f, 0.0f, 0.0f, 0.0f); +} + +tint_symbol_1 vertex_main() { + const float4 inner_result = vertex_main_inner(); + tint_symbol_1 wrapper_result = (tint_symbol_1)0; + wrapper_result.value = inner_result; + return wrapper_result; +} + +void fragment_main() { + doTextureLoad(); + return; +} + +[numthreads(1, 1, 1)] +void compute_main() { + doTextureLoad(); + return; +} diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl.expected.msl b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.msl new file mode 100644 index 0000000000..cee9f8ef51 --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.msl @@ -0,0 +1,37 @@ +#include + +using namespace metal; +int4 textureLoad2d(texture2d tint_symbol, int2 coords, int level) { + return tint_symbol.read(uint2(coords), level); +} + +void doTextureLoad(texture2d tint_symbol_2) { + int4 res = textureLoad2d(tint_symbol_2, int2(), 0); +} + +struct tint_symbol_1 { + float4 value [[position]]; +}; + +float4 vertex_main_inner(texture2d tint_symbol_3) { + doTextureLoad(tint_symbol_3); + return float4(); +} + +vertex tint_symbol_1 vertex_main(texture2d tint_symbol_4 [[texture(0)]]) { + float4 const inner_result = vertex_main_inner(tint_symbol_4); + tint_symbol_1 wrapper_result = {}; + wrapper_result.value = inner_result; + return wrapper_result; +} + +fragment void fragment_main(texture2d tint_symbol_5 [[texture(0)]]) { + doTextureLoad(tint_symbol_5); + return; +} + +kernel void compute_main(texture2d tint_symbol_6 [[texture(0)]]) { + doTextureLoad(tint_symbol_6); + return; +} + diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl.expected.spvasm b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.spvasm new file mode 100644 index 0000000000..b292e121d8 --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.spvasm @@ -0,0 +1,90 @@ +; SPIR-V +; Version: 1.3 +; Generator: Google Tint Compiler; 0 +; Bound: 47 +; Schema: 0 + OpCapability Shader + OpMemoryModel Logical GLSL450 + OpEntryPoint Vertex %vertex_main "vertex_main" %value %vertex_point_size + OpEntryPoint Fragment %fragment_main "fragment_main" + OpEntryPoint GLCompute %compute_main "compute_main" + OpExecutionMode %fragment_main OriginUpperLeft + OpExecutionMode %compute_main LocalSize 1 1 1 + OpName %value "value" + OpName %vertex_point_size "vertex_point_size" + OpName %arg_0 "arg_0" + OpName %textureLoad2d "textureLoad2d" + OpName %texture "texture" + OpName %coords "coords" + OpName %level "level" + OpName %doTextureLoad "doTextureLoad" + OpName %res "res" + OpName %vertex_main_inner "vertex_main_inner" + OpName %vertex_main "vertex_main" + OpName %fragment_main "fragment_main" + OpName %compute_main "compute_main" + OpDecorate %value BuiltIn Position + OpDecorate %vertex_point_size BuiltIn PointSize + OpDecorate %arg_0 DescriptorSet 1 + OpDecorate %arg_0 Binding 0 + %float = OpTypeFloat 32 + %v4float = OpTypeVector %float 4 +%_ptr_Output_v4float = OpTypePointer Output %v4float + %5 = OpConstantNull %v4float + %value = OpVariable %_ptr_Output_v4float Output %5 +%_ptr_Output_float = OpTypePointer Output %float + %8 = OpConstantNull %float +%vertex_point_size = OpVariable %_ptr_Output_float Output %8 + %int = OpTypeInt 32 1 + %11 = OpTypeImage %int 2D 0 0 0 1 Unknown +%_ptr_UniformConstant_11 = OpTypePointer UniformConstant %11 + %arg_0 = OpVariable %_ptr_UniformConstant_11 UniformConstant + %v4int = OpTypeVector %int 4 + %v2int = OpTypeVector %int 2 + %13 = OpTypeFunction %v4int %11 %v2int %int + %void = OpTypeVoid + %22 = OpTypeFunction %void + %28 = OpConstantNull %v2int + %int_0 = OpConstant %int 0 +%_ptr_Function_v4int = OpTypePointer Function %v4int + %32 = OpConstantNull %v4int + %33 = OpTypeFunction %v4float + %float_1 = OpConstant %float 1 +%textureLoad2d = OpFunction %v4int None %13 + %texture = OpFunctionParameter %11 + %coords = OpFunctionParameter %v2int + %level = OpFunctionParameter %int + %20 = OpLabel + %21 = OpImageFetch %v4int %texture %coords Lod %level + OpReturnValue %21 + OpFunctionEnd +%doTextureLoad = OpFunction %void None %22 + %25 = OpLabel + %res = OpVariable %_ptr_Function_v4int Function %32 + %27 = OpLoad %11 %arg_0 + %26 = OpFunctionCall %v4int %textureLoad2d %27 %28 %int_0 + OpStore %res %26 + OpReturn + OpFunctionEnd +%vertex_main_inner = OpFunction %v4float None %33 + %35 = OpLabel + %36 = OpFunctionCall %void %doTextureLoad + OpReturnValue %5 + OpFunctionEnd +%vertex_main = OpFunction %void None %22 + %38 = OpLabel + %39 = OpFunctionCall %v4float %vertex_main_inner + OpStore %value %39 + OpStore %vertex_point_size %float_1 + OpReturn + OpFunctionEnd +%fragment_main = OpFunction %void None %22 + %42 = OpLabel + %43 = OpFunctionCall %void %doTextureLoad + OpReturn + OpFunctionEnd +%compute_main = OpFunction %void None %22 + %45 = OpLabel + %46 = OpFunctionCall %void %doTextureLoad + OpReturn + OpFunctionEnd diff --git a/test/tint/builtins/textureLoad/texture_param.wgsl.expected.wgsl b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.wgsl new file mode 100644 index 0000000000..13f6d82c09 --- /dev/null +++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.wgsl @@ -0,0 +1,25 @@ +@group(1) @binding(0) var arg_0 : texture_2d; + +fn textureLoad2d(texture : texture_2d, coords : vec2, level : i32) -> vec4 { + return textureLoad(texture, coords, level); +} + +fn doTextureLoad() { + var res : vec4 = textureLoad2d(arg_0, vec2(), 0); +} + +@stage(vertex) +fn vertex_main() -> @builtin(position) vec4 { + doTextureLoad(); + return vec4(); +} + +@stage(fragment) +fn fragment_main() { + doTextureLoad(); +} + +@stage(compute) @workgroup_size(1) +fn compute_main() { + doTextureLoad(); +}