WebGPU 的新变化 (Chrome 123)

François Beaufort
François Beaufort
GitHub

WGSL 中的 DP4a 内置函数支持

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

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

建议使用 requires-directive，通过 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
  }`,
});

特别感谢 Intel 位于上海的 Web 图形团队推动该规范和实现顺利完成！

WGSL 中的无限制指针参数

"unrestricted_pointer_parameters" WGSL 语言扩展放宽了对可传递到 WGSL 函数的指针的限制：

  • storage、uniform 和 workgroup 地址空间的参数指针，指向用户声明的函数。

  • 将指向结构成员和数组元素的指针传递给用户声明的函数。

请参阅指针作为函数参数 |WGSL 导览，了解详情。

您可以使用 navigator.gpu.wgslLanguageFeatures 对此特征进行特征检测。建议始终使用 requires-directive，通过 WGSL 着色器代码顶部的 requires unrestricted_pointer_parameters; 发出不可移植性的信号。请参阅以下示例、WGSL 规范变更和 issue 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-directive，通过 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() 设置的未捕获错误回调。这是开发者一直以来对调试的期望和期望。请参阅 change dawn:173620。

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

本指南仅涵盖部分重要内容。查看详尽的提交内容列表。

WebGPU 的新变化

WebGPU 新变化系列涵盖的所有内容的列表。

Chrome 127

Chrome 126

Chrome 125

Chrome 124

Chrome 123

Chrome 122

Chrome 121

Chrome 120

Chrome 119

Chrome 118

Chrome 117

Chrome 116

Chrome 115

Chrome 114

Chrome 113