blob: c6ecc772e9862528d7480933f3bc7bab10f6bf2b [file]
<script src="../../../resources/js-test-pre.js"></script>
<script>
var output;
const BUFFER_SIZE = 0x4000;
const SHADER = `
struct OOBStruct {
field0: array<u32, 0x10000000>,
field1: array<u32, 0x10000000>,
field2: array<u32, 0x10000000>,
field3: array<u32, 0x10000000>,
field4: array<u32, 0x1000>,
}
@group(0) @binding(0) var<storage, read_write> g_oob_struct: OOBStruct;
@compute @workgroup_size(1)
fn c() {
g_oob_struct.field0[0x1000] = 1234;
g_oob_struct.field0[0x2000] = 1234;
g_oob_struct.field0[0x3000] = 1234;
g_oob_struct.field0[0x4000] = 1234;
}
`;
async function main() {
const adapter = await navigator.gpu.requestAdapter({});
const device = await adapter.requestDevice({});
device.pushErrorScope('validation');
const shaderModule = device.createShaderModule({code: SHADER});
const bindGroupLayout = device.createBindGroupLayout({
entries: [
{binding: 0, buffer: {type: 'storage'}, visibility: GPUShaderStage.COMPUTE},
]
});
const oobBuffer = device.createBuffer({
usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_SRC,
size: BUFFER_SIZE
});
const overflownBuffer = device.createBuffer({
usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_SRC,
size: BUFFER_SIZE
});
const outputBuffer = device.createBuffer({
usage: GPUBufferUsage.COPY_DST | GPUBufferUsage.MAP_READ,
size: BUFFER_SIZE
});
const bindGroup = device.createBindGroup({
layout: bindGroupLayout, entries: [
{binding: 0, resource: {buffer: oobBuffer}},
],
});
const commandEncoder = device.createCommandEncoder();
const computePassEncoder = commandEncoder.beginComputePass({});
const pipelineLayout = device.createPipelineLayout({bindGroupLayouts: [bindGroupLayout]});
const computePipeline = device.createComputePipeline({layout: pipelineLayout, compute: {module: shaderModule}});
computePassEncoder.setPipeline(computePipeline);
computePassEncoder.setBindGroup(0, bindGroup);
computePassEncoder.dispatchWorkgroups(1);
computePassEncoder.end();
commandEncoder.copyBufferToBuffer(overflownBuffer, 0, outputBuffer, 0, BUFFER_SIZE);
device.queue.submit([commandEncoder.finish()]);
await device.queue.onSubmittedWorkDone();
await outputBuffer.mapAsync(GPUMapMode.READ);
output = new Uint32Array(outputBuffer.getMappedRange());
debug(output);
shouldBe('output[0]', '0');
outputBuffer.unmap();
}
globalThis.testRunner?.waitUntilDone();
main().catch(e => {
debug(e);
}).finally(() => {
globalThis.testRunner?.notifyDone();
});
</script>