More GLSL fixes.
Remove register-and-space decoration.
Add "main" to GLSL keywords, to force renaming.
Output correct compute width decoration (not numthreads).
Remove static keyword.
Bug: tint:1218 tint:1219 tint:1220
Change-Id: I171f183690b6531c76218414e0d81f6ef5e22e6b
Reviewed-on: https://dawn-review.googlesource.com/c/tint/+/66340
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Stephen White <senorblanco@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
diff --git a/src/transform/renamer.cc b/src/transform/renamer.cc
index 6bf953f..9a89cf8 100644
--- a/src/transform/renamer.cc
+++ b/src/transform/renamer.cc
@@ -136,6 +136,7 @@
"layout",
"long",
"lowp",
+ "main",
"mat2",
"mat2x2",
"mat2x3",
diff --git a/src/writer/glsl/generator_impl.cc b/src/writer/glsl/generator_impl.cc
index 4bf06b2..55b5098 100644
--- a/src/writer/glsl/generator_impl.cc
+++ b/src/writer/glsl/generator_impl.cc
@@ -93,22 +93,6 @@
}
}
-// Helper for writing " : register(RX, spaceY)", where R is the register, X is
-// the binding point binding value, and Y is the binding point group value.
-struct RegisterAndSpace {
- RegisterAndSpace(char r, ast::Variable::BindingPoint bp)
- : reg(r), binding_point(bp) {}
-
- char const reg;
- ast::Variable::BindingPoint const binding_point;
-};
-
-std::ostream& operator<<(std::ostream& s, const RegisterAndSpace& rs) {
- s << " : register(" << rs.reg << rs.binding_point.binding->value()
- << ", space" << rs.binding_point.group->value() << ")";
- return s;
-}
-
} // namespace
GeneratorImpl::GeneratorImpl(const Program* program) : TextGenerator(program) {}
@@ -1591,16 +1575,13 @@
return false;
}
- out << RegisterAndSpace(var->Access() == ast::Access::kRead ? 't' : 'u',
- decl->binding_point())
- << ";";
+ out << ";";
return true;
}
bool GeneratorImpl::EmitHandleVariable(const sem::Variable* var) {
auto* decl = var->Declaration();
- auto* unwrapped_type = var->Type()->UnwrapRef();
auto out = line();
auto name = builder_.Symbols().NameFor(decl->symbol());
@@ -1609,25 +1590,6 @@
return false;
}
- const char* register_space = nullptr;
-
- if (unwrapped_type->Is<sem::Texture>()) {
- register_space = "t";
- if (auto* storage_tex = unwrapped_type->As<sem::StorageTexture>()) {
- if (storage_tex->access() != ast::Access::kRead) {
- register_space = "u";
- }
- }
- } else if (unwrapped_type->Is<sem::Sampler>()) {
- register_space = "s";
- }
-
- if (register_space) {
- auto bp = decl->binding_point();
- out << " : register(" << register_space << bp.binding->value() << ", space"
- << bp.group->value() << ")";
- }
-
out << ";";
return true;
}
@@ -1636,8 +1598,6 @@
auto* decl = var->Declaration();
auto out = line();
- out << "static ";
-
auto name = builder_.Symbols().NameFor(decl->symbol());
auto* type = var->Type()->UnwrapRef();
if (!EmitTypeAndName(out, type, var->StorageClass(), var->Access(), name)) {
@@ -1778,13 +1738,14 @@
{
auto out = line();
if (func->pipeline_stage() == ast::PipelineStage::kCompute) {
- // Emit the workgroup_size attribute.
+ // Emit the layout(local_size) attributes.
auto wgsize = func_sem->workgroup_size();
- out << "[numthreads(";
+ out << "layout(";
for (int i = 0; i < 3; i++) {
if (i > 0) {
out << ", ";
}
+ out << "local_size_" << (i == 0 ? "x" : i == 1 ? "y" : "z") << " = ";
if (wgsize[i].overridable_const) {
auto* global = builder_.Sem().Get<sem::GlobalVariable>(
@@ -1798,7 +1759,7 @@
out << std::to_string(wgsize[i].value);
}
}
- out << ")]" << std::endl;
+ out << ") in;" << std::endl;
}
out << func->return_type()->FriendlyName(builder_.Symbols());
@@ -2641,7 +2602,7 @@
line() << "#endif";
{
auto out = line();
- out << "static const ";
+ out << "const ";
if (!EmitTypeAndName(out, type, sem->StorageClass(), sem->Access(),
builder_.Symbols().NameFor(var->symbol()))) {
return false;
@@ -2650,7 +2611,7 @@
}
} else {
auto out = line();
- out << "static const ";
+ out << "const ";
if (!EmitTypeAndName(out, type, sem->StorageClass(), sem->Access(),
builder_.Symbols().NameFor(var->symbol()))) {
return false;
diff --git a/src/writer/glsl/generator_impl_function_test.cc b/src/writer/glsl/generator_impl_function_test.cc
index e0ed9d9..51c3cd6 100644
--- a/src/writer/glsl/generator_impl_function_test.cc
+++ b/src/writer/glsl/generator_impl_function_test.cc
@@ -537,7 +537,7 @@
precision mediump float;
-Data coord : register(u0, space1);
+Data coord;
void frag_main() {
float v = coord.b;
@@ -586,7 +586,7 @@
precision mediump float;
-Data coord : register(t0, space1);
+Data coord;
void frag_main() {
float v = coord.b;
@@ -631,7 +631,7 @@
precision mediump float;
-Data coord : register(u0, space1);
+Data coord;
void frag_main() {
coord.b = 2.0f;
@@ -677,7 +677,7 @@
precision mediump float;
-Data coord : register(u0, space1);
+Data coord;
void frag_main() {
coord.b = 2.0f;
@@ -782,7 +782,7 @@
precision mediump float;
-S coord : register(u0, space1);
+S coord;
float sub_func(float param) {
return coord.x;
@@ -837,7 +837,7 @@
EXPECT_EQ(gen.result(), R"(#version 310 es
precision mediump float;
-[numthreads(1, 1, 1)]
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
return;
}
@@ -863,7 +863,7 @@
EXPECT_EQ(gen.result(), R"(#version 310 es
precision mediump float;
-[numthreads(2, 4, 6)]
+layout(local_size_x = 2, local_size_y = 4, local_size_z = 6) in;
void main() {
return;
}
@@ -892,11 +892,11 @@
EXPECT_EQ(gen.result(), R"(#version 310 es
precision mediump float;
-static const int width = int(2);
-static const int height = int(3);
-static const int depth = int(4);
+const int width = int(2);
+const int height = int(3);
+const int depth = int(4);
-[numthreads(2, 3, 4)]
+layout(local_size_x = 2, local_size_y = 3, local_size_z = 4) in;
void main() {
return;
}
@@ -928,17 +928,17 @@
#ifndef WGSL_SPEC_CONSTANT_7
#define WGSL_SPEC_CONSTANT_7 int(2)
#endif
-static const int width = WGSL_SPEC_CONSTANT_7;
+const int width = WGSL_SPEC_CONSTANT_7;
#ifndef WGSL_SPEC_CONSTANT_8
#define WGSL_SPEC_CONSTANT_8 int(3)
#endif
-static const int height = WGSL_SPEC_CONSTANT_8;
+const int height = WGSL_SPEC_CONSTANT_8;
#ifndef WGSL_SPEC_CONSTANT_9
#define WGSL_SPEC_CONSTANT_9 int(4)
#endif
-static const int depth = WGSL_SPEC_CONSTANT_9;
+const int depth = WGSL_SPEC_CONSTANT_9;
-[numthreads(WGSL_SPEC_CONSTANT_7, WGSL_SPEC_CONSTANT_8, WGSL_SPEC_CONSTANT_9)]
+layout(local_size_x = WGSL_SPEC_CONSTANT_7, local_size_y = WGSL_SPEC_CONSTANT_8, local_size_z = WGSL_SPEC_CONSTANT_9) in;
void main() {
return;
}
@@ -1047,9 +1047,9 @@
precision mediump float;
-Data data : register(u0, space0);
+Data data;
-[numthreads(1, 1, 1)]
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void a() {
float v = data.d;
return;
@@ -1060,7 +1060,7 @@
-[numthreads(1, 1, 1)]
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void b() {
float v = data.d;
return;
diff --git a/src/writer/glsl/generator_impl_intrinsic_test.cc b/src/writer/glsl/generator_impl_intrinsic_test.cc
index 6d1c392..fc8183b 100644
--- a/src/writer/glsl/generator_impl_intrinsic_test.cc
+++ b/src/writer/glsl/generator_impl_intrinsic_test.cc
@@ -545,7 +545,7 @@
GeneratorImpl& gen = Build();
ASSERT_TRUE(gen.Generate()) << gen.error();
- EXPECT_EQ(gen.result(), R"([numthreads(1, 1, 1)]
+ EXPECT_EQ(gen.result(), R"(layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
DeviceMemoryBarrierWithGroupSync();
return;
@@ -564,7 +564,7 @@
GeneratorImpl& gen = Build();
ASSERT_TRUE(gen.Generate()) << gen.error();
- EXPECT_EQ(gen.result(), R"([numthreads(1, 1, 1)]
+ EXPECT_EQ(gen.result(), R"(layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
GroupMemoryBarrierWithGroupSync();
return;
@@ -590,7 +590,7 @@
return ((a + b) * c);
}
-[numthreads(1, 1, 1)]
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void main() {
f(1, 2, 3);
return;
diff --git a/src/writer/glsl/generator_impl_member_accessor_test.cc b/src/writer/glsl/generator_impl_member_accessor_test.cc
index 808c211..9d19353 100644
--- a/src/writer/glsl/generator_impl_member_accessor_test.cc
+++ b/src/writer/glsl/generator_impl_member_accessor_test.cc
@@ -139,9 +139,9 @@
float mem;
};
-static Data str = Data(0.0f);
+Data str = Data(0.0f);
-[numthreads(1, 1, 1)]
+layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;
void test_function() {
float expr = str.mem;
return;
@@ -297,14 +297,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
data.b = mat2x3(0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f);
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -340,14 +340,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
float x = data.a[2][1];
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -381,14 +381,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
int x = data.a[2];
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -423,14 +423,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
int x = data.a[((2 + 4) - 3)];
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -462,14 +462,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
data.a[2] = 2;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -512,14 +512,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
vec3 x = data.c[2].b;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -565,14 +565,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
vec2 x = data.c[2].b.xy;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -618,14 +618,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
float x = data.c[2].b.g;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -671,14 +671,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
float x = data.c[2].b[1];
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -720,14 +720,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
data.c[2].b = vec3(1.0f, 2.0f, 3.0f);
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -773,14 +773,14 @@
precision mediump float;
-Data data : register(u0, space1);
+Data data;
-void main() {
+void tint_symbol() {
data.c[2].b.y = 1.0f;
return;
}
void main() {
- main();
+ tint_symbol();
}
diff --git a/src/writer/glsl/generator_impl_module_constant_test.cc b/src/writer/glsl/generator_impl_module_constant_test.cc
index 85bcab9..dc6db0d 100644
--- a/src/writer/glsl/generator_impl_module_constant_test.cc
+++ b/src/writer/glsl/generator_impl_module_constant_test.cc
@@ -29,8 +29,7 @@
GeneratorImpl& gen = Build();
ASSERT_TRUE(gen.EmitProgramConstVariable(var)) << gen.error();
- EXPECT_EQ(gen.result(),
- "static const float pos[3] = float[3](1.0f, 2.0f, 3.0f);\n");
+ EXPECT_EQ(gen.result(), "const float pos[3] = float[3](1.0f, 2.0f, 3.0f);\n");
}
TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant) {
@@ -45,7 +44,7 @@
EXPECT_EQ(gen.result(), R"(#ifndef WGSL_SPEC_CONSTANT_23
#define WGSL_SPEC_CONSTANT_23 3.0f
#endif
-static const float pos = WGSL_SPEC_CONSTANT_23;
+const float pos = WGSL_SPEC_CONSTANT_23;
)");
}
@@ -61,7 +60,7 @@
EXPECT_EQ(gen.result(), R"(#ifndef WGSL_SPEC_CONSTANT_23
#error spec constant required for constant id 23
#endif
-static const float pos = WGSL_SPEC_CONSTANT_23;
+const float pos = WGSL_SPEC_CONSTANT_23;
)");
}
@@ -82,11 +81,11 @@
EXPECT_EQ(gen.result(), R"(#ifndef WGSL_SPEC_CONSTANT_0
#define WGSL_SPEC_CONSTANT_0 3.0f
#endif
-static const float a = WGSL_SPEC_CONSTANT_0;
+const float a = WGSL_SPEC_CONSTANT_0;
#ifndef WGSL_SPEC_CONSTANT_1
#define WGSL_SPEC_CONSTANT_1 2.0f
#endif
-static const float b = WGSL_SPEC_CONSTANT_1;
+const float b = WGSL_SPEC_CONSTANT_1;
)");
}
diff --git a/src/writer/glsl/generator_impl_sanitizer_test.cc b/src/writer/glsl/generator_impl_sanitizer_test.cc
index 5bf694a..d35ac15 100644
--- a/src/writer/glsl/generator_impl_sanitizer_test.cc
+++ b/src/writer/glsl/generator_impl_sanitizer_test.cc
@@ -52,7 +52,7 @@
precision mediump float;
-my_struct b : register(t1, space2);
+my_struct b;
void a_func() {
uint tint_symbol_1 = 0u;
@@ -101,7 +101,7 @@
precision mediump float;
-my_struct b : register(t1, space2);
+my_struct b;
void a_func() {
uint tint_symbol_1 = 0u;
@@ -152,7 +152,7 @@
precision mediump float;
-my_struct b : register(t1, space2);
+my_struct b;
void a_func() {
uint tint_symbol_1 = 0u;
@@ -192,13 +192,13 @@
auto* expect = R"(#version 310 es
precision mediump float;
-void main() {
- int tint_symbol[4] = int[4](1, 2, 3, 4);
- int pos = tint_symbol[3];
+void tint_symbol() {
+ int tint_symbol_1[4] = int[4](1, 2, 3, 4);
+ int pos = tint_symbol_1[3];
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -239,13 +239,13 @@
int c;
};
-void main() {
- S tint_symbol = S(1, vec3(2.0f, 3.0f, 4.0f), 4);
- vec3 pos = tint_symbol.b;
+void tint_symbol() {
+ S tint_symbol_1 = S(1, vec3(2.0f, 3.0f, 4.0f), 4);
+ vec3 pos = tint_symbol_1.b;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -280,13 +280,13 @@
auto* expect = R"(#version 310 es
precision mediump float;
-void main() {
+void tint_symbol() {
int v = 0;
int x = v;
return;
}
void main() {
- main();
+ tint_symbol();
}
@@ -331,13 +331,13 @@
auto* expect = R"(#version 310 es
precision mediump float;
-void main() {
+void tint_symbol() {
mat4 m = mat4(0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f);
float f = m[2][1];
return;
}
void main() {
- main();
+ tint_symbol();
}
diff --git a/src/writer/glsl/generator_impl_type_test.cc b/src/writer/glsl/generator_impl_type_test.cc
index 0709dae..7bdb286 100644
--- a/src/writer/glsl/generator_impl_type_test.cc
+++ b/src/writer/glsl/generator_impl_type_test.cc
@@ -352,14 +352,12 @@
GlslGeneratorImplTest_Type,
GlslDepthTexturesTest,
testing::Values(
- GlslDepthTextureData{ast::TextureDimension::k2d,
- "Texture2D tex : register(t1, space2);"},
+ GlslDepthTextureData{ast::TextureDimension::k2d, "Texture2D tex;"},
GlslDepthTextureData{ast::TextureDimension::k2dArray,
- "Texture2DArray tex : register(t1, space2);"},
- GlslDepthTextureData{ast::TextureDimension::kCube,
- "TextureCube tex : register(t1, space2);"},
+ "Texture2DArray tex;"},
+ GlslDepthTextureData{ast::TextureDimension::kCube, "TextureCube tex;"},
GlslDepthTextureData{ast::TextureDimension::kCubeArray,
- "TextureCubeArray tex : register(t1, space2);"}));
+ "TextureCubeArray tex;"}));
using GlslDepthMultisampledTexturesTest = TestHelper;
TEST_F(GlslDepthMultisampledTexturesTest, Emit) {
@@ -377,8 +375,7 @@
GeneratorImpl& gen = Build();
ASSERT_TRUE(gen.Generate()) << gen.error();
- EXPECT_THAT(gen.result(),
- HasSubstr("Texture2DMS<float4> tex : register(t1, space2);"));
+ EXPECT_THAT(gen.result(), HasSubstr("Texture2DMS<float4> tex;"));
}
enum class TextureDataType { F32, U32, I32 };
@@ -424,100 +421,99 @@
ASSERT_TRUE(gen.Generate()) << gen.error();
EXPECT_THAT(gen.result(), HasSubstr(params.result));
}
-INSTANTIATE_TEST_SUITE_P(
- GlslGeneratorImplTest_Type,
- GlslSampledTexturesTest,
- testing::Values(
- GlslSampledTextureData{
- ast::TextureDimension::k1d,
- TextureDataType::F32,
- "Texture1D<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2d,
- TextureDataType::F32,
- "Texture2D<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2dArray,
- TextureDataType::F32,
- "Texture2DArray<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k3d,
- TextureDataType::F32,
- "Texture3D<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCube,
- TextureDataType::F32,
- "TextureCube<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCubeArray,
- TextureDataType::F32,
- "TextureCubeArray<float4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k1d,
- TextureDataType::U32,
- "Texture1D<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2d,
- TextureDataType::U32,
- "Texture2D<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2dArray,
- TextureDataType::U32,
- "Texture2DArray<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k3d,
- TextureDataType::U32,
- "Texture3D<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCube,
- TextureDataType::U32,
- "TextureCube<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCubeArray,
- TextureDataType::U32,
- "TextureCubeArray<uint4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k1d,
- TextureDataType::I32,
- "Texture1D<int4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2d,
- TextureDataType::I32,
- "Texture2D<int4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k2dArray,
- TextureDataType::I32,
- "Texture2DArray<int4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::k3d,
- TextureDataType::I32,
- "Texture3D<int4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCube,
- TextureDataType::I32,
- "TextureCube<int4> tex : register(t1, space2);",
- },
- GlslSampledTextureData{
- ast::TextureDimension::kCubeArray,
- TextureDataType::I32,
- "TextureCubeArray<int4> tex : register(t1, space2);",
- }));
+INSTANTIATE_TEST_SUITE_P(GlslGeneratorImplTest_Type,
+ GlslSampledTexturesTest,
+ testing::Values(
+ GlslSampledTextureData{
+ ast::TextureDimension::k1d,
+ TextureDataType::F32,
+ "Texture1D<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2d,
+ TextureDataType::F32,
+ "Texture2D<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2dArray,
+ TextureDataType::F32,
+ "Texture2DArray<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k3d,
+ TextureDataType::F32,
+ "Texture3D<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCube,
+ TextureDataType::F32,
+ "TextureCube<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCubeArray,
+ TextureDataType::F32,
+ "TextureCubeArray<float4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k1d,
+ TextureDataType::U32,
+ "Texture1D<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2d,
+ TextureDataType::U32,
+ "Texture2D<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2dArray,
+ TextureDataType::U32,
+ "Texture2DArray<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k3d,
+ TextureDataType::U32,
+ "Texture3D<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCube,
+ TextureDataType::U32,
+ "TextureCube<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCubeArray,
+ TextureDataType::U32,
+ "TextureCubeArray<uint4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k1d,
+ TextureDataType::I32,
+ "Texture1D<int4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2d,
+ TextureDataType::I32,
+ "Texture2D<int4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k2dArray,
+ TextureDataType::I32,
+ "Texture2DArray<int4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::k3d,
+ TextureDataType::I32,
+ "Texture3D<int4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCube,
+ TextureDataType::I32,
+ "TextureCube<int4> tex;",
+ },
+ GlslSampledTextureData{
+ ast::TextureDimension::kCubeArray,
+ TextureDataType::I32,
+ "TextureCubeArray<int4> tex;",
+ }));
TEST_F(GlslGeneratorImplTest_Type, EmitMultisampledTexture) {
auto* f32 = create<sem::F32>();
@@ -564,46 +560,45 @@
INSTANTIATE_TEST_SUITE_P(
GlslGeneratorImplTest_Type,
GlslStorageTexturesTest,
- testing::Values(
- GlslStorageTextureData{
- ast::TextureDimension::k1d, ast::ImageFormat::kRgba8Unorm,
- "RWTexture1D<float4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k2d, ast::ImageFormat::kRgba16Float,
- "RWTexture2D<float4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k2dArray, ast::ImageFormat::kR32Float,
- "RWTexture2DArray<float4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k3d, ast::ImageFormat::kRg32Float,
- "RWTexture3D<float4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k1d, ast::ImageFormat::kRgba32Float,
- "RWTexture1D<float4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k2d, ast::ImageFormat::kRgba16Uint,
- "RWTexture2D<uint4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k2dArray, ast::ImageFormat::kR32Uint,
- "RWTexture2DArray<uint4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k3d, ast::ImageFormat::kRg32Uint,
- "RWTexture3D<uint4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k1d, ast::ImageFormat::kRgba32Uint,
- "RWTexture1D<uint4> tex : register(u1, space2);"},
- GlslStorageTextureData{ast::TextureDimension::k2d,
- ast::ImageFormat::kRgba16Sint,
- "RWTexture2D<int4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k2dArray, ast::ImageFormat::kR32Sint,
- "RWTexture2DArray<int4> tex : register(u1, space2);"},
- GlslStorageTextureData{ast::TextureDimension::k3d,
- ast::ImageFormat::kRg32Sint,
- "RWTexture3D<int4> tex : register(u1, space2);"},
- GlslStorageTextureData{
- ast::TextureDimension::k1d, ast::ImageFormat::kRgba32Sint,
- "RWTexture1D<int4> tex : register(u1, space2);"}));
+ testing::Values(GlslStorageTextureData{ast::TextureDimension::k1d,
+ ast::ImageFormat::kRgba8Unorm,
+ "RWTexture1D<float4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2d,
+ ast::ImageFormat::kRgba16Float,
+ "RWTexture2D<float4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2dArray,
+ ast::ImageFormat::kR32Float,
+ "RWTexture2DArray<float4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k3d,
+ ast::ImageFormat::kRg32Float,
+ "RWTexture3D<float4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k1d,
+ ast::ImageFormat::kRgba32Float,
+ "RWTexture1D<float4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2d,
+ ast::ImageFormat::kRgba16Uint,
+ "RWTexture2D<uint4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2dArray,
+ ast::ImageFormat::kR32Uint,
+ "RWTexture2DArray<uint4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k3d,
+ ast::ImageFormat::kRg32Uint,
+ "RWTexture3D<uint4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k1d,
+ ast::ImageFormat::kRgba32Uint,
+ "RWTexture1D<uint4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2d,
+ ast::ImageFormat::kRgba16Sint,
+ "RWTexture2D<int4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k2dArray,
+ ast::ImageFormat::kR32Sint,
+ "RWTexture2DArray<int4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k3d,
+ ast::ImageFormat::kRg32Sint,
+ "RWTexture3D<int4> tex;"},
+ GlslStorageTextureData{ast::TextureDimension::k1d,
+ ast::ImageFormat::kRgba32Sint,
+ "RWTexture1D<int4> tex;"}));
} // namespace
} // namespace glsl
diff --git a/src/writer/glsl/generator_impl_variable_decl_statement_test.cc b/src/writer/glsl/generator_impl_variable_decl_statement_test.cc
index 3db868b..8d5cf90 100644
--- a/src/writer/glsl/generator_impl_variable_decl_statement_test.cc
+++ b/src/writer/glsl/generator_impl_variable_decl_statement_test.cc
@@ -76,7 +76,7 @@
gen.increment_indent();
ASSERT_TRUE(gen.Generate()) << gen.error();
- EXPECT_THAT(gen.result(), HasSubstr(" static float a = 0.0f;\n"));
+ EXPECT_THAT(gen.result(), HasSubstr(" float a = 0.0f;\n"));
}
TEST_F(GlslGeneratorImplTest_VariableDecl,
