(...args: Array<GPUPointer | number>)
| 714 | // Function to create the pipeline. |
| 715 | const createShaderFunc = (pipeline: GPUComputePipeline): Function => { |
| 716 | const submitShader = (...args: Array<GPUPointer | number>): void => { |
| 717 | if (this.debugShaderSubmitLimit != -1 && |
| 718 | this.shaderSubmitCounter >= this.debugShaderSubmitLimit) { |
| 719 | this.shaderSubmitCounter += 1; |
| 720 | return; |
| 721 | } |
| 722 | |
| 723 | // Reuse a single command encoder across dispatches; only flush on sync/readback. |
| 724 | if (!this.pendingEncoder) { |
| 725 | this.pendingEncoder = this.device.createCommandEncoder(); |
| 726 | } |
| 727 | |
| 728 | const compute = this.pendingEncoder.beginComputePass(); |
| 729 | compute.setPipeline(pipeline); |
| 730 | const bindGroupEntries: Array<GPUBindGroupEntry> = []; |
| 731 | const numBufferOrPodArgs = bufferArgIndices.length + podArgIndices.length; |
| 732 | |
| 733 | assert(args.length == numBufferOrPodArgs + dispatchToDim.length); |
| 734 | |
| 735 | const workDim: Array<number> = [1, 1, 1, 1, 1, 1]; |
| 736 | for (let i = 0; i < dispatchToDim.length; ++i) { |
| 737 | workDim[dispatchToDim[i]] = args[numBufferOrPodArgs + i]; |
| 738 | } |
| 739 | |
| 740 | // get around 65535 restriction of blockIdx.x |
| 741 | if (workDim[2] != 1) { |
| 742 | throw Error("WebGPU: blockIdx.z is reserved for internal use"); |
| 743 | } |
| 744 | const packDimX = workDim[0]; |
| 745 | // spread thinsg out into blockIdx.z |
| 746 | if (workDim[0] >= (1 << 16)) { |
| 747 | let wl_x = workDim[0]; |
| 748 | let wl_z = workDim[2]; |
| 749 | |
| 750 | while (wl_x >= (1 << 16)) { |
| 751 | if (wl_x % 2 == 0) { |
| 752 | wl_x = wl_x / 2; |
| 753 | } else { |
| 754 | // pad up |
| 755 | wl_x = (wl_x + 1) / 2; |
| 756 | } |
| 757 | wl_z *= 2; |
| 758 | } |
| 759 | workDim[0] = wl_x; |
| 760 | workDim[2] = wl_z; |
| 761 | assert(wl_x * wl_z >= packDimX); |
| 762 | } |
| 763 | |
| 764 | for (let i = 0; i < bufferArgIndices.length; ++i) { |
| 765 | bindGroupEntries.push({ |
| 766 | binding: i, |
| 767 | resource: { |
| 768 | buffer: this.gpuBufferFromPtr(args[bufferArgIndices[i]]) |
| 769 | } |
| 770 | }); |
| 771 | } |
| 772 | |
| 773 | const sizeOfI32 = 4; |
nothing calls this directly
no test coverage detected