wgsl
npx skills add https://github.com/juzi5201314/cruft --skill wgsl
Agent 安装分布
Skill 文档
WGSL Skillï¼WebGPU / wgpu / Bevy éç¨ï¼
æ¬ skill çç®æ æ¯ï¼æâéæ±æè¿°â稳å®è½¬æå¯ç¼è¯ã坿¥å ¥ç WGSLï¼å¹¶æç®¡çº¿å¯¹æ¥ä¿¡æ¯ï¼bindings / locations / entry pointsï¼ä¸å¹¶è¾åºï¼åå°æ¥åè¯éã
ð¯ 硬约æï¼å¿ é¡»éµå®ï¼
- åªåç°ä»£ WGSLï¼ä½¿ç¨
@vertex/@fragment/@computeä¸@location/@builtinçæ°è¯æ³ï¼ä¸å任使§ç/å ¼å®¹åæ³ã - WGSL ä¸¥æ ¼ç±»åï¼æ¯ä¸ªåé/åæ°/è¿åå¼å¿
须坿¨å¯¼ææ¾å¼æ 注ï¼éè¦æ¶å¿
é¡»æ¾å¼è½¬æ¢ï¼ä¾å¦
f32(i32_value)ï¼ã - é»è®¤ä¸ä½¿ç¨
f16ï¼å®æ¯å¯éç¹æ§ï¼ï¼é¤éç¨æ·æç¡®è¦æ±å¹¶è¯´æå·²å¯ç¨å¯¹åº featureã - ä»£ç æ è¯ç¬¦ç¨è±æï¼è§£éç¨ä¸æï¼å¯æ··ç¨è±ææ¯è¯ï¼ã
ð§ ä½¿ç¨æµç¨ï¼æ¯æ¬¡å WGSL 齿è¿ä¸ªèµ°ï¼
- æç¡®ç®æ ï¼
- shader stageï¼
vertex/fragment/compute - entry point åç§°ï¼ä¾å¦
vs_main/fs_main/cs_mainï¼ - è¾å ¥/è¾åºï¼é¡¶ç¹å±æ§å¸å±ãinter-stage åéãfragment è¾åº attachment æ°é
- èµæºï¼æåªäº buffer/texture/samplerï¼åå«ç¨äºåªä¸ª stage
- shader stageï¼
- å
ç»âæ¥å£âååâç®æ³âï¼
- ç¨
structå®ä¹ VSIn / VSOut / FSIn / FSOutï¼æç´æ¥ç¨åæ°ä¸è¿åæ æ³¨ï¼ - 忏
ææ
@location(n)ä¸@builtin(...) - 忏
ææ
@group(g) @binding(b)声æ
- ç¨
- åå®ç°é»è¾ï¼å½æ° + æ§å¶æµï¼ï¼å¹¶åâå¯ç¼è¯èªæ£æ¸ åâï¼è§ææ«ï¼ã
- è¾åºæ¶å¿
须带ä¸ï¼
- 宿´ WGSL
- Bindings æè¦è¡¨ï¼group/binding/kind/type/stageï¼
- IO æè¦è¡¨ï¼locations/builtins/ç±»åï¼
ð§± è¯è¨æ ¸å¿éæ¥ï¼å对æ¯åå¿«æ´éè¦ï¼
1) ç±»åä¸åé¢é
- æ éï¼
i32/u32/f32/boolï¼f16å¯éï¼ - æ°ååç¼ï¼å¸¸ç¨ï¼ï¼
3uâu32ï¼2iâi32ï¼4f/4.5fâf32ï¼5hâf166/7.0å¯è½æ¯æ½è±¡æ°ï¼ç¼è¯ææ¨å¯¼ï¼ï¼ä½ä¸è¦ä¾èµå®è·¨è¡¨è¾¾å¼èªå¨âå¸®ä½ è½¬æ¢â
- 常è§åï¼
let a = 1; let b = 2.0; a + bä¼ç±»åä¸å¹é ï¼æ¹æf32(a) + bææä¸¤è¾¹ç»ä¸æåç±»åã
2) let / var / const
letï¼ä¸å¯åå¼ï¼å¸¸ç¨ï¼varï¼å¯ååéï¼æåå¨ï¼è½èµå¼ï¼constï¼ç¼è¯æå¸¸éï¼åªè½ç±ç¼è¯æè¡¨è¾¾å¼ç»æï¼ä¸è½ä¾èµè¿è¡æ¶å¼ï¼overrideï¼ç®¡çº¿å¸¸éï¼specialization constantï¼ï¼ç±å®¿ä¸»å¨å建 pipeline æ¶æä¾ï¼éååå¼å ³/常éåæ°/å·¥ä½ç»å¤§å°çï¼è§ä¸ææ¨¡æ¿ï¼
3) åéä¸ç©éµ
- åéï¼
vec2<T>/vec3<T>/vec4<T>ï¼å¸¸ç¨å«åï¼vec2f/vec3f/vec4fï¼f32ï¼ãvec4uï¼u32ï¼ç - 访é®ï¼
.x/.y/.z/.wã.r/.g/.b/.aãa[i] - swizzleï¼
a.zxãa.zzyç - ç©éµï¼
matCxR<T>ï¼ååéæ°ç»ï¼ï¼å¸¸ç¨mat4x4fm[2]å第ä¸ååémat4x4f * vec4fåæ³ï¼è¿åvec4f
4) æ°ç»
- åºå®å¤§å°ï¼
array<T, N> - æé å¨ï¼
array(v0, v1, ...)æarray<T, N>(...) - runtime-sized arrayï¼
- ä» å è®¸å¨æ ¹ä½ç¨å storage 声æï¼æä½ä¸ºæ ¹ä½ç¨å struct çæåä¸ä¸ªå段ï¼å¹¶ç»å®ä¸º storageï¼
- ç¨
arrayLength(&arr)æ¥è¯¢é¿åº¦ï¼ä» 对 runtime-sized array ææä¹ï¼
5) 彿°ä¸å ¥å£ç¹
- 彿°ï¼
fn name(params) -> returnType { ... } - å
¥å£ç¹ï¼
@vertex/@fragment/@compute - æå¿½ç¥ä½å¾å
³é®ï¼
- shader âéè¦åªäº bindingsâï¼åªç± entry point å¯è¾¾ç访é®è·¯å¾å³å®ï¼æªè¢«å ¥å£ç¹è®¿é®å°çå ¨å±èµæºä¸ä¼æä¸ºå¿ éç»å®ï¼
- å¦æä½ æ³å¼ºå¶æä¸ª binding åºç°å¨ç®¡çº¿éæ±éï¼å¯ç¨
_ = resource;åâå使ç¨â
6) overrideï¼æ¨èï¼æå¯è°åæ°é½åæå®ï¼
override USE_FOG: bool = false;
override WORKGROUP_SIZE_X: u32 = 64u;
@compute @workgroup_size(WORKGROUP_SIZE_X)
fn cs_main(@builtin(global_invocation_id) gid: vec3u) {
if (USE_FOG) {
// ...
}
}
è¦ç¹ï¼
overrideæ¯â管线å建æ¶å³å®âç常éï¼ä¸æ¯è¿è¡æ¶åéã- éåï¼åæ¯å¼å ³ãæ°ç»ä¸éãworkgroup sizeãç®æ³å¸¸éçã
ð 管线 IO ä¸ Attributesï¼æå¸¸åºéçå°æ¹ï¼
@location(n)ï¼ç¨æ·èªå®ä¹ IO
- VS è¾å
¥ï¼entry point åæ°ä¸ç
@location(n)ï¼æåæ° struct åæ®µä¸ç@location(n) - VSâFSï¼VS è¾åº struct åæ®µ
@location(n)ä¸ FS è¾å ¥ struct åæ®µ@location(n)å¿ é¡»å¹é ï¼ååå¯ä¸åï¼location å¿ é¡»å¯¹ï¼ - FS è¾åºï¼
@location(n)对åºç¬¬ n 个 color attachment
@builtin(name)ï¼å
建 IO
常ç¨ä¾åï¼
- VS è¾å
¥ï¼
@builtin(vertex_index)ã@builtin(instance_index) - VS è¾åºï¼
@builtin(position)ï¼è£åªç©ºé´ positionï¼é常vec4fï¼
æå¼æ§å¶ï¼inter-stage ç»èï¼å对äºè½å°å¾å¤â奿ªéªç/é¢è²ä¸å¯¹âï¼
- é»è®¤ä¼å¯¹ float varyings åéè§æ£ç¡®æå¼ï¼perspective interpolationï¼ã
- 对 æ´å/æä¸¾/ID è¿ç±»ä¸è¯¥æå¼çå¼ï¼æ¾å¼ä½¿ç¨
@interpolate(flat)ï¼å¦å伿¥éæäº§çä¸ç¬¦å颿çç»æï¼ã - éè¦å±å¹ç©ºé´çº¿æ§æå¼æ¶ï¼å¯ç¨
@interpolate(linear)ï¼æé使ç¨ï¼ã - MSAA ä¸çæå¼éæ ·ä½ç½®ï¼fragmentï¼ï¼
@interpolate(center)ï¼é»è®¤ï¼æåç´ ä¸å¿æå¼@interpolate(centroid)ï¼ä¿è¯éæ ·ç¹å¨å¾å è¦çåºåå ï¼æ´éåè¾¹ç¼ï¼@interpolate(sample)ï¼æ¯ä¸ª sample 齿§è¡ä¸æ¬¡ fragment shaderï¼æ´è´µï¼ä½å¯ç²¾ç¡®å° sampleï¼
Builtins å ¨è¡¨ï¼æ stage/æ¹å使ç¨ï¼é¿å åéï¼
说æï¼ä»¥ä¸è¡¨æ WebGPU 常ç¨å 建æ´çã
positionå¨ VS è¾åºä¸ FS è¾å ¥è¯ä¹ä¸åã
| builtin | stage | io | type | 夿³¨ |
|---|---|---|---|---|
vertex_index |
vertex | in | u32 |
å½å draw çé¡¶ç¹ç´¢å¼ï¼ä¸ index buffer / baseVertex çç¸å ³ï¼ |
instance_index |
vertex | in | u32 |
å½å draw ç instance ç´¢å¼ |
position |
vertex | out | vec4f |
è£åªç©ºé´åæ ï¼é½æ¬¡åæ ï¼ |
position |
fragment | in | vec4f |
framebuffer 空é´ä½ç½®ï¼w 为 1ï¼xy åç´ åæ è¯ä¹ï¼ |
front_facing |
fragment | in | bool |
æ¯å¦æ£é¢ |
frag_depth |
fragment | out | f32 |
åå ¥æ·±åº¦ï¼ä¸ååç¨åºå®ç®¡çº¿æ·±åº¦ï¼ |
sample_index |
fragment | in | u32 |
å½å sampleï¼MSAAï¼ |
sample_mask |
fragment | in | u32 |
è¦çå°ç samples mask |
sample_mask |
fragment | out | u32 |
è¾åº samples maskï¼å¯å±è½æäº samples åå ¥ï¼ |
local_invocation_id |
compute | in | vec3u |
workgroup å 线ç¨åæ |
local_invocation_index |
compute | in | u32 |
workgroup å 线æ§ç´¢å¼ |
global_invocation_id |
compute | in | vec3u |
å ¨å±çº¿ç¨åæ |
workgroup_id |
compute | in | vec3u |
workgroup åæ |
num_workgroups |
compute | in | vec3u |
dispatch ç workgroup æ°é |
ð§© èµæºç»å®ï¼Bindingsï¼åæ³æ¨¡æ¿
注æï¼æäºæ§æç« /示ä¾ä¼åæ
var<uniforms>ï¼ç°ä»£ WGSL å ³é®åæ¯var<uniform>ã
struct Uniforms {
color: vec4f,
};
@group(0) @binding(0)
var<uniform> uniforms: Uniforms;
@group(0) @binding(1)
var my_sampler: sampler;
@group(0) @binding(2)
var my_tex: texture_2d<f32>;
storage bufferï¼ç¤ºä¾ï¼ï¼
@group(0) @binding(3)
var<storage, read_write> data: array<vec4f>;
Address spaces / access modeï¼ç»å¸¸æ¼å/åéï¼
- module-scope èµæºåéå¿
é¡»æ¯
var<address_space, access>è¿ç§å½¢å¼ï¼ä¾å¦var<uniform>ãvar<storage, read>ãvar<storage, read_write>ï¼ã arrayLength(&x)éç&æ¯âåå¼ç¨/æéâï¼å¾å¤å 建éè¦ä¼ æéï¼å°¤å ¶é对 runtime-sized arrayï¼ã- compute ä¸éè¦è·¨çº¿ç¨å
±äº«ä¸é´ç»ææ¶ï¼ç¨
var<workgroup>ï¼åªå¨å½å invocation å ç¨ä¸´æ¶åéæ¶ï¼ç¨å½æ°åvarå³å¯ã
纹ç/éæ ·ï¼å³å®äºä½ è¯¥ç¨ sample è¿æ¯ loadï¼
常ç¨ç±»åï¼
- å¯éæ · 2Dï¼
texture_2d<f32>+samplerâ 常ç¨textureSample(...) - ä¸å¯è¿æ»¤ï¼unfilterableï¼æéè¦ç²¾ç¡® texelï¼ä¼å
textureLoad(...)ï¼åæ éå¸¸æ¯æ´æ° texel + mipï¼ - 深度ï¼
texture_depth_2då¾å¾æé sampler_comparisonä¸textureSampleCompare(...) - Storage textureï¼
texture_storage_2d<format, access>ï¼ç¨äºè¯»ååç´ ï¼åéæ ·çº¹çæ¯ä¸¤å¥ä¸è¥¿ï¼
èªæ£æç¤ºï¼é常å®ç¨ï¼ï¼
textureSampleéè¦âå¯è¿æ»¤/å¯éæ ·âççº¹çæ ¼å¼ä¸å¹é ç samplerï¼å¦æéªè¯å±æ¥âfiltering sampler / sampleType ä¸å¹é âï¼ä¼å åå° bind group layout æ£æ¥ sampleType ä¸ sampler ç±»åãtextureLoadæ¯ä¸ç»è¿éæ ·å¨çï¼å¸¸ç¨äº G-bufferãå¾åå¤çãunfilterableãstorage texture ç读åçåºæ¯ã
纹çå®¶æä¸æ¥è¯¢å½æ°ï¼æâè½ä¸è½ç¨â䏿¬¡è¯´æ¸ ï¼
å¸¸è§ sampled texturesï¼é
samplerï¼ï¼
texture_1d<f32>ï¼è¾å°è§ï¼texture_2d<f32>/texture_2d_array<f32>texture_3d<f32>texture_cube<f32>/texture_cube_array<f32>texture_multisampled_2d<f32>ï¼MSAA colorï¼texture_externalï¼å¤é¨å¾åæºï¼åéæ´å¤ï¼
å¸¸è§ depth texturesï¼é
sampler æ sampler_comparisonï¼ï¼
texture_depth_2d/texture_depth_2d_arraytexture_depth_cube/texture_depth_cube_arraytexture_depth_multisampled_2d
å¸¸è§ storage texturesï¼ä¸é samplerï¼ç´æ¥ load/storeï¼ï¼
texture_storage_2d<F, access>/texture_storage_2d_array<F, access>texture_storage_3d<F, access>
è¦ç¹ï¼
- storage texture çè¯»åæ¯âé texelâçï¼ç¨
textureLoad/textureStoreã - storage texture é常åªä½ç¨å¨ mip level 0ï¼æ¥å£ä¸ä¸éè¦/䏿¯æä¼ mip levelï¼ï¼ä¸è¦æ sampled texture ç
textureLoad(..., level)ä¹ æ¯å¸¦è¿æ¥ã
é«é¢âæ¥è¯¢/è¾ å©â彿°ï¼å shader æ¶å¸¸ç¨æ¥é¿å 硬ç¼ç ï¼ï¼
textureDimensions(...)ï¼çº¹çå°ºå¯¸ï¼æç»´åº¦è¿åu32/vec2u/vec3uï¼mip çæ¬ä¹æï¼textureNumLevels(...)ï¼mip 屿°textureNumLayers(...)ï¼array layerstextureNumSamples(...)ï¼MSAA sample æ°
éæ ·å½æ°éæ©ï¼æå¸¸è§çæ£ç¡®å§¿å¿ï¼
å ³é®ç¹ï¼å¾å¤éæ ·å½æ°å¨âé uniform control flowâéä¼åºé®é¢ï¼è§åé¢ç Uniformity ç« èï¼ã
- æå¸¸ç¨ï¼
textureSample(t, s, coords)ï¼éå¼ LODï¼é常ä¾èµå¯¼æ°ï¼ - æ¾å¼ LODï¼
textureSampleLevel(t, s, coords, level)ï¼ä¸è®¡ç®å¯¼æ°ï¼éå忝ééæ ·ï¼ - æ¾å¼å¯¼æ°ï¼
textureSampleGrad(t, s, coords, dpdx, dpdy)ï¼èªå·±æä¾å¯¼æ°ï¼ - åç½® LODï¼
textureSampleBias(t, s, coords, bias)ï¼åæ ·å¯è½å uniformity 约æå½±åï¼ - 精确 texelï¼
- sampled/depthï¼
textureLoad(t, coords, level)ï¼æ´æ° texel åæ ï¼MSAA åä½ä¼å¸¦sample_indexï¼ - storageï¼
textureLoad(t, coords)ï¼é texel 读åï¼é常åªå¯¹åº level 0ï¼
- sampled/depthï¼
- åå
¥ storageï¼
textureStore(t, coords, value)ï¼storage texture æè½åï¼
texture_external æç¤ºï¼
- 常ç¨äºè§é¢/å¤é¨å¾åæºï¼å¯ç¨
textureLoad(texture_external, coords)ã - éæ ·æ¶å¸¸ç¨
textureSampleBaseClampToEdge(...)ï¼ç¨äºå¤é¨çº¹ççåééæ ·åºæ¯ï¼ã
深度/模æ¿ï¼Depth & Stencilï¼æç¼
- 深度âéæ ·âï¼
textureSample(texture_depth_*, sampler, coords)è¿åf32深度å¼- æ·±åº¦âæ¯è¾éæ ·âï¼
textureSampleCompare(texture_depth_*, sampler_comparison, coords, depth_ref)è¿åæ¯è¾ç»æï¼0..1ï¼ - æäºåºæ¯å¯ç¨
textureSampleCompareLevel(..., level=0)ï¼å®ä¸è®¡ç®å¯¼æ°ã䏿²¡æ uniform control flow éå¶ï¼å¹¶ä¸å¯å¨ä»»æ stage è°ç¨ï¼éåé¿å¼ uniformity é®é¢ï¼ã
- å
frag_depthï¼- fragment è¾åº
@builtin(frag_depth) depth: f32æç´æ¥è¿å弿 注 builtinï¼åå³äºä½ ç¨ struct è¿æ¯è¿åæ æ³¨ï¼ã
- fragment è¾åº
- stencilï¼
- stencil æ´å¤ç±ç®¡çº¿ç¶ææ§å¶ï¼compare/op/writeMask çï¼ï¼WGSL ä¾§é常ä¸â读å stencil 弿¬èº«âã
宿主侧ç»å®çº¦æï¼WebGPU 常è§éªè¯éè¯¯æ¥æºï¼
- 妿 pipeline ç¨
layout: 'auto'ï¼WebGPU 常è§é»è®¤ï¼ï¼WebGPU 伿 ¹æ® WGSL èªå¨æ¨å¯¼ bind group layoutï¼è¿ä¼å¯¹çº¹çç sampleTypeãsampler ç±»åçååºæ´ä¸¥æ ¼çè¦æ±ã - æ³å¨å¤ä¸ª pipeline ä¹é´å¤ç¨åä¸ä¸ª bind group æ¶ï¼é常éè¦æ¾å¼åå»ºå¹¶å ±äº«åä¸ä¸ª bind group layout / pipeline layoutï¼é¿å âauto layout ä¸å ¼å®¹âï¼ã
- ä½¿ç¨ dynamic offsets æ¶ï¼éè¦å¨ bind group layout 䏿¾å¼å¼å¯å¯¹åº binding ç卿åç§»è½åï¼å¦å
setBindGroupä¼ offsets ä¼ç´æ¥æ¥éã
ð¦ æ°æ®å¸å±ä¸å¯¹é½ï¼uniform/storage æå¸¸è§ bugï¼
- buffer éç
struct/arrayæå¯¹é½ä¸ size è§åï¼CPU ä¾§åå ¥ç bytes å¿ é¡»å WGSL ä¾§å¸å±ä¸è´ã - é«é¢åï¼
vec3<f32>ç对é½éå¸¸æ¯ 16 bytesï¼æä»¥å®å¨ struct/array éç»å¸¸ä¼äº§ç paddingï¼âçèµ·æ¥è¿ç»âçåæ®µå¨å åéå¯è½å¹¶ä¸è¿ç»ã - è¿ç±»é误é常ä¸ä¼æ¥éï¼ä½ä¼è®© shader 读å°éè¯¯æ°æ®ï¼è¡¨ç°ä¸ºæ¨¡åä¸è§äº/æ°å¼è®¡ç®æªå¼ï¼ã
- å®æå»ºè®®ï¼
- ä¸ç¡®å®å¸å±æ¶ï¼ä¼å
ç¨
vec4f对é½ï¼æå¨ struct 䏿¾å¼å padding åæ®µï¼ä¾å¦pad0: f32ï¼ã - ç¨å·¥å ·/计ç®å¨èªå¨ç® offset/sizeï¼é¿å æç®ï¼ã
- ä¸ç¡®å®å¸å±æ¶ï¼ä¼å
ç¨
𧪠常ç¨éª¨æ¶ï¼ç´æ¥æ¹å°±è½ç¨ï¼
Vertex + Fragmentï¼æå°å¯æ¥å ¥éª¨æ¶ï¼
struct VSIn {
@location(0) position: vec3f,
@location(1) uv: vec2f,
};
struct VSOut {
@builtin(position) clip_pos: vec4f,
@location(0) uv: vec2f,
};
@vertex
fn vs_main(v: VSIn) -> VSOut {
var o: VSOut;
o.clip_pos = vec4f(v.position, 1.0);
o.uv = v.uv;
return o;
}
@group(0) @binding(0) var my_sampler: sampler;
@group(0) @binding(1) var my_tex: texture_2d<f32>;
struct FSOut {
@location(0) color: vec4f,
};
@fragment
fn fs_main(@location(0) uv: vec2f) -> FSOut {
var o: FSOut;
o.color = textureSample(my_tex, my_sampler, uv);
return o;
}
Computeï¼æå°éª¨æ¶ï¼
@group(0) @binding(0)
var<storage, read_write> out_data: array<u32>;
@compute @workgroup_size(64)
fn cs_main(@builtin(global_invocation_id) gid: vec3u) {
let idx = gid.x;
if (idx < arrayLength(&out_data)) {
out_data[idx] = idx;
}
}
Compute å¹¶è¡æ¨¡åè¡¥å ï¼é¿å æ°æ®ç«äºï¼
- å¸¸ç¨ builtinsï¼
@builtin(global_invocation_id)ï¼å ¨å±çº¿ç¨ idï¼è·¨å·¥ä½ç»å¯ä¸ï¼@builtin(local_invocation_id)ï¼å·¥ä½ç»å çº¿ç¨ id@builtin(local_invocation_index)ï¼å·¥ä½ç»å 线æ§ç´¢å¼@builtin(workgroup_id)ï¼å·¥ä½ç» id
- å storage buffer æ¶é»è®¤å°±æå¹¶ååå
¥é£é©ï¼
- æ¯ä¸ª invocation åèªå·±ç¬å ç indexï¼æç®åãæå®å ¨ï¼
- éè¦æ±æ»/è®¡æ°æ¶ï¼ç¨
atomic<T>ä¸atomicAdd/atomicMin/...ï¼å¦åä¼åºç°ç«æï¼ - éè¦å·¥ä½ç»å
å
±äº«æ°æ®æ¶ï¼ç¨
var<workgroup>å¹¶å¨è¯»åé¶æ®µä¹é´æå ¥workgroupBarrier()
â ï¸ è¯æ³å·®å¼ä¸æéç¹æ¸ åï¼åå®éæ¡å¯¹ï¼
- 没æä¸å
è¿ç®ç¬¦ï¼ç¨
select(falseValue, trueValue, cond)ã - æ§å¶æµä¸å¼ºå¶æ¬å·ï¼
if cond { ... }åæ³ï¼å«ææ¬å·å½æå¿ é¡»ï¼ã ++/--æ¯è¯å¥ä¸æ¯è¡¨è¾¾å¼ï¼ä¸è½ålet x = a++;ï¼åªè½åç¬ä¸è¡a++;ã+=/-=æ¯èµå¼è¯å¥ï¼ä¸æ¯è¡¨è¾¾å¼ï¼ä¸è½ålet x = (a += 1);ã- swizzle ä¸è½åºç°å¨èµå¼å·¦ä¾§ï¼ä¸è½å
color.rgb = ...ï¼è¦æé æ°åéåæ´ä½èµå¼ã - éå°ä½ç§»/æ¯è¾ç¸å ³è§£æé误æ¶ï¼ä¼å ç»å表达å¼å æ¬å·ï¼WGSL çè¿ç®ç¬¦ä¼å 级/è§£æè§åæ¯å¾å¤è¯è¨æ´âä¸¥æ ¼âï¼ã
discardåªè½å¨ fragment shader 使ç¨ãswitchåªæ¯æi32/u32ï¼caseå¿ é¡»æ¯å¸¸éã- éå°âæä¸ª binding ä¸å¨é误æç¤ºé/ä¸è¢«è¦æ±âï¼
- æ£æ¥å®æ¯å¦çç被 entry point å¯è¾¾ä»£ç 访é®å°ï¼å¿
è¦æ¶
_ = resource;å¼ºå¶æ 记使ç¨ã
- æ£æ¥å®æ¯å¦çç被 entry point å¯è¾¾ä»£ç 访é®å°ï¼å¿
è¦æ¶
ð§ Uniformity / Derivativesï¼å¾å¤âè½ç¼è¯ä½ç»é¢æª/ä¸åæºå¨ä¸ä¸è´âçæ ¹æºï¼
- 导æ°å
建ï¼
dpdx/dpdy/...ï¼ä»¥åä¾èµå¯¼æ°çæä½ï¼å¨é uniform control flow ä¸ä¼äº§çä¸ç¡®å®ç»æï¼å¸¸è§è¡¨ç°ï¼éªè¯å±æ¥éãæç»æ indeterminateï¼ã textureSample*é大éåä½é½ä¼æ 注 âReturns an indeterminate value if called in non-uniform control flow.âï¼å°¤å ¶æ¯éå¼ LOD çé£äºï¼ã- 宿çç¥ï¼
- åæ¯éæ³éæ ·ï¼ä¼å
ç¨
textureSampleLevelï¼æ¾å¼ levelï¼ætextureLoadï¼æ texelï¼ - å¿ é¡»éå¼ LODï¼æéæ ·æªå°åæ¯å¤ï¼æä¿è¯æ§å¶æµå¯¹åä¸ quad ç invocations ä¸è´ï¼æ´é¾ï¼
- 深度æ¯è¾ä¸æ uniformityï¼å¯èè
textureSampleCompareLevelï¼åºå® level=0ï¼ä¸è®¡ç®å¯¼æ°ï¼
- åæ¯éæ³éæ ·ï¼ä¼å
ç¨
ð§µ ååä¸åæ¥ï¼æâé«çº§å¹¶è¡/忥âçº³å ¥èå´ï¼
WGSL å å»ºåæ¥å½æ°ï¼æ¥èª WGSL Function Referenceï¼ï¼
storageBarrier()ï¼å½±åstorageå°å空é´çå åä¸ atomicworkgroupBarrier()ï¼å½±åworkgroupå°å空é´çå åä¸ atomicworkgroupUniformLoad(ptr<workgroup, T>) -> Tï¼æ workgroup å æå°åçå¼âç»ä¸å¹¿æâ为 uniform å¼ï¼æéæ¬èº«å¿ é¡»æ¯ uniformï¼
ååç±»å䏿ä½è¦ç¹ï¼
- ååç±»åæ¯
atomic<T>ï¼å¸¸è§atomic<u32>/atomic<i32>ï¼ï¼ç¨äºå¨å¤ä¸ª invocations é´å®å ¨ç´¯å /计æ°/æå¼çã - åå + barrier ç使ç¨è¦æç¡®é¶æ®µï¼
- workgroup å
å
±äº«ï¼
var<workgroup>+workgroupBarrier()åâåå®å读â - storage çè·¨ workgroup éä¿¡é常åéï¼é常éè¿åå/åé¶æ®µ pass 设计解å³ï¼å«ææ barrier è®©å ¨å±æåºï¼
- workgroup å
å
±äº«ï¼
ð§© å¯éç¹æ§ä¸ enableï¼ç¡¬åæ¢ï¼ä¸ç¨å°±ä¸åï¼ç¨å°±åå ¨ï¼
f16ï¼- WGSL ä¸
f16æ¯å¯éè½åï¼å¦æè¦ç¨ï¼shader 顶鍿¾å¼åenable f16;ï¼å¹¶ä¸å®¿ä¸»ä¾§å¿ é¡»å¨å建 device æ¶è¯·æ±å¯¹åº featureï¼å¦åä¼ç¼è¯/å建失败ï¼ã
- WGSL ä¸
- storage texture ç
bgra8unormï¼bgra8unormé»è®¤ä¸è½ä½ä¸º storage textureï¼éè¦å®¿ä¸»è¯·æ± WebGPU featureï¼bgra8unorm-storageï¼è¯¦è§ Storage Textures æç« ï¼ã
- ååï¼é常éè¦ï¼ï¼
- ä¸è¦é»è®¤å¼å¯ä»»ä½å¯éç¹æ§ï¼å¿ é¡»ç±ç¨æ·æç¡®è¦æ±ï¼ææç¡®è¯´æå®¿ä¸»å·²å¯ç¨ï¼ã
- 䏿¦å¯ç¨ï¼ä¸å fallbackï¼hard cutoverï¼ã
enable ...;æ¾å¨æ¨¡åä½ç¨åçé¡¶é¨ï¼å¨å£°æ/彿°ä¹åï¼ï¼å¹¶ä¸åªåä½ ç¡®å®è¦ç¨ç项ã- éå°âå®éªæ§ enable åç§°/feature åç§°âä¸è¦çï¼è®©ç¨æ·ç»åºåç¡®å符串ï¼åæå ¶è¦æ±åå ¥ï¼å¹¶åæ¥æ´æ° Bindings/ç®¡çº¿éæ±ï¼ã
ð å¯ç¼è¯èªæ£æ¸ åï¼åå®å¿ åï¼
-
å ¥å£ç¹ç¾åï¼
@vertex/@fragment/@computeæ¯å¦é½å ¨ï¼è¿åå¼/è¾åº struct ç@location/@builtinæ¯å¦å®æ´ï¼ -
ç±»åä¸è´æ§ï¼ææç®æ¯ä¸¤ç«¯ç±»åæ¯å¦ä¸è´ï¼éè¦æ¶æ¯å¦åäº
f32(...)/u32(...)çæ¾å¼è½¬æ¢ï¼ -
Bindings 宿´ï¼
@group/@bindingæ¯å¦å¯ä¸ä¸ä¸å²çªï¼var<uniform>/var<storage, ...>/texture_*/sampler*çç±»åæ¯å¦å宿主侧 bind group layout ä¸è´ï¼
-
纹ç彿°å¹é ï¼
- ç¨äº
textureSampleççº¹çæ¯å¦âå¯éæ ·ä¸å¯è¿æ»¤âï¼sampler ç±»åæ¯å¦å¹é ï¼ - ç¨äº
textureLoadçåæ /level ç±»åæ¯å¦æ£ç¡®ï¼éå¸¸æ¯æ´æ° texel + mipï¼ï¼
- ç¨äº
-
æ°æ®å¸å±ï¼uniform/storage éç struct æ¯å¦èèäº paddingï¼æ¯å¦é¿å äº
vec3f带æ¥çéå¼ padding é·é±ï¼ -
Compute å®å ¨ï¼
- 对 output buffer åå
¥æ¯å¦åäºè¾¹çä¿æ¤ï¼
idx < arrayLength(...)ï¼ï¼ - æ¯å¦åå¨å¤ä¸ª invocation ååä¸ä½ç½®ï¼è¥æï¼æ¯å¦ä½¿ç¨ atomic ææ¹å为æ å²çªåå ¥ï¼
- 对 output buffer åå
¥æ¯å¦åäºè¾¹çä¿æ¤ï¼
-
Uniformityï¼fragment 䏿¯å¦å¨åæ¯/循ç¯éè°ç¨äº
dpdx/dpdy/textureSample*è¿ç±»å¯¹ uniformity ææç彿°ï¼è½å¦æ¹ætextureSampleLevel/textureLoadï¼
â è¾åºæ ¼å¼ï¼å¼ºå¶ï¼
å½ä½¿ç¨æ¬ skill çæ WGSL æ¶ï¼è¾åºå¿ é¡»å å« 3 段ï¼
WGSLï¼å®æ´ä»£ç åï¼wgsl ...ï¼Bindingsï¼Markdown 表ï¼group/binding/kind/wgsl type/stages/noteï¼IOï¼Markdown 表ï¼stage/location or builtin/name/typeï¼