From bbabda35908676d0085e34475de460f06f650e48 Mon Sep 17 00:00:00 2001 From: Ryan Harrison Date: Thu, 18 Mar 2021 17:20:48 +0000 Subject: [PATCH] Fix Storage Buffers in WGSL tests & examples Converts var to var Adds in [[access(read_write)]] when no access qualifier was supplied. BUG=dawn:699 Change-Id: I9da95366d0bb8734f1d134ee2dc764a7c16ef0cf Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/45300 Commit-Queue: Ryan Harrison Commit-Queue: Corentin Wallez Auto-Submit: Ryan Harrison Reviewed-by: Corentin Wallez Reviewed-by: Austin Eng --- examples/ComputeBoids.cpp | 4 ++-- src/dawn_native/QueryHelper.cpp | 4 ++-- src/tests/end2end/BindGroupTests.cpp | 16 +++++++------- src/tests/end2end/BufferZeroInitTests.cpp | 2 +- .../end2end/ComputeCopyStorageBufferTests.cpp | 12 +++++----- src/tests/end2end/ComputeDispatchTests.cpp | 2 +- .../ComputeStorageBufferBarrierTests.cpp | 14 ++++++------ .../end2end/CopyTextureForBrowserTests.cpp | 2 +- .../end2end/CreatePipelineAsyncTests.cpp | 4 ++-- src/tests/end2end/D3D12CachingTests.cpp | 2 +- .../end2end/DepthStencilSamplingTests.cpp | 10 ++++----- .../end2end/DynamicBufferOffsetTests.cpp | 8 +++---- src/tests/end2end/EntryPointTests.cpp | 2 +- src/tests/end2end/FirstIndexOffsetTests.cpp | 2 +- .../end2end/GpuMemorySynchronizationTests.cpp | 22 +++++++++---------- .../end2end/MultisampledSamplingTests.cpp | 2 +- src/tests/end2end/OpArrayLengthTests.cpp | 8 +++---- src/tests/end2end/ShaderTests.cpp | 2 +- src/tests/end2end/StorageTextureTests.cpp | 6 ++--- src/tests/end2end/TextureZeroInitTests.cpp | 2 +- .../validation/BindGroupValidationTests.cpp | 18 +++++++-------- .../GetBindGroupLayoutValidationTests.cpp | 12 +++++----- .../RenderBundleValidationTests.cpp | 2 +- .../RenderPipelineValidationTests.cpp | 2 +- .../validation/ResourceUsageTrackingTests.cpp | 4 ++-- .../StorageTextureValidationTests.cpp | 2 +- 26 files changed, 83 insertions(+), 83 deletions(-) diff --git a/examples/ComputeBoids.cpp b/examples/ComputeBoids.cpp index 484105ffbe..272e28bb25 100644 --- a/examples/ComputeBoids.cpp +++ b/examples/ComputeBoids.cpp @@ -168,8 +168,8 @@ void initSim() { particles : array; }; [[binding(0), group(0)]] var params : SimParams; - [[binding(1), group(0)]] var particlesA : [[access(read)]] Particles; - [[binding(2), group(0)]] var particlesB : [[access(read_write)]] Particles; + [[binding(1), group(0)]] var particlesA : [[access(read)]] Particles; + [[binding(2), group(0)]] var particlesB : [[access(read_write)]] Particles; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; // https://github.com/austinEng/Project6-Vulkan-Flocking/blob/master/data/shaders/computeparticles/particle.comp diff --git a/src/dawn_native/QueryHelper.cpp b/src/dawn_native/QueryHelper.cpp index 9b84e307d8..4d8bc3fadf 100644 --- a/src/dawn_native/QueryHelper.cpp +++ b/src/dawn_native/QueryHelper.cpp @@ -53,9 +53,9 @@ namespace dawn_native { }; [[group(0), binding(0)]] - var timestamps : [[access(read_write)]] TimestampArr; + var timestamps : [[access(read_write)]] TimestampArr; [[group(0), binding(1)]] - var availability : [[access(read)]] AvailabilityArr; + var availability : [[access(read)]] AvailabilityArr; [[group(0), binding(2)]] var params : TimestampParams; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; diff --git a/src/tests/end2end/BindGroupTests.cpp b/src/tests/end2end/BindGroupTests.cpp index 5f97b0f834..a518e89062 100644 --- a/src/tests/end2end/BindGroupTests.cpp +++ b/src/tests/end2end/BindGroupTests.cpp @@ -76,7 +76,7 @@ class BindGroupTests : public DawnTest { << " : Buffer" << i << ";"; break; case wgpu::BufferBindingType::Storage: - fs << "\n[[group(" << i << "), binding(0)]] var buffer" << i + fs << "\n[[group(" << i << "), binding(0)]] var buffer" << i << " : [[access(read)]] Buffer" << i << ";"; break; default: @@ -854,9 +854,9 @@ TEST_P(BindGroupTests, DynamicOffsetOrder) { }; [[group(0), binding(2)]] var buffer2 : Buffer2; - [[group(0), binding(3)]] var buffer3 : [[access(read)]] Buffer3; - [[group(0), binding(0)]] var buffer0 : [[access(read)]] Buffer0; - [[group(0), binding(4)]] var outputBuffer : [[access(read_write)]] OutputBuffer; + [[group(0), binding(3)]] var buffer3 : [[access(read)]] Buffer3; + [[group(0), binding(0)]] var buffer0 : [[access(read)]] Buffer0; + [[group(0), binding(4)]] var outputBuffer : [[access(read_write)]] OutputBuffer; [[stage(compute)]] fn main() -> void { outputBuffer.value = vec3(buffer0.value, buffer2.value, buffer3.value); @@ -1123,7 +1123,7 @@ TEST_P(BindGroupTests, ReadonlyStorage) { [[block]] struct Buffer0 { color : vec4; }; - [[group(0), binding(0)]] var buffer0 : [[access(read)]] Buffer0; + [[group(0), binding(0)]] var buffer0 : [[access(read)]] Buffer0; [[location(0)]] var fragColor : vec4; [[stage(fragment)]] fn main() -> void { @@ -1261,8 +1261,8 @@ TEST_P(BindGroupTests, ReallyLargeBindGroup) { }; )"; interface << "[[group(0), binding(" << binding++ << ")]] " - << "var sbuf" << i << " : [[access(read)]] ReadOnlyStorageBuffer" - << i << ";\n"; + << "var sbuf" << i << " : [[access(read)]] ReadOnlyStorageBuffer" << i + << ";\n"; body << "if (sbuf" << i << ".value != " << expectedValue++ << "u) {\n"; body << " return;\n"; @@ -1278,7 +1278,7 @@ TEST_P(BindGroupTests, ReallyLargeBindGroup) { }; )"; interface << "[[group(0), binding(" << binding++ << ")]] " - << "var result : [[access(read_write)]] ReadWriteStorageBuffer;\n"; + << "var result : [[access(read_write)]] ReadWriteStorageBuffer;\n"; body << "result.value = 1u;\n"; diff --git a/src/tests/end2end/BufferZeroInitTests.cpp b/src/tests/end2end/BufferZeroInitTests.cpp index 8806881121..327723e43f 100644 --- a/src/tests/end2end/BufferZeroInitTests.cpp +++ b/src/tests/end2end/BufferZeroInitTests.cpp @@ -1037,7 +1037,7 @@ TEST_P(BufferZeroInitTest, BoundAsReadonlyStorageBuffer) { [[block]] struct SSBO { value : vec4; }; - [[group(0), binding(0)]] var ssbo : SSBO; + [[group(0), binding(0)]] var ssbo : [[access(read_write)]] SSBO; [[group(0), binding(1)]] var outImage : [[access(write)]] texture_storage_2d; [[stage(compute)]] fn main() -> void { diff --git a/src/tests/end2end/ComputeCopyStorageBufferTests.cpp b/src/tests/end2end/ComputeCopyStorageBufferTests.cpp index c73bdf521d..132d71d7f1 100644 --- a/src/tests/end2end/ComputeCopyStorageBufferTests.cpp +++ b/src/tests/end2end/ComputeCopyStorageBufferTests.cpp @@ -96,8 +96,8 @@ TEST_P(ComputeCopyStorageBufferTests, SizedArrayOfBasic) { }; // TODO(crbug.com/tint/386): Use the same struct type - [[set(0), binding(0)]] var src : Buf1; - [[set(0), binding(1)]] var dst : Buf2; + [[set(0), binding(0)]] var src : [[access(read_write)]] Buf1; + [[set(0), binding(1)]] var dst : [[access(read_write)]] Buf2; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; @@ -124,8 +124,8 @@ TEST_P(ComputeCopyStorageBufferTests, SizedArrayOfStruct) { }; // TODO(crbug.com/tint/386): Use the same struct type - [[set(0), binding(0)]] var src : Buf1; - [[set(0), binding(1)]] var dst : Buf2; + [[set(0), binding(0)]] var src : [[access(read_write)]] Buf1; + [[set(0), binding(1)]] var dst : [[access(read_write)]] Buf2; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; @@ -147,8 +147,8 @@ TEST_P(ComputeCopyStorageBufferTests, UnsizedArrayOfBasic) { }; // TODO(crbug.com/tint/386): Use the same struct type - [[set(0), binding(0)]] var src : Buf1; - [[set(0), binding(1)]] var dst : Buf2; + [[set(0), binding(0)]] var src : [[access(read_write)]] Buf1; + [[set(0), binding(1)]] var dst : [[access(read_write)]] Buf2; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; diff --git a/src/tests/end2end/ComputeDispatchTests.cpp b/src/tests/end2end/ComputeDispatchTests.cpp index 779ed70a32..1886372e7e 100644 --- a/src/tests/end2end/ComputeDispatchTests.cpp +++ b/src/tests/end2end/ComputeDispatchTests.cpp @@ -37,7 +37,7 @@ class ComputeDispatchTests : public DawnTest { }; [[group(0), binding(0)]] var input : InputBuf; - [[group(0), binding(1)]] var output : OutputBuf; + [[group(0), binding(1)]] var output : [[access(read_write)]] OutputBuf; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; diff --git a/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp b/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp index c28fa0fdce..f8e98d5d79 100644 --- a/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp +++ b/src/tests/end2end/ComputeStorageBufferBarrierTests.cpp @@ -36,7 +36,7 @@ TEST_P(ComputeStorageBufferBarrierTests, AddIncrement) { data : array; }; - [[group(0), binding(0)]] var buf : [[access(read_write)]] Buf; + [[group(0), binding(0)]] var buf : [[access(read_write)]] Buf; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; @@ -92,8 +92,8 @@ TEST_P(ComputeStorageBufferBarrierTests, AddPingPong) { data : array; }; - [[group(0), binding(0)]] var src : [[access(read_write)]] Src; - [[group(0), binding(1)]] var dst : [[access(read_write)]] Dst; + [[group(0), binding(0)]] var src : [[access(read_write)]] Src; + [[group(0), binding(1)]] var dst : [[access(read_write)]] Dst; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; [[stage(compute)]] fn main() -> void { @@ -163,8 +163,8 @@ TEST_P(ComputeStorageBufferBarrierTests, StorageAndReadonlyStoragePingPongInOneP data : array; }; - [[group(0), binding(0)]] var src : [[access(read)]] Src; - [[group(0), binding(1)]] var dst : [[access(read_write)]] Dst; + [[group(0), binding(0)]] var src : [[access(read)]] Src; + [[group(0), binding(1)]] var dst : [[access(read_write)]] Dst; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; @@ -233,7 +233,7 @@ TEST_P(ComputeStorageBufferBarrierTests, UniformToStorageAddPingPong) { }; [[group(0), binding(0)]] var src : Buf; - [[group(0), binding(1)]] var dst : [[access(read_write)]] Buf; + [[group(0), binding(1)]] var dst : [[access(read_write)]] Buf; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; [[stage(compute)]] fn main() -> void { @@ -301,7 +301,7 @@ TEST_P(ComputeStorageBufferBarrierTests, UniformToStorageAddPingPongInOnePass) { }; [[group(0), binding(0)]] var src : Buf; - [[group(0), binding(1)]] var dst : [[access(read_write)]] Buf; + [[group(0), binding(1)]] var dst : [[access(read_write)]] Buf; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; [[stage(compute)]] fn main() -> void { diff --git a/src/tests/end2end/CopyTextureForBrowserTests.cpp b/src/tests/end2end/CopyTextureForBrowserTests.cpp index 3f6220fc75..deb3a7002e 100644 --- a/src/tests/end2end/CopyTextureForBrowserTests.cpp +++ b/src/tests/end2end/CopyTextureForBrowserTests.cpp @@ -80,7 +80,7 @@ class CopyTextureForBrowserTests : public DawnTest { }; [[group(0), binding(0)]] var src : texture_2d; [[group(0), binding(1)]] var dst : texture_2d; - [[group(0), binding(2)]] var output : OutputBuf; + [[group(0), binding(2)]] var output : [[access(read_write)]] OutputBuf; [[group(0), binding(3)]] var uniforms : Uniforms; [[builtin(global_invocation_id)]] var GlobalInvocationID : vec3; [[stage(compute), workgroup_size(1, 1, 1)]] diff --git a/src/tests/end2end/CreatePipelineAsyncTests.cpp b/src/tests/end2end/CreatePipelineAsyncTests.cpp index e5c6525f8b..88cd729a35 100644 --- a/src/tests/end2end/CreatePipelineAsyncTests.cpp +++ b/src/tests/end2end/CreatePipelineAsyncTests.cpp @@ -38,7 +38,7 @@ TEST_P(CreatePipelineAsyncTest, BasicUseOfCreateComputePipelineAsync) { [[block]] struct SSBO { value : u32; }; - [[group(0), binding(0)]] var ssbo : SSBO; + [[group(0), binding(0)]] var ssbo : [[access(read_write)]] SSBO; [[stage(compute)]] fn main() -> void { ssbo.value = 1u; @@ -105,7 +105,7 @@ TEST_P(CreatePipelineAsyncTest, CreateComputePipelineFailed) { [[block]] struct SSBO { value : u32; }; - [[group(0), binding(0)]] var ssbo : SSBO; + [[group(0), binding(0)]] var ssbo : [[access(read_write)]] SSBO; [[stage(compute)]] fn main() -> void { ssbo.value = 1u; diff --git a/src/tests/end2end/D3D12CachingTests.cpp b/src/tests/end2end/D3D12CachingTests.cpp index efe86bafc6..ea43ad17ae 100644 --- a/src/tests/end2end/D3D12CachingTests.cpp +++ b/src/tests/end2end/D3D12CachingTests.cpp @@ -229,7 +229,7 @@ TEST_P(D3D12CachingTests, ReuseShaderWithMultipleEntryPoints) { [[block]] struct Data { data : u32; }; - [[binding(0), group(0)]] var data : Data; + [[binding(0), group(0)]] var data : [[access(read_write)]] Data; [[stage(compute)]] fn write1() -> void { data.data = 1u; diff --git a/src/tests/end2end/DepthStencilSamplingTests.cpp b/src/tests/end2end/DepthStencilSamplingTests.cpp index bc1d153073..c410be9ee7 100644 --- a/src/tests/end2end/DepthStencilSamplingTests.cpp +++ b/src/tests/end2end/DepthStencilSamplingTests.cpp @@ -139,8 +139,8 @@ class DepthStencilSamplingTest : public DawnTest { << " : texture_2d;\n"; shaderSource << "[[group(0), binding(" << 2 * index + 1 - << ")]] var result" << index - << " : DepthResult;\n"; + << ")]] var result" << index + << " : [[access(read_write)]] DepthResult;\n"; shaderBody << "\nresult" << index << ".value = textureLoad(tex" << index << ", vec2(0, 0), 0)[" << componentIndex << "];"; @@ -150,8 +150,8 @@ class DepthStencilSamplingTest : public DawnTest { << " : texture_2d;\n"; shaderSource << "[[group(0), binding(" << 2 * index + 1 - << ")]] var result" << index - << " : StencilResult;\n"; + << ")]] var result" << index + << " : [[access(read_write)]] StencilResult;\n"; shaderBody << "\nresult" << index << ".value = textureLoad(tex" << index << ", vec2(0, 0), 0)[" << componentIndex << "];"; @@ -223,7 +223,7 @@ class DepthStencilSamplingTest : public DawnTest { [[block]] struct SamplerResult { value : f32; }; - [[group(0), binding(3)]] var samplerResult : SamplerResult; + [[group(0), binding(3)]] var samplerResult : [[access(read_write)]] SamplerResult; [[stage(compute)]] fn main() -> void { samplerResult.value = textureSampleCompare(tex, samp, vec2(0.5, 0.5), uniforms.compareRef); diff --git a/src/tests/end2end/DynamicBufferOffsetTests.cpp b/src/tests/end2end/DynamicBufferOffsetTests.cpp index dc730aa93a..dbaa8bdf9c 100644 --- a/src/tests/end2end/DynamicBufferOffsetTests.cpp +++ b/src/tests/end2end/DynamicBufferOffsetTests.cpp @@ -126,9 +126,9 @@ class DynamicBufferOffsetTests : public DawnTest { }; [[group(0), binding(0)]] var uBufferNotDynamic : Buffer1; - [[group(0), binding(1)]] var sBufferNotDynamic : [[access(read_write)]] Buffer2; + [[group(0), binding(1)]] var sBufferNotDynamic : [[access(read_write)]] Buffer2; [[group(0), binding(3)]] var uBuffer : Buffer3; - [[group(0), binding(4)]] var sBuffer : [[access(read_write)]] Buffer4; + [[group(0), binding(4)]] var sBuffer : [[access(read_write)]] Buffer4; )"; if (isInheritedPipeline) { @@ -195,9 +195,9 @@ class DynamicBufferOffsetTests : public DawnTest { }; [[group(0), binding(0)]] var uBufferNotDynamic : Buffer1; - [[group(0), binding(1)]] var sBufferNotDynamic : [[access(read_write)]] Buffer2; + [[group(0), binding(1)]] var sBufferNotDynamic : [[access(read_write)]] Buffer2; [[group(0), binding(3)]] var uBuffer : Buffer3; - [[group(0), binding(4)]] var sBuffer : [[access(read_write)]] Buffer4; + [[group(0), binding(4)]] var sBuffer : [[access(read_write)]] Buffer4; )"; if (isInheritedPipeline) { diff --git a/src/tests/end2end/EntryPointTests.cpp b/src/tests/end2end/EntryPointTests.cpp index e09eb86f4e..50cef2ebff 100644 --- a/src/tests/end2end/EntryPointTests.cpp +++ b/src/tests/end2end/EntryPointTests.cpp @@ -74,7 +74,7 @@ TEST_P(EntryPointTests, TwoComputeInModule) { [[block]] struct Data { data : u32; }; - [[binding(0), group(0)]] var data : Data; + [[binding(0), group(0)]] var data : [[access(read_write)]] Data; [[stage(compute)]] fn write1() -> void { data.data = 1u; diff --git a/src/tests/end2end/FirstIndexOffsetTests.cpp b/src/tests/end2end/FirstIndexOffsetTests.cpp index 6e3a2f8e73..3d9f7d02a0 100644 --- a/src/tests/end2end/FirstIndexOffsetTests.cpp +++ b/src/tests/end2end/FirstIndexOffsetTests.cpp @@ -119,7 +119,7 @@ void FirstIndexOffsetTests::TestImpl(DrawMode mode, instance_index : u32; }; - [[group(0), binding(0)]] var idx_vals : [[access(read_write)]] IndexVals; + [[group(0), binding(0)]] var idx_vals : [[access(read_write)]] IndexVals; [[stage(fragment)]] fn main() -> void { )"; diff --git a/src/tests/end2end/GpuMemorySynchronizationTests.cpp b/src/tests/end2end/GpuMemorySynchronizationTests.cpp index 8386fb2ca4..59de5d7130 100644 --- a/src/tests/end2end/GpuMemorySynchronizationTests.cpp +++ b/src/tests/end2end/GpuMemorySynchronizationTests.cpp @@ -39,7 +39,7 @@ class GpuMemorySyncTests : public DawnTest { [[block]] struct Data { a : i32; }; - [[group(0), binding(0)]] var data : [[access(read_write)]] Data; + [[group(0), binding(0)]] var data : [[access(read_write)]] Data; [[stage(compute)]] fn main() -> void { data.a = data.a + 1; })"); @@ -67,7 +67,7 @@ class GpuMemorySyncTests : public DawnTest { [[block]] struct Data { i : i32; }; - [[group(0), binding(0)]] var data : [[access(read_write)]] Data; + [[group(0), binding(0)]] var data : [[access(read_write)]] Data; [[location(0)]] var fragColor : vec4; [[stage(fragment)]] fn main() -> void { data.i = data.i + 1; @@ -257,7 +257,7 @@ TEST_P(GpuMemorySyncTests, SampledAndROStorageTextureInComputePass) { sampledOut: u32; storageOut: u32; }; - [[group(0), binding(0)]] var output : [[access(write)]] Output; + [[group(0), binding(0)]] var output : [[access(write)]] Output; [[group(0), binding(1)]] var sampledTex : texture_2d; [[group(0), binding(2)]] var storageTex : [[access(read)]] texture_storage_2d; @@ -317,7 +317,7 @@ class StorageToUniformSyncTests : public DawnTest { [[block]] struct Data { a : f32; }; - [[group(0), binding(0)]] var data : [[access(read_write)]] Data; + [[group(0), binding(0)]] var data : [[access(read_write)]] Data; [[stage(compute)]] fn main() -> void { data.a = 1.0; })"); @@ -516,12 +516,12 @@ TEST_P(MultipleWriteThenMultipleReadTests, SeparateBuffers) { [[block]] struct VBContents { pos : array, 4>; }; - [[group(0), binding(0)]] var vbContents : [[access(read_write)]] VBContents; + [[group(0), binding(0)]] var vbContents : [[access(read_write)]] VBContents; [[block]] struct IBContents { indices : array, 2>; }; - [[group(0), binding(1)]] var ibContents : [[access(read_write)]] IBContents; + [[group(0), binding(1)]] var ibContents : [[access(read_write)]] IBContents; // TODO(crbug.com/tint/386): Use the same struct. [[block]] struct ColorContents1 { @@ -530,8 +530,8 @@ TEST_P(MultipleWriteThenMultipleReadTests, SeparateBuffers) { [[block]] struct ColorContents2 { color : f32; }; - [[group(0), binding(2)]] var uniformContents : [[access(read_write)]] ColorContents1; - [[group(0), binding(3)]] var storageContents : [[access(read_write)]] ColorContents2; + [[group(0), binding(2)]] var uniformContents : [[access(read_write)]] ColorContents1; + [[group(0), binding(3)]] var storageContents : [[access(read_write)]] ColorContents2; [[stage(compute)]] fn main() -> void { vbContents.pos[0] = vec4(-1.0, 1.0, 0.0, 1.0); @@ -586,7 +586,7 @@ TEST_P(MultipleWriteThenMultipleReadTests, SeparateBuffers) { }; [[group(0), binding(0)]] var uniformBuffer : Buf; - [[group(0), binding(1)]] var storageBuffer : [[access(read)]] Buf; + [[group(0), binding(1)]] var storageBuffer : [[access(read)]] Buf; [[location(0)]] var fragColor : vec4; [[stage(fragment)]] fn main() -> void { @@ -648,7 +648,7 @@ TEST_P(MultipleWriteThenMultipleReadTests, OneBuffer) { [[align(256)]] color1 : f32; }; - [[group(0), binding(0)]] var contents : [[access(read_write)]] Contents; + [[group(0), binding(0)]] var contents : [[access(read_write)]] Contents; [[stage(compute)]] fn main() -> void { contents.pos[0] = vec4(-1.0, 1.0, 0.0, 1.0); @@ -703,7 +703,7 @@ TEST_P(MultipleWriteThenMultipleReadTests, OneBuffer) { color : f32; }; [[group(0), binding(0)]] var uniformBuffer : Buf; - [[group(0), binding(1)]] var storageBuffer : [[access(read)]] Buf; + [[group(0), binding(1)]] var storageBuffer : [[access(read)]] Buf; [[location(0)]] var fragColor : vec4; [[stage(fragment)]] fn main() -> void { diff --git a/src/tests/end2end/MultisampledSamplingTests.cpp b/src/tests/end2end/MultisampledSamplingTests.cpp index 5a6e1cf95d..37ec434e7c 100644 --- a/src/tests/end2end/MultisampledSamplingTests.cpp +++ b/src/tests/end2end/MultisampledSamplingTests.cpp @@ -96,7 +96,7 @@ class MultisampledSamplingTest : public DawnTest { colorSamples : array; depthSamples : array; }; - [[group(0), binding(2)]] var results : [[access(read_write)]] Results; + [[group(0), binding(2)]] var results : [[access(read_write)]] Results; [[stage(compute)]] fn main() -> void { for (var i : i32 = 0; i < 4; i = i + 1) { diff --git a/src/tests/end2end/OpArrayLengthTests.cpp b/src/tests/end2end/OpArrayLengthTests.cpp index 814e158824..3917f477ac 100644 --- a/src/tests/end2end/OpArrayLengthTests.cpp +++ b/src/tests/end2end/OpArrayLengthTests.cpp @@ -63,10 +63,10 @@ class OpArrayLengthTest : public DawnTest { }; // The length should be 1 because the buffer is 4-byte long. - [[group(0), binding(0)]] var buffer1 : [[access(read)]] DataBuffer1; + [[group(0), binding(0)]] var buffer1 : [[access(read)]] DataBuffer1; // The length should be 64 because the buffer is 256 bytes long. - [[group(0), binding(1)]] var buffer2 : [[access(read)]] DataBuffer2; + [[group(0), binding(1)]] var buffer2 : [[access(read)]] DataBuffer2; // The length should be (512 - 16*4) / 8 = 56 because the buffer is 512 bytes long // and the structure is 8 bytes big. @@ -79,7 +79,7 @@ class OpArrayLengthTest : public DawnTest { [[size(64)]] garbage : mat4x4; data : [[stride(8)]] array; }; - [[group(0), binding(2)]] var buffer3 : [[access(read)]] Buffer3; + [[group(0), binding(2)]] var buffer3 : [[access(read)]] Buffer3; )"; // See comments in the shader for an explanation of these values @@ -128,7 +128,7 @@ TEST_P(OpArrayLengthTest, Compute) { [[block]] struct ResultBuffer { data : [[stride(4)]] array; }; - [[group(1), binding(0)]] var result : [[access(read_write)]] ResultBuffer; + [[group(1), binding(0)]] var result : [[access(read_write)]] ResultBuffer; )" + mShaderInterface + R"( [[stage(compute)]] fn main() -> void { result.data[0] = arrayLength(buffer1.data); diff --git a/src/tests/end2end/ShaderTests.cpp b/src/tests/end2end/ShaderTests.cpp index 5b86ae0c59..3ae22c5371 100644 --- a/src/tests/end2end/ShaderTests.cpp +++ b/src/tests/end2end/ShaderTests.cpp @@ -35,7 +35,7 @@ TEST_P(ShaderTests, ComputeLog2) { data : array; }; -[[group(0), binding(0)]] var buf : [[access(read_write)]] Buf; +[[group(0), binding(0)]] var buf : [[access(read_write)]] Buf; [[stage(compute)]] fn main() -> void { const factor : f32 = 1.0001; diff --git a/src/tests/end2end/StorageTextureTests.cpp b/src/tests/end2end/StorageTextureTests.cpp index 9cfbfd3511..946ee01f33 100644 --- a/src/tests/end2end/StorageTextureTests.cpp +++ b/src/tests/end2end/StorageTextureTests.cpp @@ -712,7 +712,7 @@ TEST_P(StorageTextureTests, ReadonlyStorageTextureInComputeShader) { result : u32; }; -[[group(0), binding(1)]] var dstBuffer : DstBuffer; +[[group(0), binding(1)]] var dstBuffer : [[access(read_write)]] DstBuffer; )" << CommonReadOnlyTestCode(format) << R"( [[stage(compute)]] fn main() -> void { @@ -937,7 +937,7 @@ TEST_P(StorageTextureTests, Readonly2DArrayStorageTexture) { result : u32; }; -[[group(0), binding(1)]] var dstBuffer : DstBuffer; +[[group(0), binding(1)]] var dstBuffer : [[access(read_write)]] DstBuffer; )" << CommonReadOnlyTestCode(kTextureFormat, true) << R"( [[stage(compute)]] fn main() -> void { @@ -1207,7 +1207,7 @@ TEST_P(StorageTextureZeroInitTests, ReadonlyStorageTextureClearsToZeroInComputeP }; [[group(0), binding(0)]] var srcImage : [[access(read)]] texture_storage_2d; -[[group(0), binding(1)]] var dstBuffer : DstBuffer; +[[group(0), binding(1)]] var dstBuffer : [[access(read_write)]] DstBuffer; )") + kCommonReadOnlyZeroInitTestCode + R"( [[stage(compute)]] fn main() -> void { if (doTest()) { diff --git a/src/tests/end2end/TextureZeroInitTests.cpp b/src/tests/end2end/TextureZeroInitTests.cpp index 05c74f1a1b..8076c13743 100644 --- a/src/tests/end2end/TextureZeroInitTests.cpp +++ b/src/tests/end2end/TextureZeroInitTests.cpp @@ -977,7 +977,7 @@ TEST_P(TextureZeroInitTest, ComputePassSampledTextureClear) { [[block]] struct Result { value : vec4; }; - [[group(0), binding(1)]] var result : Result; + [[group(0), binding(1)]] var result : [[access(read_write)]] Result; [[stage(compute)]] fn main() -> void { result.value = textureLoad(tex, vec2(0,0), 0); } diff --git a/src/tests/unittests/validation/BindGroupValidationTests.cpp b/src/tests/unittests/validation/BindGroupValidationTests.cpp index 9f524c304e..71a9b92cbc 100644 --- a/src/tests/unittests/validation/BindGroupValidationTests.cpp +++ b/src/tests/unittests/validation/BindGroupValidationTests.cpp @@ -1094,8 +1094,8 @@ class SetBindGroupValidationTest : public ValidationTest { [[group(0), binding(0)]] var uBufferDynamic : S; [[group(0), binding(1)]] var uBuffer : S; - [[group(0), binding(2)]] var sBufferDynamic : [[access(read_write)]] S; - [[group(0), binding(3)]] var sReadonlyBufferDynamic : [[access(read)]] S; + [[group(0), binding(2)]] var sBufferDynamic : [[access(read_write)]] S; + [[group(0), binding(3)]] var sReadonlyBufferDynamic : [[access(read)]] S; [[stage(fragment)]] fn main() -> void { })"); @@ -1117,8 +1117,8 @@ class SetBindGroupValidationTest : public ValidationTest { [[group(0), binding(0)]] var uBufferDynamic : S; [[group(0), binding(1)]] var uBuffer : S; - [[group(0), binding(2)]] var sBufferDynamic : [[access(read_write)]] S; - [[group(0), binding(3)]] var sReadonlyBufferDynamic : [[access(read)]] S; + [[group(0), binding(2)]] var sBufferDynamic : [[access(read_write)]] S; + [[group(0), binding(3)]] var sReadonlyBufferDynamic : [[access(read)]] S; [[stage(compute), workgroup_size(4, 4, 1)]] fn main() -> void { })"); @@ -1543,7 +1543,7 @@ class SetBindGroupPersistenceValidationTest : public ValidationTest { ss << "[[group(" << l << "), binding(" << b << ")]] "; switch (binding) { case wgpu::BufferBindingType::Storage: - ss << "var set" << l << "_binding" << b + ss << "var set" << l << "_binding" << b << " : [[access(read_write)]] S;"; break; case wgpu::BufferBindingType::Uniform: @@ -1711,8 +1711,8 @@ class BindGroupLayoutCompatibilityTest : public ValidationTest { value : vec2; }; - [[group(0), binding(0)]] var sBufferDynamic : [[access(read_write)]] S; - [[group(1), binding(0)]] var sReadonlyBufferDynamic : [[access(read)]] S; + [[group(0), binding(0)]] var sBufferDynamic : [[access(read_write)]] S; + [[group(1), binding(0)]] var sReadonlyBufferDynamic : [[access(read)]] S; [[stage(fragment)]] fn main() -> void { })", @@ -1744,8 +1744,8 @@ class BindGroupLayoutCompatibilityTest : public ValidationTest { value : vec2; }; - [[group(0), binding(0)]] var sBufferDynamic : [[access(read_write)]] S; - [[group(1), binding(0)]] var sReadonlyBufferDynamic : [[access(read)]] S; + [[group(0), binding(0)]] var sBufferDynamic : [[access(read_write)]] S; + [[group(1), binding(0)]] var sReadonlyBufferDynamic : [[access(read)]] S; [[stage(compute), workgroup_size(4, 4, 1)]] fn main() -> void { })", diff --git a/src/tests/unittests/validation/GetBindGroupLayoutValidationTests.cpp b/src/tests/unittests/validation/GetBindGroupLayoutValidationTests.cpp index 485b3708fc..eacd5c137a 100644 --- a/src/tests/unittests/validation/GetBindGroupLayoutValidationTests.cpp +++ b/src/tests/unittests/validation/GetBindGroupLayoutValidationTests.cpp @@ -62,7 +62,7 @@ TEST_F(GetBindGroupLayoutTests, SameObject) { [[block]] struct S3 { pos : mat4x4; }; - [[group(3), binding(0)]] var storage3 : S3; + [[group(3), binding(0)]] var storage3 : [[access(read_write)]] S3; [[stage(fragment)]] fn main() -> void { })"); @@ -194,7 +194,7 @@ TEST_F(GetBindGroupLayoutTests, BindingType) { [[block]] struct S { pos : vec4; }; - [[group(0), binding(0)]] var ssbo : S; + [[group(0), binding(0)]] var ssbo : [[access(read_write)]] S; [[stage(fragment)]] fn main() -> void { })"); @@ -219,7 +219,7 @@ TEST_F(GetBindGroupLayoutTests, BindingType) { [[block]] struct S { pos : vec4; }; - [[group(0), binding(0)]] var ssbo : [[access(read)]] S; + [[group(0), binding(0)]] var ssbo : [[access(read)]] S; [[stage(fragment)]] fn main() -> void { })"); @@ -638,7 +638,7 @@ TEST_F(GetBindGroupLayoutTests, ConflictingBindingType) { [[block]] struct S { pos : vec4; }; - [[group(0), binding(0)]] var ssbo : S; + [[group(0), binding(0)]] var ssbo : [[access(read_write)]] S; [[stage(fragment)]] fn main() -> void { })"); @@ -827,8 +827,8 @@ TEST_F(GetBindGroupLayoutTests, DISABLED_FromCorrectEntryPoint) { [[block]] struct Data { data : f32; }; - [[binding 0, set 0]] var data0 : Data; - [[binding 1, set 0]] var data1 : Data; + [[binding 0, set 0]] var data0 : [[access(read_write)]] Data; + [[binding 1, set 0]] var data1 : [[access(read_write)]] Data; fn compute0() -> void { data0.data = 0.0; diff --git a/src/tests/unittests/validation/RenderBundleValidationTests.cpp b/src/tests/unittests/validation/RenderBundleValidationTests.cpp index fc27c08992..2790b44cb5 100644 --- a/src/tests/unittests/validation/RenderBundleValidationTests.cpp +++ b/src/tests/unittests/validation/RenderBundleValidationTests.cpp @@ -47,7 +47,7 @@ namespace { [[block]] struct Storage { dummy : array; }; - [[group(1), binding(1)]] var ssbo : [[access(read_write)]] Storage; + [[group(1), binding(1)]] var ssbo : [[access(read_write)]] Storage; [[stage(fragment)]] fn main() -> void { })"); diff --git a/src/tests/unittests/validation/RenderPipelineValidationTests.cpp b/src/tests/unittests/validation/RenderPipelineValidationTests.cpp index bd8da600b0..93175ea556 100644 --- a/src/tests/unittests/validation/RenderPipelineValidationTests.cpp +++ b/src/tests/unittests/validation/RenderPipelineValidationTests.cpp @@ -486,7 +486,7 @@ TEST_F(RenderPipelineValidationTest, StorageBufferInVertexShaderNoLayout) { [[block]] struct Dst { data : array; }; - [[group(0), binding(0)]] var dst : [[access(read_write)]] Dst; + [[group(0), binding(0)]] var dst : [[access(read_write)]] Dst; [[builtin(vertex_index)]] var VertexIndex : u32; [[stage(vertex)]] fn main() -> void { dst.data[VertexIndex] = 0x1234u; diff --git a/src/tests/unittests/validation/ResourceUsageTrackingTests.cpp b/src/tests/unittests/validation/ResourceUsageTrackingTests.cpp index 5f46c23be3..b15780735f 100644 --- a/src/tests/unittests/validation/ResourceUsageTrackingTests.cpp +++ b/src/tests/unittests/validation/ResourceUsageTrackingTests.cpp @@ -779,7 +779,7 @@ namespace { [[block]] struct RBuffer { value : f32; }; - [[group(0), binding(0)]] var rBuffer : [[access(read)]] RBuffer; + [[group(0), binding(0)]] var rBuffer : [[access(read)]] RBuffer; [[stage(fragment)]] fn main() -> void { })"); utils::ComboRenderPipelineDescriptor2 pipelineDescriptor; @@ -819,7 +819,7 @@ namespace { [[block]] struct RBuffer { value : f32; }; - [[group(0), binding(0)]] var rBuffer : [[access(read)]] RBuffer; + [[group(0), binding(0)]] var rBuffer : [[access(read)]] RBuffer; [[stage(compute)]] fn main() -> void { })"); wgpu::ComputePipelineDescriptor pipelineDescriptor; diff --git a/src/tests/unittests/validation/StorageTextureValidationTests.cpp b/src/tests/unittests/validation/StorageTextureValidationTests.cpp index 5abe2a5382..4b6c04240a 100644 --- a/src/tests/unittests/validation/StorageTextureValidationTests.cpp +++ b/src/tests/unittests/validation/StorageTextureValidationTests.cpp @@ -197,7 +197,7 @@ TEST_F(StorageTextureValidationTests, ComputePipeline) { [[block]] struct Buf { data : f32; }; - [[group(0), binding(1)]] var buf : [[access(read_write)]] Buf; + [[group(0), binding(1)]] var buf : [[access(read_write)]] Buf; [[stage(compute)]] fn main() -> void { buf.data = textureLoad(image0, vec2(LocalInvocationID.xy)).x;