in src/webgpu/shader/execution/flow_control/harness.ts [128:232]
stack: Error().stack,
});
// Expectation id starts from 1 to distinguish from initialization 0.
return `push_output(${expectations.length}); // expect_not_reached()`;
},
});
const built_wgsl =
typeof build_wgsl_result === 'string'
? { entrypoint: build_wgsl_result, extra: '' }
: build_wgsl_result;
const main_wgsl = built_wgsl.entrypoint !== undefined ? built_wgsl : built_wgsl.entrypoint;
const wgsl = `
struct Outputs {
count : u32,
data : array<u32>,
};
@group(0) @binding(0) var<storage, read> inputs : array<i32>;
@group(0) @binding(1) var<storage, read_write> outputs : Outputs;
fn push_output(value : u32) {
outputs.data[outputs.count] = value;
outputs.count++;
}
@compute @workgroup_size(1)
fn main() {
_ = &inputs;
_ = &outputs;
${main_wgsl.entrypoint}
}
${main_wgsl.extra}
`;
const pipeline = t.device.createComputePipeline({
layout: 'auto',
compute: {
module: t.device.createShaderModule({ code: wgsl }),
entryPoint: 'main',
},
});
// If there are no inputs, just put a single value in the buffer to keep
// makeBufferWithContents() happy.
if (inputData.length === 0) {
inputData.push(0);
}
const inputBuffer = t.makeBufferWithContents(new Uint32Array(inputData), GPUBufferUsage.STORAGE);
const maxOutputValues = 1000;
const outputBuffer = t.createBufferTracked({
size: 4 * (1 + maxOutputValues),
usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_SRC,
});
const bindGroup = t.device.createBindGroup({
layout: pipeline.getBindGroupLayout(0),
entries: [
{ binding: 0, resource: { buffer: inputBuffer } },
{ binding: 1, resource: { buffer: outputBuffer } },
],
});
// Run the shader.
const encoder = t.device.createCommandEncoder({ label: 'runFlowControlTest' });
const pass = encoder.beginComputePass();
pass.setPipeline(pipeline);
pass.setBindGroup(0, bindGroup);
pass.dispatchWorkgroups(1);
pass.end();
t.queue.submit([encoder.finish()]);
t.eventualExpectOK(
t
.readGPUBufferRangeTyped(outputBuffer, {
type: Uint32Array,
typedLength: outputBuffer.size / 4,
})
.then(outputs => {
// outputs[0] is the number of outputted values
// outputs[1..N] holds the outputted values
const outputCount = outputs.data[0];
if (outputCount > maxOutputValues) {
return new Error(
`output data count (${outputCount}) exceeds limit of ${maxOutputValues}`
);
}
// returns an Error with the given message and WGSL source
const fail = (err: string) => Error(`${err}\nWGSL:\n${Colors.dim(Colors.blue(wgsl))}`);
// returns a string that shows the outputted values to help understand the whole trace.
const print_output_value = () => {
const subarray = outputs.data.subarray(1, outputCount + 1);
return `Output values (length: ${outputCount}): ${subarray.join(', ')}`;
};
// returns a colorized string of the expect_order() call, highlighting
// the event number that caused an error.
const expect_order_err = (expectation: ExpectedEvents, err_idx: number) => {
let out = 'expect_order(';
for (let i = 0; i < expectation.values.length; i++) {