From 3fae862a029c250ec0b0d3e4bd05a1102937022d Mon Sep 17 00:00:00 2001 From: Greggman Date: Fri, 10 Jan 2025 06:22:38 +0900 Subject: [PATCH] Compat: refactor state_tracking test for 0 frag buffers. (#4112) This is a first attempt. Feel free to push back and/or give ideas. The original tests use 2 read-only-storage buffers and 1 read-write storage buffer. Each has a single i32 in it and generally they substract the first 2 from the 2nd. Storage buffers in the fragment stage might not exist on some compat devices so the question is how to work around that and still test. This solution is to add subcases, `storage` and `uniform`. The `storage` case is unchanged. The compute pass case will run in compat always. The render pass and render bundle cases only run in compat if the device supports storage buffers in the fragment stage. The uniform cases use 2 uniform buffers and render to a single pixel r32sint texture. They then copy that texture to the `out` buffer that the original test was checking. This path needs no storage buffers in the fragment shader and so always runs. This works but it's effectively only checking 2 bindings, not 3. So, the question is, should I add 3rd buffer and change the algo to out = a - b - c etc.... so that we can shuffle more bindings? Or is this good enough? Or should I do something completely different. Also note: the last test 'compatible_pipelines' is unchagned and so only runs the comput pass unless the device supports storage buffers in fragment shaders. I didn't update it yet because for it to work requires either (a) two render passes to render to 2 different render targets. Or it needs some viewport settings to render to 2 different pixels in the same target. Or something..., all of which seem like the might require some big refactors. In the `createEncoder` infra in gpu_test.ts or else they'd just have to do their own thing entirely. Maybe that change doesn't need to happen in this PR but ideas are welcome. --- .../programmable/programmable_state_test.ts | 144 +++++++++++--- .../programmable/state_tracking.spec.ts | 188 ++++++++++++++---- src/webgpu/gpu_test.ts | 8 +- 3 files changed, 275 insertions(+), 65 deletions(-) diff --git a/src/webgpu/api/operation/command_buffer/programmable/programmable_state_test.ts b/src/webgpu/api/operation/command_buffer/programmable/programmable_state_test.ts index 19cf91419c16..a8222807b7bd 100644 --- a/src/webgpu/api/operation/command_buffer/programmable/programmable_state_test.ts +++ b/src/webgpu/api/operation/command_buffer/programmable/programmable_state_test.ts @@ -1,5 +1,5 @@ import { unreachable } from '../../../../../common/util/util.js'; -import { GPUTest } from '../../../../gpu_test.js'; +import { GPUTest, GPUTestBase } from '../../../../gpu_test.js'; import { EncoderType } from '../../../../util/command_buffer_maker.js'; interface BindGroupIndices { @@ -8,38 +8,81 @@ interface BindGroupIndices { out: number; } +type CreateEncoderType = ReturnType< + typeof GPUTestBase.prototype.createEncoder<'compute pass' | 'render pass' | 'render bundle'> +>['encoder']; + export class ProgrammableStateTest extends GPUTest { private commonBindGroupLayouts: Map = new Map(); - getBindGroupLayout(type: GPUBufferBindingType): GPUBindGroupLayout { - if (!this.commonBindGroupLayouts.has(type)) { + skipIfNeedsStorageBuffersInFragmentStageAndHaveNone( + type: GPUBufferBindingType, + encoderType: EncoderType + ) { + if (!this.isCompatibility) { + return; + } + + const needsStorageBuffersInFragmentStage = + type === 'storage' && (encoderType === 'render bundle' || encoderType === 'render pass'); + + this.skipIf( + needsStorageBuffersInFragmentStage && + !(this.device.limits.maxStorageBuffersInFragmentStage! >= 3), + `maxStorageBuffersInFragmentStage(${this.device.limits.maxStorageBuffersInFragmentStage}) < 3` + ); + } + + getBindGroupLayout( + type: GPUBufferBindingType, + visibility: GPUShaderStageFlags + ): GPUBindGroupLayout { + const id = `${type}:${visibility}`; + if (!this.commonBindGroupLayouts.has(id)) { this.commonBindGroupLayouts.set( - type, + id, this.device.createBindGroupLayout({ entries: [ { binding: 0, - visibility: GPUShaderStage.COMPUTE | GPUShaderStage.FRAGMENT, + visibility, buffer: { type }, }, ], }) ); } - return this.commonBindGroupLayouts.get(type)!; + return this.commonBindGroupLayouts.get(id)!; } - getBindGroupLayouts(indices: BindGroupIndices): GPUBindGroupLayout[] { + getVisibilityForEncoderType(encoderType: EncoderType) { + return encoderType === 'compute pass' ? GPUShaderStage.COMPUTE : GPUShaderStage.FRAGMENT; + } + + getBindGroupLayouts( + indices: BindGroupIndices, + type: GPUBufferBindingType, + encoderType: EncoderType + ): GPUBindGroupLayout[] { const bindGroupLayouts: GPUBindGroupLayout[] = []; - bindGroupLayouts[indices.a] = this.getBindGroupLayout('read-only-storage'); - bindGroupLayouts[indices.b] = this.getBindGroupLayout('read-only-storage'); - bindGroupLayouts[indices.out] = this.getBindGroupLayout('storage'); + const inputType = type === 'storage' ? 'read-only-storage' : 'uniform'; + const visibility = this.getVisibilityForEncoderType(encoderType); + bindGroupLayouts[indices.a] = this.getBindGroupLayout(inputType, visibility); + bindGroupLayouts[indices.b] = this.getBindGroupLayout(inputType, visibility); + if (type === 'storage' || encoderType === 'compute pass') { + bindGroupLayouts[indices.out] = this.getBindGroupLayout('storage', visibility); + } return bindGroupLayouts; } - createBindGroup(buffer: GPUBuffer, type: GPUBufferBindingType): GPUBindGroup { + createBindGroup( + buffer: GPUBuffer, + type: GPUBufferBindingType, + encoderType: EncoderType + ): GPUBindGroup { + const visibility = this.getVisibilityForEncoderType(encoderType); return this.device.createBindGroup({ - layout: this.getBindGroupLayout(type), + layout: this.getBindGroupLayout(type, visibility), entries: [{ binding: 0, resource: { buffer } }], }); } @@ -57,6 +100,7 @@ export class ProgrammableStateTest extends GPUTest { createBindingStatePipeline( encoderType: T, groups: BindGroupIndices, + type: GPUBufferBindingType, algorithm: string = 'a.value - b.value' ): GPUComputePipeline | GPURenderPipeline { switch (encoderType) { @@ -65,8 +109,8 @@ export class ProgrammableStateTest extends GPUTest { value : i32 }; - @group(${groups.a}) @binding(0) var a : Data; - @group(${groups.b}) @binding(0) var b : Data; + @group(${groups.a}) @binding(0) var<${type}> a : Data; + @group(${groups.b}) @binding(0) var<${type}> b : Data; @group(${groups.out}) @binding(0) var out : Data; @compute @workgroup_size(1) fn main() { @@ -77,7 +121,7 @@ export class ProgrammableStateTest extends GPUTest { return this.device.createComputePipeline({ layout: this.device.createPipelineLayout({ - bindGroupLayouts: this.getBindGroupLayouts(groups), + bindGroupLayouts: this.getBindGroupLayouts(groups, type, encoderType), }), compute: { module: this.device.createShaderModule({ @@ -92,7 +136,7 @@ export class ProgrammableStateTest extends GPUTest { const wgslShaders = { vertex: ` @vertex fn vert_main() -> @builtin(position) vec4 { - return vec4(0.5, 0.5, 0.0, 1.0); + return vec4(0, 0, 0, 1); } `, @@ -101,20 +145,23 @@ export class ProgrammableStateTest extends GPUTest { value : i32 }; - @group(${groups.a}) @binding(0) var a : Data; - @group(${groups.b}) @binding(0) var b : Data; + @group(${groups.a}) @binding(0) var<${type}> a : Data; + @group(${groups.b}) @binding(0) var<${type}> b : Data; @group(${groups.out}) @binding(0) var out : Data; - @fragment fn frag_main() -> @location(0) vec4 { + @fragment fn frag_main_storage() -> @location(0) vec4 { out.value = ${algorithm}; - return vec4(1.0, 0.0, 0.0, 1.0); + return vec4(1, 0, 0, 1); + } + @fragment fn frag_main_uniform() -> @location(0) vec4 { + return vec4(${algorithm}); } `, }; return this.device.createRenderPipeline({ layout: this.device.createPipelineLayout({ - bindGroupLayouts: this.getBindGroupLayouts(groups), + bindGroupLayouts: this.getBindGroupLayouts(groups, type, encoderType), }), vertex: { module: this.device.createShaderModule({ @@ -126,8 +173,8 @@ export class ProgrammableStateTest extends GPUTest { module: this.device.createShaderModule({ code: wgslShaders.fragment, }), - entryPoint: 'frag_main', - targets: [{ format: 'rgba8unorm' }], + entryPoint: type === 'uniform' ? 'frag_main_uniform' : 'frag_main_storage', + targets: [{ format: 'r32sint' }], }, primitive: { topology: 'point-list' }, }); @@ -137,6 +184,57 @@ export class ProgrammableStateTest extends GPUTest { } } + createEncoderForStateTest( + type: GPUBufferBindingType, + out: GPUBuffer, + ...params: Parameters + ): { + encoder: CreateEncoderType; + validateFinishAndSubmit: (shouldBeValid: boolean, submitShouldSucceedIfValid: boolean) => void; + } { + const encoderType = params[0]; + const renderTarget = this.createTextureTracked({ + size: [1, 1], + format: 'r32sint', + usage: GPUTextureUsage.RENDER_ATTACHMENT | GPUTextureUsage.COPY_SRC, + }); + + // Note: This nightmare of gibberish is trying the result of 2 hours of + // trying to get typescript to accept the code. Originally the code was + // effectively just + // + // const { encoder, validateFinishAndSubmit } = this.createEncoder(...); + // const fn = (b0, b1) => { validateFinishAndSubmit(b1, b1); if (...) { ... copyT2B ... } } + // return { encoder: e__, validateFinishAndSubmit: fn }; + // + // But TS didn't like it. I couldn't figure out why. + const encoderAndFinish = this.createEncoder(encoderType, { + attachmentInfo: { colorFormats: ['r32sint'] }, + targets: [renderTarget.createView()], + }); + + const validateFinishAndSubmit = ( + shouldBeValid: boolean, + submitShouldSucceedIfValid: boolean + ) => { + encoderAndFinish.validateFinishAndSubmit(shouldBeValid, submitShouldSucceedIfValid); + + if ( + type === 'uniform' && + (encoderType === 'render pass' || encoderType === 'render bundle') + ) { + const encoder = this.device.createCommandEncoder(); + encoder.copyTextureToBuffer({ texture: renderTarget }, { buffer: out }, [1, 1]); + this.device.queue.submit([encoder.finish()]); + } + }; + + return { + encoder: encoderAndFinish.encoder as CreateEncoderType, + validateFinishAndSubmit, + }; + } + setPipeline(pass: GPUBindingCommandsMixin, pipeline: GPUComputePipeline | GPURenderPipeline) { if (pass instanceof GPUComputePassEncoder) { pass.setPipeline(pipeline as GPUComputePipeline); diff --git a/src/webgpu/api/operation/command_buffer/programmable/state_tracking.spec.ts b/src/webgpu/api/operation/command_buffer/programmable/state_tracking.spec.ts index fe8ef3d4374f..3dd8b9f5392f 100644 --- a/src/webgpu/api/operation/command_buffer/programmable/state_tracking.spec.ts +++ b/src/webgpu/api/operation/command_buffer/programmable/state_tracking.spec.ts @@ -5,13 +5,18 @@ times in different orders) for setBindGroup and setPipeline. import { makeTestGroup } from '../../../../../common/framework/test_group.js'; import { GPUConst } from '../../../../constants.js'; +import { MaxLimitsTestMixin } from '../../../../gpu_test.js'; import { kProgrammableEncoderTypes } from '../../../../util/command_buffer_maker.js'; import { ProgrammableStateTest } from './programmable_state_test.js'; -export const g = makeTestGroup(ProgrammableStateTest); +export const g = makeTestGroup(MaxLimitsTestMixin(ProgrammableStateTest)); -const kBufferUsage = GPUConst.BufferUsage.COPY_SRC | GPUConst.BufferUsage.STORAGE; +const kBufferUsage = + GPUConst.BufferUsage.COPY_SRC | + GPUConst.BufferUsage.COPY_DST | + GPUConst.BufferUsage.STORAGE | + GPUConst.BufferUsage.UNIFORM; g.test('bind_group_indices') .desc( @@ -24,6 +29,7 @@ g.test('bind_group_indices') u // .combine('encoderType', kProgrammableEncoderTypes) .beginSubcases() + .combine('type', ['storage', 'uniform'] as const) .combine('groupIndices', [ { a: 0, b: 1, out: 2 }, { a: 1, b: 2, out: 0 }, @@ -34,24 +40,40 @@ g.test('bind_group_indices') ]) ) .fn(t => { - const { encoderType, groupIndices } = t.params; + const { encoderType, groupIndices, type } = t.params; + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone(type, encoderType); - const pipeline = t.createBindingStatePipeline(encoderType, groupIndices); + const pipeline = t.createBindingStatePipeline( + encoderType, + groupIndices, + type, + 'a.value - b.value' + ); + const inputType: GPUBufferBindingType = type === 'storage' ? 'read-only-storage' : 'uniform'; const out = t.makeBufferWithContents(new Int32Array([0]), kBufferUsage); const bindGroups = { a: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), b: t.createBindGroup( t.makeBufferWithContents(new Int32Array([2]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), - out: t.createBindGroup(out, 'storage'), + out: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(out, 'storage', encoderType) + : null, }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoderForStateTest( + type, + out, + encoderType + ); t.setPipeline(encoder, pipeline); encoder.setBindGroup(groupIndices.a, bindGroups.a); @@ -73,6 +95,7 @@ g.test('bind_group_order') u // .combine('encoderType', kProgrammableEncoderTypes) .beginSubcases() + .combine('type', ['storage', 'uniform'] as const) .combine('setOrder', [ ['a', 'b', 'out'], ['b', 'out', 'a'], @@ -83,25 +106,41 @@ g.test('bind_group_order') ] as const) ) .fn(t => { - const { encoderType, setOrder } = t.params; + const { encoderType, setOrder, type } = t.params; + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone(type, encoderType); const groupIndices = { a: 0, b: 1, out: 2 }; - const pipeline = t.createBindingStatePipeline(encoderType, groupIndices); + const pipeline = t.createBindingStatePipeline( + encoderType, + groupIndices, + type, + 'a.value - b.value' + ); const out = t.makeBufferWithContents(new Int32Array([0]), kBufferUsage); + const inputType: GPUBufferBindingType = type === 'storage' ? 'read-only-storage' : 'uniform'; const bindGroups = { a: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), b: t.createBindGroup( t.makeBufferWithContents(new Int32Array([2]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), - out: t.createBindGroup(out, 'storage'), + out: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(out, 'storage', encoderType) + : null, }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoderForStateTest( + type, + out, + encoderType + ); t.setPipeline(encoder, pipeline); for (const bindingName of setOrder) { @@ -124,6 +163,7 @@ g.test('bind_group_before_pipeline') u // .combine('encoderType', kProgrammableEncoderTypes) .beginSubcases() + .combine('type', ['storage', 'uniform'] as const) .combineWithParams([ { setBefore: ['a', 'b'], setAfter: ['out'] }, { setBefore: ['a'], setAfter: ['b', 'out'] }, @@ -132,24 +172,41 @@ g.test('bind_group_before_pipeline') ] as const) ) .fn(t => { - const { encoderType, setBefore, setAfter } = t.params; + const { encoderType, type, setBefore, setAfter } = t.params; + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone(type, encoderType); + const groupIndices = { a: 0, b: 1, out: 2 }; - const pipeline = t.createBindingStatePipeline(encoderType, groupIndices); + const pipeline = t.createBindingStatePipeline( + encoderType, + groupIndices, + type, + 'a.value - b.value' + ); const out = t.makeBufferWithContents(new Int32Array([0]), kBufferUsage); + const inputType: GPUBufferBindingType = type === 'storage' ? 'read-only-storage' : 'uniform'; const bindGroups = { a: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), b: t.createBindGroup( t.makeBufferWithContents(new Int32Array([2]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), - out: t.createBindGroup(out, 'storage'), + out: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(out, 'storage', encoderType) + : null, }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoderForStateTest( + type, + out, + encoderType + ); for (const bindingName of setBefore) { encoder.setBindGroup(groupIndices[bindingName], bindGroups[bindingName]); @@ -176,21 +233,39 @@ g.test('one_bind_group_multiple_slots') .params(u => u // .combine('encoderType', kProgrammableEncoderTypes) + .beginSubcases() + .combine('type', ['storage', 'uniform'] as const) ) .fn(t => { - const { encoderType } = t.params; - const pipeline = t.createBindingStatePipeline(encoderType, { a: 0, b: 1, out: 2 }); + const { encoderType, type } = t.params; + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone(type, encoderType); + + const pipeline = t.createBindingStatePipeline( + encoderType, + { a: 0, b: 1, out: 2 }, + type, + 'a.value - b.value' + ); const out = t.makeBufferWithContents(new Int32Array([1]), kBufferUsage); + const inputType: GPUBufferBindingType = type === 'storage' ? 'read-only-storage' : 'uniform'; const bindGroups = { ab: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), - out: t.createBindGroup(out, 'storage'), + out: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(out, 'storage', encoderType) + : null, }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoderForStateTest( + type, + out, + encoderType + ); t.setPipeline(encoder, pipeline); encoder.setBindGroup(0, bindGroups.ab); @@ -212,31 +287,54 @@ g.test('bind_group_multiple_sets') .params(u => u // .combine('encoderType', kProgrammableEncoderTypes) + .beginSubcases() + .combine('type', ['storage', 'uniform'] as const) ) .fn(t => { - const { encoderType } = t.params; - const pipeline = t.createBindingStatePipeline(encoderType, { a: 0, b: 1, out: 2 }); + const { encoderType, type } = t.params; + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone(type, encoderType); + + const pipeline = t.createBindingStatePipeline( + encoderType, + { a: 0, b: 1, out: 2 }, + type, + 'a.value - b.value' + ); const badOut = t.makeBufferWithContents(new Int32Array([-1]), kBufferUsage); const out = t.makeBufferWithContents(new Int32Array([0]), kBufferUsage); + const inputType: GPUBufferBindingType = type === 'storage' ? 'read-only-storage' : 'uniform'; const bindGroups = { a: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), b: t.createBindGroup( t.makeBufferWithContents(new Int32Array([2]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), c: t.createBindGroup( t.makeBufferWithContents(new Int32Array([5]), kBufferUsage), - 'read-only-storage' + inputType, + encoderType ), - badOut: t.createBindGroup(badOut, 'storage'), - out: t.createBindGroup(out, 'storage'), + badOut: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(badOut, 'storage', encoderType) + : null, + out: + encoderType === 'compute pass' || type === 'storage' + ? t.createBindGroup(out, 'storage', encoderType) + : null, }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoderForStateTest( + type, + out, + encoderType + ); encoder.setBindGroup(1, bindGroups.c); @@ -265,10 +363,18 @@ g.test('compatible_pipelines') ) .fn(t => { const { encoderType } = t.params; - const pipelineA = t.createBindingStatePipeline(encoderType, { a: 0, b: 1, out: 2 }); + t.skipIfNeedsStorageBuffersInFragmentStageAndHaveNone('storage', encoderType); + + const pipelineA = t.createBindingStatePipeline( + encoderType, + { a: 0, b: 1, out: 2 }, + 'storage', + 'a.value - b.value' + ); const pipelineB = t.createBindingStatePipeline( encoderType, { a: 0, b: 1, out: 2 }, + 'storage', 'a.value + b.value' ); @@ -277,17 +383,21 @@ g.test('compatible_pipelines') const bindGroups = { a: t.createBindGroup( t.makeBufferWithContents(new Int32Array([3]), kBufferUsage), - 'read-only-storage' + 'read-only-storage', + encoderType ), b: t.createBindGroup( t.makeBufferWithContents(new Int32Array([2]), kBufferUsage), - 'read-only-storage' + 'read-only-storage', + encoderType ), - outA: t.createBindGroup(outA, 'storage'), - outB: t.createBindGroup(outB, 'storage'), + outA: t.createBindGroup(outA, 'storage', encoderType), + outB: t.createBindGroup(outB, 'storage', encoderType), }; - const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType); + const { encoder, validateFinishAndSubmit } = t.createEncoder(encoderType, { + attachmentInfo: { colorFormats: ['r32sint'] }, + }); encoder.setBindGroup(0, bindGroups.a); encoder.setBindGroup(1, bindGroups.b); diff --git a/src/webgpu/gpu_test.ts b/src/webgpu/gpu_test.ts index 1e7fac4a984e..9c63f5d15d2c 100644 --- a/src/webgpu/gpu_test.ts +++ b/src/webgpu/gpu_test.ts @@ -1131,9 +1131,11 @@ export class GPUTestBase extends Fixture { { attachmentInfo, occlusionQuerySet, + targets, }: { attachmentInfo?: GPURenderBundleEncoderDescriptor; occlusionQuerySet?: GPUQuerySet; + targets?: GPUTextureView[]; } = {} ): CommandBufferMaker { const fullAttachmentInfo = { @@ -1155,7 +1157,7 @@ export class GPUTestBase extends Fixture { case 'render bundle': { const device = this.device; const rbEncoder = device.createRenderBundleEncoder(fullAttachmentInfo); - const pass = this.createEncoder('render pass', { attachmentInfo }); + const pass = this.createEncoder('render pass', { attachmentInfo, targets }); return new CommandBufferMaker(this, rbEncoder, () => { pass.encoder.executeBundles([rbEncoder.finish()]); @@ -1205,10 +1207,10 @@ export class GPUTestBase extends Fixture { } } const passDesc: GPURenderPassDescriptor = { - colorAttachments: Array.from(fullAttachmentInfo.colorFormats, format => + colorAttachments: Array.from(fullAttachmentInfo.colorFormats, (format, i) => format ? { - view: makeAttachmentView(format), + view: targets ? targets[i] : makeAttachmentView(format), clearValue: [0, 0, 0, 0], loadOp: 'clear', storeOp: 'store',