blob: fcacf92d72d753558a329165fe94128fa6d80213 [file] [log] [blame]
<!DOCTYPE html><!-- webkit-test-runner [ experimental:WebGPUEnabled=true ] -->
<meta charset=utf-8>
<title>Execute a simple compute shader with Web GPU.</title>
<body>
<script src="../resources/testharness.js"></script>
<script src="../resources/testharnessreport.js"></script>
<script src="js/webgpu-functions.js"></script>
<script>
if (window.testRunner)
testRunner.waitUntilDone();
const data = new Uint32Array([2, 3, 4, 5, 6, 7, 8, 9, 10]);
const dataBinding = 0;
const bindGroupIndex = 0;
const shaderCode = `
#include <metal_stdlib>
struct Data {
device unsigned* numbers [[id(${dataBinding})]];
};
kernel void compute(device Data& data [[buffer(${bindGroupIndex})]], unsigned gid [[thread_position_in_grid]])
{
if (gid >= ${data.length})
return;
unsigned original = data.numbers[gid];
data.numbers[gid] = original * original;
}
`
promise_test(async () => {
const device = await getBasicDevice();
const shaderModule = device.createShaderModule({ code: shaderCode, isWHLSL: false });
const computeStageDescriptor = { module: shaderModule, entryPoint: "compute" };
const pipeline = device.createComputePipeline({ computeStage: computeStageDescriptor });
const dataBuffer = device.createBuffer({ size: data.byteLength, usage: GPUBufferUsage.STORAGE | GPUBufferUsage.TRANSFER_DST | GPUBufferUsage.MAP_READ });
dataBuffer.setSubData(0, data.buffer);
const bgLayoutBinding = { binding: dataBinding, visibility: GPUShaderStageBit.COMPUTE, type: "storage-buffer" };
const bgLayout = device.createBindGroupLayout({ bindings: [bgLayoutBinding] });
const bufferBinding = { buffer: dataBuffer, size: data.byteLength };
const bgBinding = { binding: dataBinding, resource: bufferBinding };
const bindGroupDescriptor = { layout: bgLayout, bindings: [bgBinding] };
const bindGroup = device.createBindGroup(bindGroupDescriptor);
const commandEncoder = device.createCommandEncoder();
const passEncoder = commandEncoder.beginComputePass();
passEncoder.setBindGroup(bindGroupIndex, bindGroup);
passEncoder.setPipeline(pipeline);
// One thread group.
passEncoder.dispatch(1, 1, 1);
passEncoder.endPass();
device.getQueue().submit([commandEncoder.finish()]);
const readDataArrayBuffer = await dataBuffer.mapReadAsync();
assert_not_equals(readDataArrayBuffer, null, "Async read promise resolved successfully");
const readData = new Uint32Array(readDataArrayBuffer);
for (var i = 0; i < readData.length; ++i)
assert_equals(readData[i], data[i] * data[i], "Data was succesfully squared");
if (window.testRunner)
testRunner.notifyDone();
}, "Successfully executed a basic compute pass");
</script>
</body>