WebGPU 的新变化 (Chrome 123)

François Beaufort

WGSL 中对 DP4a 内置函数的支持

DP4a（4 个元素的点积和累加）是指一组 GPU 指令，通常用于深度学习推理以进行量化。它可以高效地执行 8 位整数点积，以加快此类 int8 量化模型的计算速度。与 f32 版本相比，它可以节省（高达 75%）内存和网络带宽，并提高任何机器学习模型在推理方面的性能。因此，它现在被许多热门 AI 框架广泛使用。

当 "packed_4x8_integer_dot_product" WGSL 语言扩展在 navigator.gpu.wgslLanguageFeatures 中存在时，您现在可以使用 32 位整数标量将 4 分量向量的 8 位整数打包为 WGSL 着色器代码中点积指令的输入，并使用 dot4U8Packed 和 dot4I8Packed 内置函数。您还可以使用打包和解包指令以及打包的 4 分量向量的 8 位整数，并使用 pack4xI8、pack4xU8、pack4xI8Clamp、pack4xU8Clamp、unpack4xI8 和 unpack4xU8 WGSL 内置函数。

建议使用 requires 指令来表明可能存在不可移植性，方法是在 WGSL 着色器代码的顶部使用 requires packed_4x8_integer_dot_product;。请参阅以下示例和问题 tint:1497。

if (!navigator.gpu.wgslLanguageFeatures.has("packed_4x8_integer_dot_product")) {
  throw new Error(`DP4a built-in functions are not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires packed_4x8_integer_dot_product;

  fn main() {
    const result: u32 = dot4U8Packed(0x01020304u, 0x02040405u); // 42
  }`,
});

特别感谢英特尔上海 Web 图形团队推动此规范和实现工作顺利完成！

WGSL 中的无限制指针形参

通过 "unrestricted_pointer_parameters" WGSL 语言扩展，可以放宽对可传递给 WGSL 函数的指针的限制：

storage、uniform 和 workgroup 地址空间的用户声明函数的形参指针。
将指向结构成员和数组元素的指针传递给用户声明的函数。

如需了解详情，请查看 Pointers As Function Parameters | Tour of WGSL。

您可以使用 navigator.gpu.wgslLanguageFeatures 检测此功能。建议始终使用 requires 指令来表明可能存在不可移植性，方法是在 WGSL 着色器代码的顶部使用 requires unrestricted_pointer_parameters;。请参阅以下示例、WGSL 规范更改和问题 tint:2053。

if (!navigator.gpu.wgslLanguageFeatures.has("unrestricted_pointer_parameters")) {
  throw new Error(`Unrestricted pointer parameters are not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires unrestricted_pointer_parameters;

  @group(0) @binding(0) var<storage, read_write> S : i32;

  fn func(pointer : ptr<storage, i32, read_write>) {
    *pointer = 42;
  }

  @compute @workgroup_size(1)
  fn main() {
    func(&S);
  }`
});

WGSL 中用于取消引用复合类型的语法糖

当 navigator.gpu.wgslLanguageFeatures 中存在 "pointer_composite_access" WGSL 语言扩展时，无论您是直接处理数据还是处理指向数据的指针，您的 WGSL 着色器代码现在都支持使用相同的点 (.) 语法访问复杂数据类型的组件。其工作原理如下：

如果 foo 是一个指针，则 foo.bar 是一种更方便的 (*foo).bar 编写方式。通常需要使用星号 (*) 将指针转换为可以取消引用的“引用”，但现在指针和引用非常相似，几乎可以互换。
如果 foo 不是指针，则点 (.) 运算符的工作方式与您习惯的直接访问成员的方式完全相同。

同样，如果 pa 是一个指针，用于存储数组的起始地址，那么使用 pa[i] 可以直接访问存储该数组的第 'i' 个元素的内存位置。

建议使用 requires 指令来表明可能存在不可移植性，方法是在 WGSL 着色器代码的顶部使用 requires pointer_composite_access;。请参阅以下示例和问题 tint:2113。

if (!navigator.gpu.wgslLanguageFeatures.has("pointer_composite_access")) {
  throw new Error(`Pointer composite access is not available`);
}

const adapter = await navigator.gpu.requestAdapter();
const device = await adapter.requestDevice();

const shaderModule = device.createShaderModule({ code: `
  requires pointer_composite_access;

  fn main() {
    var a = vec3f();
    let p : ptr<function, vec3f> = &a;
    let r1 = (*p).x; // always valid.
    let r2 = p.x; // requires pointer composite access.
  }`
});

为模板和深度方面单独设置只读状态

以前，渲染通道中的只读深度模板附件要求深度和模板这两个方面都为只读。此限制已解除。现在，您可以采用只读方式使用深度方面（例如，用于接触阴影跟踪），同时写入模板缓冲区以识别像素以供进一步处理。请参阅问题 dawn:2146。

Dawn 更新

现在，当发生错误时，系统会立即调用使用 wgpuDeviceSetUncapturedErrorCallback() 设置的未捕获错误回调。这是开发者一直期望并希望用于调试的功能。请参阅更改 dawn:173620。

已实现 webgpu.h API 中的 wgpuSurfaceGetPreferredFormat() 方法。请参阅问题 dawn:1362。

这仅涵盖了一些关键亮点。请查看提交的完整列表。