(limits …)
By default a payload runs against WebGPU’s default device limits. Some
shaders need more, most commonly a compute entry whose @workgroup_size
exceeds the default maxComputeWorkgroupSizeX /
maxComputeInvocationsPerWorkgroup (both 256).
Declare the raised limits with a single top-level (limits …) form. Each key is
a WebGPU GPUSupportedLimits member in kebab-case, taking a non-negative
integer: a literal, an expression, or a bare (define …) name.
Syntax
Section titled “Syntax”(limits :max-compute-workgroup-size-x 512 :max-compute-invocations-per-workgroup 512)
(shader-module :name code :code """@compute @workgroup_size(512, 1, 1) fn main() { /* … */ }""")Without the form the compiler rejects @workgroup_size(512): compute entry
‘main’ @workgroup_size(512, 1, 1) exceeds WebGPU limits (x exceeds
maxComputeWorkgroupSizeX). That is the same failure you would otherwise meet
only as an unexplained pipeline-creation error at load time.
It is an optional singleton: declare it at most once, and it takes no :name.
Requirements, not hints
Section titled “Requirements, not hints”Authored limits are passed verbatim to requestDevice’s requiredLimits,
on the browser viewer (pngine, the full worker-based runtime), the --html
export and the native --frame renderer alike. Unlike device features, which
the runtime requests opportunistically, an unsatisfiable limit fails
loudly: requestDevice rejects (surfaced through the worker error channel,
or a nonzero exit natively). It is never silently clamped to what the adapter
offers.
A few newer limits (for instance :max-bind-groups-plus-vertex-buffers) have
uneven browser support and will reject on a device that lacks them.
The schema names every GPUSupportedLimits member, and nothing else: a key it
does not know is an unknown_key diagnostic rather than a limit the device
never hears about. The ones that come up most:
| Key | WebGPU default |
|---|---|
:max-compute-workgroup-size-x |
256 |
:max-compute-workgroup-size-y |
256 |
:max-compute-workgroup-size-z |
64 |
:max-compute-invocations-per-workgroup |
256 |
:max-compute-workgroup-storage-size |
16384 |
:max-compute-workgroups-per-dimension |
65535 |
:max-buffer-size |
268435456 |
:max-storage-buffer-binding-size |
134217728 |
:max-uniform-buffer-binding-size |
65536 |
:max-texture-dimension-2d |
8192 |
:max-texture-array-layers |
256 |
:max-bind-groups |
4 |
:max-vertex-buffers |
8 |
:max-vertex-attributes |
16 |
:max-color-attachments |
8 |
:max-storage-buffers-per-shader-stage |
8 |
:max-sampled-textures-per-shader-stage |
16 |
:max-samplers-per-shader-stage |
16 |
The remaining members follow the same kebab-case rule: maxTextureDimension3D
is :max-texture-dimension-3d, minUniformBufferOffsetAlignment is
:min-uniform-buffer-offset-alignment, and so on. The compatibility split-stage
limits are there too, :max-storage-buffers-in-vertex-stage and its three
siblings, as is :max-immediate-size.
Example
Section titled “Example”A compute pass with a 512-invocation workgroup:
(limits :max-compute-workgroup-size-x 512 :max-compute-invocations-per-workgroup 512)
(shader-module :name workCode :code """@group(0) @binding(0) var<storage, read_write> colour: array<f32>;@compute @workgroup_size(512, 1, 1)fn writeColour(@builtin(local_invocation_index) idx: u32) { if (idx == 511u) { colour[0] = 0.9; colour[1] = 0.4; colour[2] = 0.1; colour[3] = 1.0; }}""")
(buffer :name colourBuf :size 16 :usage [storage])
(compute-pipeline :name workPipe :layout auto (compute :module workCode :entry writeColour))
(bind-group :name workGroup :layout workPipe :group 0 (entry :binding 0 :buffer colourBuf))
(compute-pass :name work :pipeline workPipe :bind-groups [workGroup] (dispatch :workgroups [1]))
(frame :name main :perform [work])Validation Rules
Section titled “Validation Rules”| Rule | Error |
|---|---|
Declare at most one (limits …) per document |
duplicate (limits …) form, declare device limits at most once |
The key must be a GPUSupportedLimits member |
unknown_key |
| Each value must be a non-negative integer | number_below_min |
A @workgroup_size above the effective limit is rejected at compile time |
compute entry ‘main’ @workgroup_size(512, 1, 1) exceeds WebGPU limits |
Platform notes
Section titled “Platform notes”The mini player (pngine/mini, the small main-thread runtime for --flat
payloads) refuses a limits-bearing payload at write time; I keep it requesting
a bare device. Use the standard viewer runtime, or the --html export, for
payloads that raise limits.
Where these words come from
Section titled “Where these words come from”Every key and value on this page traced to the WebGPU name it stands for, with a link to the definition. How the tracing is made, and what keeps it from rotting, is the subject of Where the Words Come From.
(limits …) mirrors GPUSupportedLimits (MDN). Each key is one attribute of GPUSupportedLimits, the limit the runtime requests from the adapter at device creation. The engine holds this key set equal to the interface, exactly.
Checked against the WebGPU specification at revision b8c0fa9; the links go to the current draft.
Related
Section titled “Related”(compute-pipeline …)- Where large workgroups appear(canvas …)- The other device-configuration singleton(buffer …)- Sizes bounded by the buffer limits above