export const description = `Validation tests for identifiers`; import { makeTestGroup } from '../../../../common/framework/test_group.js'; import { keysOf } from '../../../../common/util/data_tables.js'; import { ShaderValidationTest } from '../shader_validation_test.js'; export const g = makeTestGroup(ShaderValidationTest); g.test('function_param') .desc( `Test that a function param can shadow a builtin, but the builtin is available for other params.` ) .fn(t => { const code = ` fn f(f: i32, i32: i32, t: i32) -> i32 { return i32; } `; t.expectCompileResult(true, code); }); const kTests = { abs: { keyword: `abs`, src: `_ = abs(1);`, }, acos: { keyword: `acos`, src: `_ = acos(.2);`, }, acosh: { keyword: `acosh`, src: `_ = acosh(1.2);`, }, all: { keyword: `all`, src: `_ = all(true);`, }, any: { keyword: `any`, src: `_ = any(true);`, }, array_templated: { keyword: `array`, src: `_ = array(1, 2);`, }, array: { keyword: `array`, src: `_ = array(1, 2);`, }, array_length: { keyword: `arrayLength`, src: `_ = arrayLength(&placeholder.rt_arr);`, }, asin: { keyword: `asin`, src: `_ = asin(.2);`, }, asinh: { keyword: `asinh`, src: `_ = asinh(1.2);`, }, atan: { keyword: `atan`, src: `_ = atan(1.2);`, }, atanh: { keyword: `atanh`, src: `_ = atanh(.2);`, }, atan2: { keyword: `atan2`, src: `_ = atan2(1.2, 2.3);`, }, bool: { keyword: `bool`, src: `_ = bool(1);`, }, bitcast: { keyword: `bitcast`, src: `_ = bitcast(1i);`, }, ceil: { keyword: `ceil`, src: `_ = ceil(1.23);`, }, clamp: { keyword: `clamp`, src: `_ = clamp(1, 2, 3);`, }, cos: { keyword: `cos`, src: `_ = cos(2);`, }, cosh: { keyword: `cosh`, src: `_ = cosh(2.2);`, }, countLeadingZeros: { keyword: `countLeadingZeros`, src: `_ = countLeadingZeros(1);`, }, countOneBits: { keyword: `countOneBits`, src: `_ = countOneBits(1);`, }, countTrailingZeros: { keyword: `countTrailingZeros`, src: `_ = countTrailingZeros(1);`, }, cross: { keyword: `cross`, src: `_ = cross(vec3(1, 2, 3), vec3(4, 5, 6));`, }, degrees: { keyword: `degrees`, src: `_ = degrees(1);`, }, determinant: { keyword: `determinant`, src: `_ = determinant(mat2x2(1, 2, 3, 4));`, }, distance: { keyword: `distance`, src: `_ = distance(1, 2);`, }, dot: { keyword: `dot`, src: `_ = dot(vec2(1, 2,), vec2(2, 3));`, }, dot4U8Packed: { keyword: `dot4U8Packed`, src: `_ = dot4U8Packed(1, 2);`, }, dot4I8Packed: { keyword: `dot4I8Packed`, src: `_ = dot4I8Packed(1, 2);`, }, dpdx: { keyword: `dpdx`, src: `_ = dpdx(2);`, }, dpdxCoarse: { keyword: `dpdxCoarse`, src: `_ = dpdxCoarse(2);`, }, dpdxFine: { keyword: `dpdxFine`, src: `_ = dpdxFine(2);`, }, dpdy: { keyword: `dpdy`, src: `_ = dpdy(2);`, }, dpdyCoarse: { keyword: `dpdyCoarse`, src: `_ = dpdyCoarse(2);`, }, dpdyFine: { keyword: `dpdyFine`, src: `_ = dpdyFine(2);`, }, exp: { keyword: `exp`, src: `_ = exp(1);`, }, exp2: { keyword: `exp2`, src: `_ = exp2(2);`, }, extractBits: { keyword: `extractBits`, src: `_ = extractBits(1, 2, 3);`, }, f32: { keyword: `f32`, src: `_ = f32(1i);`, }, faceForward: { keyword: `faceForward`, src: `_ = faceForward(vec2(1, 2), vec2(3, 4), vec2(5, 6));`, }, firstLeadingBit: { keyword: `firstLeadingBit`, src: `_ = firstLeadingBit(1);`, }, firstTrailingBit: { keyword: `firstTrailingBit`, src: `_ = firstTrailingBit(1);`, }, floor: { keyword: `floor`, src: `_ = floor(1.2);`, }, fma: { keyword: `fma`, src: `_ = fma(1, 2, 3);`, }, fract: { keyword: `fract`, src: `_ = fract(1);`, }, frexp: { keyword: `frexp`, src: `_ = frexp(1);`, }, fwidth: { keyword: `fwidth`, src: `_ = fwidth(2);`, }, fwidthCoarse: { keyword: `fwidthCoarse`, src: `_ = fwidthCoarse(2);`, }, fwidthFine: { keyword: `fwidthFine`, src: `_ = fwidthFine(2);`, }, i32: { keyword: `i32`, src: `_ = i32(2u);`, }, insertBits: { keyword: `insertBits`, src: `_ = insertBits(1, 2, 3, 4);`, }, inverseSqrt: { keyword: `inverseSqrt`, src: `_ = inverseSqrt(1);`, }, ldexp: { keyword: `ldexp`, src: `_ = ldexp(1, 2);`, }, length: { keyword: `length`, src: `_ = length(1);`, }, log: { keyword: `log`, src: `_ = log(2);`, }, log2: { keyword: `log2`, src: `_ = log2(2);`, }, mat2x2_templated: { keyword: `mat2x2`, src: `_ = mat2x2(1, 2, 3, 4);`, }, mat2x2: { keyword: `mat2x2`, src: `_ = mat2x2(1, 2, 3, 4);`, }, mat2x3_templated: { keyword: `mat2x3`, src: `_ = mat2x3(1, 2, 3, 4, 5, 6);`, }, mat2x3: { keyword: `mat2x3`, src: `_ = mat2x3(1, 2, 3, 4, 5, 6);`, }, mat2x4_templated: { keyword: `mat2x4`, src: `_ = mat2x4(1, 2, 3, 4, 5, 6, 7, 8);`, }, mat2x4: { keyword: `mat2x4`, src: `_ = mat2x4(1, 2, 3, 4, 5, 6, 7, 8);`, }, mat3x2_templated: { keyword: `mat3x2`, src: `_ = mat3x2(1, 2, 3, 4, 5, 6);`, }, mat3x2: { keyword: `mat3x2`, src: `_ = mat3x2(1, 2, 3, 4, 5, 6);`, }, mat3x3_templated: { keyword: `mat3x3`, src: `_ = mat3x3(1, 2, 3, 4, 5, 6, 7, 8, 9);`, }, mat3x3: { keyword: `mat3x3`, src: `_ = mat3x3(1, 2, 3, 4, 5, 6, 7, 8, 9);`, }, mat3x4_templated: { keyword: `mat3x4`, src: `_ = mat3x4(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12);`, }, mat3x4: { keyword: `mat3x4`, src: `_ = mat3x4(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12);`, }, mat4x2_templated: { keyword: `mat4x2`, src: `_ = mat4x2(1, 2, 3, 4, 5, 6, 7, 8);`, }, mat4x2: { keyword: `mat4x2`, src: `_ = mat4x2(1, 2, 3, 4, 5, 6, 7, 8);`, }, mat4x3_templated: { keyword: `mat4x3`, src: `_ = mat4x3(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12);`, }, mat4x3: { keyword: `mat4x3`, src: `_ = mat4x3(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12);`, }, mat4x4_templated: { keyword: `mat4x4`, src: `_ = mat4x4(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16);`, }, mat4x4: { keyword: `mat4x4`, src: `_ = mat4x4(1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16);`, }, max: { keyword: `max`, src: `_ = max(1, 2);`, }, min: { keyword: `min`, src: `_ = min(1, 2);`, }, mix: { keyword: `mix`, src: `_ = mix(1, 2, 3);`, }, modf: { keyword: `modf`, src: `_ = modf(1.2);`, }, normalize: { keyword: `normalize`, src: `_ = normalize(vec2(1, 2));`, }, pack2x16snorm: { keyword: `pack2x16snorm`, src: `_ = pack2x16snorm(vec2(1, 2));`, }, pack2x16unorm: { keyword: `pack2x16unorm`, src: `_ = pack2x16unorm(vec2(1, 2));`, }, pack2x16float: { keyword: `pack2x16float`, src: `_ = pack2x16float(vec2(1, 2));`, }, pack4x8snorm: { keyword: `pack4x8snorm`, src: `_ = pack4x8snorm(vec4(1, 2, 3, 4));`, }, pack4x8unorm: { keyword: `pack4x8unorm`, src: `_ = pack4x8unorm(vec4(1, 2, 3, 4));`, }, pack4xI8: { keyword: `pack4xI8`, src: `_ = pack4xI8(vec4(1, 2, 3, 4));`, }, pack4xU8: { keyword: `pack4xU8`, src: `_ = pack4xU8(vec4(1, 2, 3, 4));`, }, pack4xI8Clamp: { keyword: `pack4xI8Clamp`, src: `_ = pack4xI8Clamp(vec4(1, 2, 3, 4));`, }, pack4xU8Clamp: { keyword: `pack4xU8Clamp`, src: `_ = pack4xU8Clamp(vec4(1, 2, 3, 4));`, }, pow: { keyword: `pow`, src: `_ = pow(1, 2);`, }, quantizeToF16: { keyword: `quantizeToF16`, src: `_ = quantizeToF16(1.2);`, }, radians: { keyword: `radians`, src: `_ = radians(1.2);`, }, reflect: { keyword: `reflect`, src: `_ = reflect(vec2(1, 2), vec2(3, 4));`, }, refract: { keyword: `refract`, src: `_ = refract(vec2(1, 1), vec2(2, 2), 3);`, }, reverseBits: { keyword: `reverseBits`, src: `_ = reverseBits(1);`, }, round: { keyword: `round`, src: `_ = round(1.2);`, }, saturate: { keyword: `saturate`, src: `_ = saturate(1);`, }, select: { keyword: `select`, src: `_ = select(1, 2, false);`, }, sign: { keyword: `sign`, src: `_ = sign(1);`, }, sin: { keyword: `sin`, src: `_ = sin(2);`, }, sinh: { keyword: `sinh`, src: `_ = sinh(3);`, }, smoothstep: { keyword: `smoothstep`, src: `_ = smoothstep(1, 2, 3);`, }, sqrt: { keyword: `sqrt`, src: `_ = sqrt(24);`, }, step: { keyword: `step`, src: `_ = step(4, 5);`, }, tan: { keyword: `tan`, src: `_ = tan(2);`, }, tanh: { keyword: `tanh`, src: `_ = tanh(2);`, }, transpose: { keyword: `transpose`, src: `_ = transpose(mat2x2(1, 2, 3, 4));`, }, trunc: { keyword: `trunc`, src: `_ = trunc(2);`, }, u32: { keyword: `u32`, src: `_ = u32(1i);`, }, unpack2x16snorm: { keyword: `unpack2x16snorm`, src: `_ = unpack2x16snorm(2);`, }, unpack2x16unorm: { keyword: `unpack2x16unorm`, src: `_ = unpack2x16unorm(2);`, }, unpack2x16float: { keyword: `unpack2x16float`, src: `_ = unpack2x16float(2);`, }, unpack4x8snorm: { keyword: `unpack4x8snorm`, src: `_ = unpack4x8snorm(4);`, }, unpack4x8unorm: { keyword: `unpack4x8unorm`, src: `_ = unpack4x8unorm(4);`, }, unpack4xI8: { keyword: `unpack4xI8`, src: `_ = unpack4xI8(4);`, }, unpack4xU8: { keyword: `unpack4xU8`, src: `_ = unpack4xU8(4);`, }, vec2_templated: { keyword: `vec2`, src: `_ = vec2(1, 2);`, }, vec2: { keyword: `vec2`, src: `_ = vec2(1, 2);`, }, vec3_templated: { keyword: `vec3`, src: `_ = vec3(1, 2, 3);`, }, vec3: { keyword: `vec3`, src: `_ = vec3(1, 2, 3);`, }, vec4_templated: { keyword: `vec4`, src: `_ = vec4(1, 2, 3, 4);`, }, vec4: { keyword: `vec4`, src: `_ = vec4(1, 2, 3, 4);`, }, }; g.test('shadow_hides_builtin') .desc(`Test that shadows hide builtins.`) .params(u => u .combine('inject', ['none', 'function', 'sibling', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kTests)) ) .fn(t => { const data = kTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : i32;` : ``; const sibling_func = t.params.inject === 'sibling' ? local : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` struct Data { rt_arr: array, } @group(0) @binding(0) var placeholder: Data; ${module_shadow} fn sibling() { ${sibling_func} } @fragment fn main() -> @location(0) vec4f { ${func} ${data.src} return vec4f(1); } `; const pass = t.params.inject === 'none' || t.params.inject === 'sibling'; t.expectCompileResult(pass, code); }); const kFloat16Tests = { f16: { keyword: `f16`, src: `_ = f16(2);`, }, }; g.test('shadow_hides_builtin_f16') .desc(`Test that shadows hide builtins when shader-f16 is enabled.`) .params(u => u .combine('inject', ['none', 'function', 'sibling', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kFloat16Tests)) ) .fn(t => { const data = kFloat16Tests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : f16;` : ``; const sibling_func = t.params.inject === 'sibling' ? local : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` enable f16; ${module_shadow} fn sibling() { ${sibling_func} } @vertex fn vtx() -> @builtin(position) vec4f { ${func} ${data.src} return vec4f(1); } `; const pass = t.params.inject === 'none' || t.params.inject === 'sibling'; t.expectCompileResult(pass, code); }); const kTextureTypeTests = { texture_1d: { keyword: `texture_1d`, src: `var t: texture_1d;`, }, texture_2d: { keyword: `texture_2d`, src: `var t: texture_2d;`, }, texture_2d_array: { keyword: `texture_2d_array`, src: `var t: texture_2d_array;`, }, texture_3d: { keyword: `texture_3d`, src: `var t: texture_3d;`, }, texture_cube: { keyword: `texture_cube`, src: `var t: texture_cube;`, }, texture_cube_array: { keyword: `texture_cube_array`, src: `var t: texture_cube_array;`, }, texture_multisampled_2d: { keyword: `texture_multisampled_2d`, src: `var t: texture_multisampled_2d;`, }, texture_depth_multisampled_2d: { keyword: `texture_depth_multisampled_2d`, src: `var t: texture_depth_multisampled_2d;`, }, texture_external: { keyword: `texture_external`, src: `var t: texture_external;`, }, texture_storage_1d: { keyword: `texture_storage_1d`, src: `var t: texture_storage_1d;`, }, texture_storage_2d: { keyword: `texture_storage_2d`, src: `var t: texture_storage_2d;`, }, texture_storage_2d_array: { keyword: `texture_storage_2d_array`, src: `var t: texture_storage_2d_array;`, }, texture_storage_3d: { keyword: `texture_storage_3d`, src: `var t: texture_storage_3d;`, }, texture_depth_2d: { keyword: `texture_depth_2d`, src: `var t: texture_depth_2d;`, }, texture_depth_2d_array: { keyword: `texture_depth_2d_array`, src: `var t: texture_depth_2d_array;`, }, texture_depth_cube: { keyword: `texture_depth_cube`, src: `var t: texture_depth_cube;`, }, texture_depth_cube_array: { keyword: `texture_depth_cube_array`, src: `var t: texture_depth_cube_array;`, }, sampler: { keyword: `sampler`, src: `var s: sampler;`, }, sampler_comparison: { keyword: `sampler_comparison`, src: `var s: sampler_comparison;`, }, }; g.test('shadow_hides_builtin_handle_type') .desc(`Test that shadows hide builtins when handle address space types are used.`) .params(u => u .combine('inject', ['none', 'function', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kTextureTypeTests)) ) .fn(t => { const data = kTextureTypeTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : f32;` : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` ${module_shadow} @group(0) @binding(0) ${data.src} fn func() { ${func} } `; const pass = t.params.inject === 'none' || t.params.inject === 'function'; t.expectCompileResult(pass, code); }); const kTextureTests = { textureDimensions: { keyword: `textureDimensions`, src: `_ = textureDimensions(t_2d);`, }, textureGather: { keyword: `textureGather`, src: `_ = textureGather(1, t_2d, s, vec2(1, 2));`, }, textureGatherCompare: { keyword: `textureGatherCompare`, src: `_ = textureGatherCompare(t_2d_depth, sc, vec2(1, 2), 3);`, }, textureLoad: { keyword: `textureLoad`, src: `_ = textureLoad(t_2d, vec2(1, 2), 1);`, }, textureNumLayers: { keyword: `textureNumLayers`, src: `_ = textureNumLayers(t_2d_array);`, }, textureNumLevels: { keyword: `textureNumLevels`, src: `_ = textureNumLevels(t_2d);`, }, textureNumSamples: { keyword: `textureNumSamples`, src: `_ = textureNumSamples(t_2d_ms);`, }, textureSample: { keyword: `textureSample`, src: `_ = textureSample(t_2d, s, vec2(1, 2));`, }, textureSampleBias: { keyword: `textureSampleBias`, src: `_ = textureSampleBias(t_2d, s, vec2(1, 2), 2);`, }, textureSampleCompare: { keyword: `textureSampleCompare`, src: `_ = textureSampleCompare(t_2d_depth, sc, vec2(1, 2), 2);`, }, textureSampleCompareLevel: { keyword: `textureSampleCompareLevel`, src: `_ = textureSampleCompareLevel(t_2d_depth, sc, vec2(1, 2), 3, vec2(1, 2));`, }, textureSampleGrad: { keyword: `textureSampleGrad`, src: `_ = textureSampleGrad(t_2d, s, vec2(1, 2), vec2(1, 2), vec2(1, 2));`, }, textureSampleLevel: { keyword: `textureSampleLevel`, src: `_ = textureSampleLevel(t_2d, s, vec2(1, 2), 3);`, }, textureSampleBaseClampToEdge: { keyword: `textureSampleBaseClampToEdge`, src: `_ = textureSampleBaseClampToEdge(t_2d, s, vec2(1, 2));`, }, }; g.test('shadow_hides_builtin_texture') .desc(`Test that shadows hide texture builtins.`) .params(u => u .combine('inject', ['none', 'function', 'sibling', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kTextureTests)) ) .fn(t => { const data = kTextureTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : i32;` : ``; const sibling_func = t.params.inject === 'sibling' ? local : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` @group(0) @binding(0) var t_2d: texture_2d; @group(0) @binding(1) var t_2d_depth: texture_depth_2d; @group(0) @binding(2) var t_2d_array: texture_2d_array; @group(0) @binding(3) var t_2d_ms: texture_multisampled_2d; @group(1) @binding(0) var s: sampler; @group(1) @binding(1) var sc: sampler_comparison; ${module_shadow} fn sibling() { ${sibling_func} } @fragment fn main() -> @location(0) vec4f { ${func} ${data.src} return vec4f(1); } `; const pass = t.params.inject === 'none' || t.params.inject === 'sibling'; t.expectCompileResult(pass, code); }); g.test('shadow_hides_builtin_atomic_type') .desc(`Test that shadows hide builtins when atomic types are used.`) .params(u => u.combine('inject', ['none', 'function', 'module'] as const).beginSubcases()) .fn(t => { const local = `let atomic = 4;`; const module_shadow = t.params.inject === 'module' ? `var atomic: i32;` : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` ${module_shadow} var val: atomic; fn func() { ${func} } `; const pass = t.params.inject === 'none' || t.params.inject === 'function'; t.expectCompileResult(pass, code); }); const kAtomicTests = { atomicLoad: { keyword: `atomicLoad`, src: `_ = atomicLoad(&a);`, }, atomicStore: { keyword: `atomicStore`, src: `atomicStore(&a, 1);`, }, atomicAdd: { keyword: `atomicAdd`, src: `_ = atomicAdd(&a, 1);`, }, atomicSub: { keyword: `atomicSub`, src: `_ = atomicSub(&a, 1);`, }, atomicMax: { keyword: `atomicMax`, src: `_ = atomicMax(&a, 1);`, }, atomicMin: { keyword: `atomicMin`, src: `_ = atomicMin(&a, 1);`, }, atomicAnd: { keyword: `atomicAnd`, src: `_ = atomicAnd(&a, 1);`, }, atomicOr: { keyword: `atomicOr`, src: `_ = atomicOr(&a, 1);`, }, atomicXor: { keyword: `atomicXor`, src: `_ = atomicXor(&a, 1);`, }, }; g.test('shadow_hides_builtin_atomic') .desc(`Test that shadows hide builtin atomic methods.`) .params(u => u .combine('inject', ['none', 'function', 'sibling', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kAtomicTests)) ) .fn(t => { const data = kAtomicTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : i32;` : ``; const sibling_func = t.params.inject === 'sibling' ? local : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` var a: atomic; ${module_shadow} fn sibling() { ${sibling_func} } @compute @workgroup_size(1) fn main() { ${func} ${data.src} } `; const pass = t.params.inject === 'none' || t.params.inject === 'sibling'; t.expectCompileResult(pass, code); }); const kBarrierTests = { storageBarrier: { keyword: `storageBarrier`, src: `storageBarrier();`, }, textureBarrier: { keyword: `textureBarrier`, src: `textureBarrier();`, }, workgroupBarrier: { keyword: `workgroupBarrier`, src: `workgroupBarrier();`, }, workgroupUniformLoad: { keyword: `workgroupUniformLoad`, src: `_ = workgroupUniformLoad(&u);`, }, }; g.test('shadow_hides_builtin_barriers') .desc(`Test that shadows hide builtin barrier methods.`) .params(u => u .combine('inject', ['none', 'function', 'sibling', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kBarrierTests)) ) .fn(t => { const data = kBarrierTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : i32;` : ``; const sibling_func = t.params.inject === 'sibling' ? local : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` var u: u32; ${module_shadow} fn sibling() { ${sibling_func} } @compute @workgroup_size(1) fn main() { ${func} ${data.src} } `; const pass = t.params.inject === 'none' || t.params.inject === 'sibling'; t.expectCompileResult(pass, code); }); const kAccessModeTests = { read: { keyword: `read`, src: `var a: i32;`, }, read_write: { keyword: `read_write`, src: `var a: i32;`, }, write: { keyword: `write`, src: `var t: texture_storage_1d;`, }, }; g.test('shadow_hides_access_mode') .desc(`Test that shadows hide access modes.`) .params(u => u .combine('inject', ['none', 'function', 'module'] as const) .beginSubcases() .combine('builtin', keysOf(kAccessModeTests)) ) .fn(t => { const data = kAccessModeTests[t.params.builtin]; const local = `let ${data.keyword} = 4;`; const module_shadow = t.params.inject === 'module' ? `var ${data.keyword} : i32;` : ``; const func = t.params.inject === 'function' ? local : ``; const code = ` ${module_shadow} @group(0) @binding(0) ${data.src} @compute @workgroup_size(1) fn main() { ${func} } `; const pass = t.params.inject === 'none' || t.params.inject === 'function'; t.expectCompileResult(pass, code); });