此页面由 Cloud Translation API 翻译。

WebGPU 的新变化 (Chrome 123)

François Beaufort

WGSL 中的 DP4a 内置函数支持

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

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

建议在 WGSL 着色器代码的顶部使用 requires-directive 来表明 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 检测此功能。建议始终在 WGSL 着色器代码的顶部使用 requires-directive 来表明 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 个元素的内存位置。

建议在 WGSL 着色器代码的顶部使用 requires-directive 来表明 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 更新