wgsl: Clean up duplicated [[block]] structs
crbug.com/tint/386 has been fixed.
We don't need these work arounds any more.
Original CL by bclayton@:
https://dawn-review.googlesource.com/c/dawn/+/54641
Bug: tint:386
Fixed: dawn:943
Fixed: dawn:975
Change-Id: I1b94d1e33b023b7d9b3153094f829885a75b5a5e
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/57743
Commit-Queue: Austin Eng <enga@chromium.org>
Auto-Submit: Austin Eng <enga@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
diff --git a/src/tests/end2end/BindGroupTests.cpp b/src/tests/end2end/BindGroupTests.cpp
index 763977c..be1a48d 100644
--- a/src/tests/end2end/BindGroupTests.cpp
+++ b/src/tests/end2end/BindGroupTests.cpp
@@ -338,18 +338,12 @@
wgpu::ShaderModule vsModule = utils::CreateShaderModule(device, R"(
// TODO(crbug.com/tint/369): Use a mat2x2 when Tint translates it correctly.
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct VertexUniformBuffer1 {
+ [[block]] struct VertexUniformBuffer {
transform : vec4<f32>;
};
- [[block]] struct VertexUniformBuffer2 {
- transform : vec4<f32>;
- };
-
- // TODO(crbug.com/tint/386): Use the same struct definition.
- [[group(0), binding(0)]] var <uniform> vertexUbo1 : VertexUniformBuffer1;
- [[group(1), binding(0)]] var <uniform> vertexUbo2 : VertexUniformBuffer2;
+ [[group(0), binding(0)]] var <uniform> vertexUbo1 : VertexUniformBuffer;
+ [[group(1), binding(0)]] var <uniform> vertexUbo2 : VertexUniformBuffer;
[[stage(vertex)]]
fn main([[builtin(vertex_index)]] VertexIndex : u32) -> [[builtin(position)]] vec4<f32> {
@@ -365,18 +359,12 @@
})");
wgpu::ShaderModule fsModule = utils::CreateShaderModule(device, R"(
- // TODO(crbug.com/tint/386): Use the same struct
- [[block]] struct FragmentUniformBuffer1 {
+ [[block]] struct FragmentUniformBuffer {
color : vec4<f32>;
};
- [[block]] struct FragmentUniformBuffer2 {
- color : vec4<f32>;
- };
-
- // TODO(crbug.com/tint/386): Use the same struct definition.
- [[group(0), binding(1)]] var <uniform> fragmentUbo1 : FragmentUniformBuffer1;
- [[group(1), binding(1)]] var <uniform> fragmentUbo2 : FragmentUniformBuffer2;
+ [[group(0), binding(1)]] var <uniform> fragmentUbo1 : FragmentUniformBuffer;
+ [[group(1), binding(1)]] var <uniform> fragmentUbo2 : FragmentUniformBuffer;
[[stage(fragment)]] fn main() -> [[location(0)]] vec4<f32> {
return fragmentUbo1.color + fragmentUbo2.color;
@@ -770,6 +758,9 @@
// Regression test for crbug.com/dawn/408 where dynamic offsets were applied in the wrong order.
// Dynamic offsets should be applied in increasing order of binding number.
TEST_P(BindGroupTests, DynamicOffsetOrder) {
+ // Does not work with SPIRV-Cross. crbug.com/dawn/975
+ DAWN_SUPPRESS_TEST_IF(IsD3D12() && !HasToggleEnabled("use_tint_generator"));
+
// We will put the following values and the respective offsets into a buffer.
// The test will ensure that the correct dynamic offset is applied to each buffer by reading the
// value from an offset binding.
@@ -819,16 +810,7 @@
wgpu::ComputePipelineDescriptor pipelineDescriptor;
pipelineDescriptor.compute.module = utils::CreateShaderModule(device, R"(
- // TODO(crbug.com/tint/386): Use the same struct
- [[block]] struct Buffer0 {
- value : u32;
- };
-
- [[block]] struct Buffer2 {
- value : u32;
- };
-
- [[block]] struct Buffer3 {
+ [[block]] struct Buffer {
value : u32;
};
@@ -836,9 +818,9 @@
value : vec3<u32>;
};
- [[group(0), binding(2)]] var<uniform> buffer2 : Buffer2;
- [[group(0), binding(3)]] var<storage, read> buffer3 : Buffer3;
- [[group(0), binding(0)]] var<storage, read> buffer0 : Buffer0;
+ [[group(0), binding(2)]] var<uniform> buffer2 : Buffer;
+ [[group(0), binding(3)]] var<storage, read> buffer3 : Buffer;
+ [[group(0), binding(0)]] var<storage, read> buffer0 : Buffer;
[[group(0), binding(4)]] var<storage, read_write> outputBuffer : OutputBuffer;
[[stage(compute), workgroup_size(1)]] fn main() {
@@ -948,23 +930,13 @@
})");
wgpu::ShaderModule fsModule = utils::CreateShaderModule(device, R"(
- // TODO(crbug.com/tint/386): Use the same struct
- [[block]] struct Ubo1 {
+ [[block]] struct Ubo {
color : vec4<f32>;
};
- [[block]] struct Ubo2 {
- color : vec4<f32>;
- };
-
- [[block]] struct Ubo3 {
- color : vec4<f32>;
- };
-
- // TODO(crbug.com/tint/386): Use the same struct definition.
- [[group(0), binding(953)]] var <uniform> ubo1 : Ubo1;
- [[group(0), binding(47)]] var <uniform> ubo2 : Ubo2;
- [[group(0), binding(111)]] var <uniform> ubo3 : Ubo3;
+ [[group(0), binding(953)]] var <uniform> ubo1 : Ubo;
+ [[group(0), binding(47)]] var <uniform> ubo2 : Ubo;
+ [[group(0), binding(111)]] var <uniform> ubo3 : Ubo;
[[stage(fragment)]] fn main() -> [[location(0)]] vec4<f32> {
return ubo1.color + 2.0 * ubo2.color + 4.0 * ubo3.color;
diff --git a/src/tests/end2end/ComputeCopyStorageBufferTests.cpp b/src/tests/end2end/ComputeCopyStorageBufferTests.cpp
index bdf9093..6b60a0e 100644
--- a/src/tests/end2end/ComputeCopyStorageBufferTests.cpp
+++ b/src/tests/end2end/ComputeCopyStorageBufferTests.cpp
@@ -88,16 +88,12 @@
// Test that a trivial compute-shader memcpy implementation works.
TEST_P(ComputeCopyStorageBufferTests, SizedArrayOfBasic) {
BasicTest(R"(
- [[block]] struct Buf1 {
- s : array<vec4<u32>, 4>;
- };
- [[block]] struct Buf2 {
+ [[block]] struct Buf {
s : array<vec4<u32>, 4>;
};
- // TODO(crbug.com/tint/386): Use the same struct type
- [[group(0), binding(0)]] var<storage, read_write> src : Buf1;
- [[group(0), binding(1)]] var<storage, read_write> dst : Buf2;
+ [[group(0), binding(0)]] var<storage, read_write> src : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> dst : Buf;
[[stage(compute), workgroup_size(1)]]
fn main([[builtin(global_invocation_id)]] GlobalInvocationID : vec3<u32>) {
@@ -115,16 +111,12 @@
b : vec2<u32>;
};
- [[block]] struct Buf1 {
- s : array<S, 4>;
- };
- [[block]] struct Buf2 {
+ [[block]] struct Buf {
s : array<S, 4>;
};
- // TODO(crbug.com/tint/386): Use the same struct type
- [[group(0), binding(0)]] var<storage, read_write> src : Buf1;
- [[group(0), binding(1)]] var<storage, read_write> dst : Buf2;
+ [[group(0), binding(0)]] var<storage, read_write> src : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> dst : Buf;
[[stage(compute), workgroup_size(1)]]
fn main([[builtin(global_invocation_id)]] GlobalInvocationID : vec3<u32>) {
@@ -137,16 +129,12 @@
// Test that a trivial compute-shader memcpy implementation works.
TEST_P(ComputeCopyStorageBufferTests, UnsizedArrayOfBasic) {
BasicTest(R"(
- [[block]] struct Buf1 {
- s : array<vec4<u32>>;
- };
- [[block]] struct Buf2 {
+ [[block]] struct Buf {
s : array<vec4<u32>>;
};
- // TODO(crbug.com/tint/386): Use the same struct type
- [[group(0), binding(0)]] var<storage, read_write> src : Buf1;
- [[group(0), binding(1)]] var<storage, read_write> dst : Buf2;
+ [[group(0), binding(0)]] var<storage, read_write> src : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> dst : Buf;
[[stage(compute), workgroup_size(1)]]
fn main([[builtin(global_invocation_id)]] GlobalInvocationID : vec3<u32>) {
diff --git a/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp b/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp
index 2cd7aa4..fd7e4ee 100644
--- a/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp
+++ b/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp
@@ -82,17 +82,12 @@
device, data.data(), bufferSize, wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc);
wgpu::ShaderModule module = utils::CreateShaderModule(device, R"(
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct Src {
+ [[block]] struct Buf {
data : array<u32, 100>;
};
- [[block]] struct Dst {
- data : array<u32, 100>;
- };
-
- [[group(0), binding(0)]] var<storage, read_write> src : Src;
- [[group(0), binding(1)]] var<storage, read_write> dst : Dst;
+ [[group(0), binding(0)]] var<storage, read_write> src : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> dst : Buf;
[[stage(compute), workgroup_size(1)]]
fn main([[builtin(global_invocation_id)]] GlobalInvocationID : vec3<u32>) {
@@ -153,17 +148,12 @@
device, data.data(), bufferSize, wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc);
wgpu::ShaderModule module = utils::CreateShaderModule(device, R"(
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct Src {
+ [[block]] struct Buf {
data : array<u32, 100>;
};
- [[block]] struct Dst {
- data : array<u32, 100>;
- };
-
- [[group(0), binding(0)]] var<storage, read> src : Src;
- [[group(0), binding(1)]] var<storage, read_write> dst : Dst;
+ [[group(0), binding(0)]] var<storage, read> src : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> dst : Buf;
[[stage(compute), workgroup_size(1)]]
fn main([[builtin(global_invocation_id)]] GlobalInvocationID : vec3<u32>) {
diff --git a/src/tests/end2end/DynamicBufferOffsetTests.cpp b/src/tests/end2end/DynamicBufferOffsetTests.cpp
index 5f91cab..ba46d37 100644
--- a/src/tests/end2end/DynamicBufferOffsetTests.cpp
+++ b/src/tests/end2end/DynamicBufferOffsetTests.cpp
@@ -107,36 +107,19 @@
std::ostringstream fs;
std::string multipleNumber = isInheritedPipeline ? "2" : "1";
fs << R"(
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct Buffer1 {
+ [[block]] struct Buf {
value : vec2<u32>;
};
- [[block]] struct Buffer2 {
- value : vec2<u32>;
- };
-
- [[block]] struct Buffer3 {
- value : vec2<u32>;
- };
-
- [[block]] struct Buffer4 {
- value : vec2<u32>;
- };
-
- [[group(0), binding(0)]] var<uniform> uBufferNotDynamic : Buffer1;
- [[group(0), binding(1)]] var<storage, read_write> sBufferNotDynamic : Buffer2;
- [[group(0), binding(3)]] var<uniform> uBuffer : Buffer3;
- [[group(0), binding(4)]] var<storage, read_write> sBuffer : Buffer4;
+ [[group(0), binding(0)]] var<uniform> uBufferNotDynamic : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> sBufferNotDynamic : Buf;
+ [[group(0), binding(3)]] var<uniform> uBuffer : Buf;
+ [[group(0), binding(4)]] var<storage, read_write> sBuffer : Buf;
)";
if (isInheritedPipeline) {
fs << R"(
- [[block]] struct Buffer5 {
- value : vec2<u32>;
- };
-
- [[group(1), binding(0)]] var<uniform> paddingBlock : Buffer5;
+ [[group(1), binding(0)]] var<uniform> paddingBlock : Buf;
)";
}
@@ -144,7 +127,7 @@
fs << R"(
[[stage(fragment)]] fn main() -> [[location(0)]] vec4<f32> {
sBufferNotDynamic.value = uBufferNotDynamic.value.xy;
- sBuffer.value = vec2<u32>(multipleNumber, multipleNumber) * (uBuffer.value.xy + sBufferNotDynamic.value.xy);
+ sBuffer.value = vec2<u32>(multipleNumber, multipleNumber) * (uBuffer.value.xy + uBufferNotDynamic.value.xy);
return vec4<f32>(f32(uBuffer.value.x) / 255.0, f32(uBuffer.value.y) / 255.0,
1.0, 1.0);
}
@@ -174,36 +157,19 @@
std::ostringstream cs;
std::string multipleNumber = isInheritedPipeline ? "2" : "1";
cs << R"(
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct Buffer1 {
+ [[block]] struct Buf {
value : vec2<u32>;
};
- [[block]] struct Buffer2 {
- value : vec2<u32>;
- };
-
- [[block]] struct Buffer3 {
- value : vec2<u32>;
- };
-
- [[block]] struct Buffer4 {
- value : vec2<u32>;
- };
-
- [[group(0), binding(0)]] var<uniform> uBufferNotDynamic : Buffer1;
- [[group(0), binding(1)]] var<storage, read_write> sBufferNotDynamic : Buffer2;
- [[group(0), binding(3)]] var<uniform> uBuffer : Buffer3;
- [[group(0), binding(4)]] var<storage, read_write> sBuffer : Buffer4;
+ [[group(0), binding(0)]] var<uniform> uBufferNotDynamic : Buf;
+ [[group(0), binding(1)]] var<storage, read_write> sBufferNotDynamic : Buf;
+ [[group(0), binding(3)]] var<uniform> uBuffer : Buf;
+ [[group(0), binding(4)]] var<storage, read_write> sBuffer : Buf;
)";
if (isInheritedPipeline) {
cs << R"(
- [[block]] struct Buffer5 {
- value : vec2<u32>;
- };
-
- [[group(1), binding(0)]] var<uniform> paddingBlock : Buffer5;
+ [[group(1), binding(0)]] var<uniform> paddingBlock : Buf;
)";
}
@@ -211,7 +177,7 @@
cs << R"(
[[stage(compute), workgroup_size(1)]] fn main() {
sBufferNotDynamic.value = uBufferNotDynamic.value.xy;
- sBuffer.value = vec2<u32>(multipleNumber, multipleNumber) * (uBuffer.value.xy + sBufferNotDynamic.value.xy);
+ sBuffer.value = vec2<u32>(multipleNumber, multipleNumber) * (uBuffer.value.xy + uBufferNotDynamic.value.xy);
}
)";
@@ -256,7 +222,7 @@
}
// Have non-zero dynamic offsets.
-TEST_P(DynamicBufferOffsetTests, SetDynamicOffestsRenderPipeline) {
+TEST_P(DynamicBufferOffsetTests, SetDynamicOffsetsRenderPipeline) {
wgpu::RenderPipeline pipeline = CreateRenderPipeline();
utils::BasicRenderPass renderPass = utils::CreateBasicRenderPass(device, kRTSize, kRTSize);
@@ -280,6 +246,9 @@
// Dynamic offsets are all zero and no effect to result.
TEST_P(DynamicBufferOffsetTests, BasicComputePipeline) {
+ // TODO(crbug.com/dawn/978): Failing on Windows Vulkan NVIDIA
+ DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsNvidia());
+
wgpu::ComputePipeline pipeline = CreateComputePipeline();
std::array<uint32_t, 2> offsets = {0, 0};
@@ -298,7 +267,10 @@
}
// Have non-zero dynamic offsets.
-TEST_P(DynamicBufferOffsetTests, SetDynamicOffestsComputePipeline) {
+TEST_P(DynamicBufferOffsetTests, SetDynamicOffsetsComputePipeline) {
+ // TODO(crbug.com/dawn/978): Failing on Windows Vulkan NVIDIA
+ DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsNvidia());
+
wgpu::ComputePipeline pipeline = CreateComputePipeline();
std::array<uint32_t, 2> offsets = {kMinUniformBufferOffsetAlignment,
@@ -319,7 +291,7 @@
}
// Test inherit dynamic offsets on render pipeline
-TEST_P(DynamicBufferOffsetTests, InheritDynamicOffestsRenderPipeline) {
+TEST_P(DynamicBufferOffsetTests, InheritDynamicOffsetsRenderPipeline) {
// Using default pipeline and setting dynamic offsets
wgpu::RenderPipeline pipeline = CreateRenderPipeline();
wgpu::RenderPipeline testPipeline = CreateRenderPipeline(true);
@@ -351,7 +323,7 @@
// TODO(shaobo.yan@intel.com) : Try this test on GTX1080 and cannot reproduce the failure.
// Suspect it is due to dawn doesn't handle sync between two dispatch and disable this case.
// Will double check root cause after got GTX1660.
-TEST_P(DynamicBufferOffsetTests, InheritDynamicOffestsComputePipeline) {
+TEST_P(DynamicBufferOffsetTests, InheritDynamicOffsetsComputePipeline) {
DAWN_SUPPRESS_TEST_IF(IsWindows());
wgpu::ComputePipeline pipeline = CreateComputePipeline();
wgpu::ComputePipeline testPipeline = CreateComputePipeline(true);
@@ -377,7 +349,7 @@
}
// Setting multiple dynamic offsets for the same bindgroup in one render pass.
-TEST_P(DynamicBufferOffsetTests, UpdateDynamicOffestsMultipleTimesRenderPipeline) {
+TEST_P(DynamicBufferOffsetTests, UpdateDynamicOffsetsMultipleTimesRenderPipeline) {
// Using default pipeline and setting dynamic offsets
wgpu::RenderPipeline pipeline = CreateRenderPipeline();
diff --git a/src/tests/end2end/GpuMemorySynchronizationTests.cpp b/src/tests/end2end/GpuMemorySynchronizationTests.cpp
index d125857..bafab67 100644
--- a/src/tests/end2end/GpuMemorySynchronizationTests.cpp
+++ b/src/tests/end2end/GpuMemorySynchronizationTests.cpp
@@ -519,15 +519,11 @@
};
[[group(0), binding(1)]] var<storage, read_write> ibContents : IBContents;
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct ColorContents1 {
+ [[block]] struct ColorContents {
color : f32;
};
- [[block]] struct ColorContents2 {
- color : f32;
- };
- [[group(0), binding(2)]] var<storage, read_write> uniformContents : ColorContents1;
- [[group(0), binding(3)]] var<storage, read_write> storageContents : ColorContents2;
+ [[group(0), binding(2)]] var<storage, read_write> uniformContents : ColorContents;
+ [[group(0), binding(3)]] var<storage, read_write> storageContents : ColorContents;
[[stage(compute), workgroup_size(1)]] fn main() {
vbContents.pos[0] = vec4<f32>(-1.0, 1.0, 0.0, 1.0);
diff --git a/src/tests/end2end/OpArrayLengthTests.cpp b/src/tests/end2end/OpArrayLengthTests.cpp
index 2f2ad52..105a754 100644
--- a/src/tests/end2end/OpArrayLengthTests.cpp
+++ b/src/tests/end2end/OpArrayLengthTests.cpp
@@ -53,20 +53,15 @@
// Common shader code to use these buffers in shaders, assuming they are in bindgroup index
// 0.
mShaderInterface = R"(
- // TODO(crbug.com/tint/386): Use the same struct.
- [[block]] struct DataBuffer1 {
- data : [[stride(4)]] array<f32>;
- };
-
- [[block]] struct DataBuffer2 {
+ [[block]] struct DataBuffer {
data : [[stride(4)]] array<f32>;
};
// The length should be 1 because the buffer is 4-byte long.
- [[group(0), binding(0)]] var<storage, read> buffer1 : DataBuffer1;
+ [[group(0), binding(0)]] var<storage, read> buffer1 : DataBuffer;
// The length should be 64 because the buffer is 256 bytes long.
- [[group(0), binding(1)]] var<storage, read> buffer2 : DataBuffer2;
+ [[group(0), binding(1)]] var<storage, read> buffer2 : DataBuffer;
// The length should be (512 - 16*4) / 8 = 56 because the buffer is 512 bytes long
// and the structure is 8 bytes big.
