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 <amaiorano@google.com>
Commit-Queue: Stephen White <senorblanco@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
diff --git a/src/tint/writer/glsl/generator_impl.cc b/src/tint/writer/glsl/generator_impl.cc
index c4722f6..05fd761 100644
--- a/src/tint/writer/glsl/generator_impl.cc
+++ b/src/tint/writer/glsl/generator_impl.cc
@@ -2488,7 +2488,8 @@
out << "out ";
break;
}
- case ast::StorageClass::kUniform: {
+ case ast::StorageClass::kUniform:
+ case ast::StorageClass::kUniformConstant: {
out << "uniform ";
break;
}
@@ -2552,7 +2553,7 @@
auto* depth_ms = tex->As<sem::DepthMultisampledTexture>();
auto* sampled = tex->As<sem::SampledTexture>();
- 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 1d29857..c6d7e86 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 @@
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 0000000..81f359e
--- /dev/null
+++ b/test/tint/builtins/textureLoad/texture_param.wgsl
@@ -0,0 +1,25 @@
+@group(1) @binding(0) var arg_0: texture_2d<i32>;
+
+fn textureLoad2d(texture: texture_2d<i32>, coords: vec2<i32>, level: i32) -> vec4<i32> {
+ return textureLoad(texture, coords, level);
+}
+
+fn doTextureLoad() {
+ var res: vec4<i32> = textureLoad2d(arg_0, vec2<i32>(), 0);
+}
+
+@stage(vertex)
+fn vertex_main() -> @builtin(position) vec4<f32> {
+ doTextureLoad();
+ return vec4<f32>();
+}
+
+@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 0000000..862fcee
--- /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 0000000..1544f5e
--- /dev/null
+++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.hlsl
@@ -0,0 +1,36 @@
+Texture2D<int4> arg_0 : register(t0, space1);
+
+int4 textureLoad2d(Texture2D<int4> 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 0000000..cee9f8e
--- /dev/null
+++ b/test/tint/builtins/textureLoad/texture_param.wgsl.expected.msl
@@ -0,0 +1,37 @@
+#include <metal_stdlib>
+
+using namespace metal;
+int4 textureLoad2d(texture2d<int, access::sample> tint_symbol, int2 coords, int level) {
+ return tint_symbol.read(uint2(coords), level);
+}
+
+void doTextureLoad(texture2d<int, access::sample> tint_symbol_2) {
+ int4 res = textureLoad2d(tint_symbol_2, int2(), 0);
+}
+
+struct tint_symbol_1 {
+ float4 value [[position]];
+};
+
+float4 vertex_main_inner(texture2d<int, access::sample> tint_symbol_3) {
+ doTextureLoad(tint_symbol_3);
+ return float4();
+}
+
+vertex tint_symbol_1 vertex_main(texture2d<int, access::sample> 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<int, access::sample> tint_symbol_5 [[texture(0)]]) {
+ doTextureLoad(tint_symbol_5);
+ return;
+}
+
+kernel void compute_main(texture2d<int, access::sample> 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 0000000..b292e12
--- /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 0000000..13f6d82
--- /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<i32>;
+
+fn textureLoad2d(texture : texture_2d<i32>, coords : vec2<i32>, level : i32) -> vec4<i32> {
+ return textureLoad(texture, coords, level);
+}
+
+fn doTextureLoad() {
+ var res : vec4<i32> = textureLoad2d(arg_0, vec2<i32>(), 0);
+}
+
+@stage(vertex)
+fn vertex_main() -> @builtin(position) vec4<f32> {
+ doTextureLoad();
+ return vec4<f32>();
+}
+
+@stage(fragment)
+fn fragment_main() {
+ doTextureLoad();
+}
+
+@stage(compute) @workgroup_size(1)
+fn compute_main() {
+ doTextureLoad();
+}
