diff options
Diffstat (limited to '')
-rw-r--r-- | testing/web-platform/mozilla/tests/webgpu/webgpu/shader/validation/expression/binary/and_or_xor.spec.js | 182 |
1 files changed, 182 insertions, 0 deletions
diff --git a/testing/web-platform/mozilla/tests/webgpu/webgpu/shader/validation/expression/binary/and_or_xor.spec.js b/testing/web-platform/mozilla/tests/webgpu/webgpu/shader/validation/expression/binary/and_or_xor.spec.js new file mode 100644 index 0000000000..9acede3030 --- /dev/null +++ b/testing/web-platform/mozilla/tests/webgpu/webgpu/shader/validation/expression/binary/and_or_xor.spec.js @@ -0,0 +1,182 @@ +/** +* AUTO-GENERATED - DO NOT EDIT. Source: https://github.com/gpuweb/cts +**/export const description = ` +Validation tests for logical and bitwise and/or/xor expressions. +`;import { makeTestGroup } from '../../../../../common/framework/test_group.js'; +import { keysOf, objectsToRecord } from '../../../../../common/util/data_tables.js'; +import { + kAllScalarsAndVectors, + ScalarType, + scalarTypeOf, + Type, + VectorType } from +'../../../../util/conversion.js'; +import { ShaderValidationTest } from '../../shader_validation_test.js'; + +export const g = makeTestGroup(ShaderValidationTest); + +// A list of operators and a flag for whether they support boolean values or not. +const kOperators = { + and: { op: '&', supportsBool: true }, + or: { op: '|', supportsBool: true }, + xor: { op: '^', supportsBool: false } +}; + +// A list of scalar and vector types. +const kScalarAndVectorTypes = objectsToRecord(kAllScalarsAndVectors); + +g.test('scalar_vector'). +desc( + ` + Validates that scalar and vector expressions are only accepted for bool or compatible integer types. + ` +). +params((u) => +u. +combine('lhs', keysOf(kScalarAndVectorTypes)). +combine( + 'rhs', + // Skip vec3 and vec4 on the RHS to keep the number of subcases down. + // vec3 + vec3 and vec4 + vec4 is tested in execution tests. + keysOf(kScalarAndVectorTypes).filter( + (value) => !(value.startsWith('vec3') || value.startsWith('vec4')) + ) +). +beginSubcases(). +combine('op', keysOf(kOperators)) +). +beforeAllSubcases((t) => { + if ( + scalarTypeOf(kScalarAndVectorTypes[t.params.lhs]) === Type.f16 || + scalarTypeOf(kScalarAndVectorTypes[t.params.rhs]) === Type.f16) + { + t.selectDeviceOrSkipTestCase('shader-f16'); + } +}). +fn((t) => { + const op = kOperators[t.params.op]; + const lhs = kScalarAndVectorTypes[t.params.lhs]; + const rhs = kScalarAndVectorTypes[t.params.rhs]; + const lhsElement = scalarTypeOf(lhs); + const rhsElement = scalarTypeOf(rhs); + const hasF16 = lhsElement === Type.f16 || rhsElement === Type.f16; + const code = ` +${hasF16 ? 'enable f16;' : ''} +const lhs = ${lhs.create(0).wgsl()}; +const rhs = ${rhs.create(0).wgsl()}; +const foo = lhs ${op.op} rhs; +`; + + // Determine if the element types are compatible. + const kIntegerTypes = [Type.abstractInt, Type.i32, Type.u32]; + let elementIsCompatible = false; + if (lhsElement === Type.abstractInt) { + // Abstract integers are compatible with any other integer type. + elementIsCompatible = kIntegerTypes.includes(rhsElement); + } else if (rhsElement === Type.abstractInt) { + // Abstract integers are compatible with any other numeric type. + elementIsCompatible = kIntegerTypes.includes(lhsElement); + } else if (kIntegerTypes.includes(lhsElement)) { + // Concrete integers are only compatible with values with the exact same type. + elementIsCompatible = lhsElement === rhsElement; + } else if (lhsElement === Type.bool) { + // Booleans are only compatible with other booleans. + elementIsCompatible = rhsElement === Type.bool; + } + + // Determine if the full type is compatible. + let valid = false; + if (lhs instanceof ScalarType && rhs instanceof ScalarType) { + valid = elementIsCompatible; + } else if (lhs instanceof VectorType && rhs instanceof VectorType) { + // Vectors are only compatible with if the vector widths match. + valid = lhs.width === rhs.width && elementIsCompatible; + } + + if (lhsElement === Type.bool) { + valid &&= op.supportsBool; + } + + t.expectCompileResult(valid, code); +}); + + + + + + + +const kInvalidTypes = { + mat2x2f: { + expr: 'm', + control: (e) => `i32(${e}[0][0])` + }, + + array: { + expr: 'arr', + control: (e) => `${e}[0]` + }, + + ptr: { + expr: '(&u)', + control: (e) => `*${e}` + }, + + atomic: { + expr: 'a', + control: (e) => `atomicLoad(&${e})` + }, + + texture: { + expr: 't', + control: (e) => `i32(textureLoad(${e}, vec2(), 0).x)` + }, + + sampler: { + expr: 's', + control: (e) => `i32(textureSampleLevel(t, ${e}, vec2(), 0).x)` + }, + + struct: { + expr: 'str', + control: (e) => `${e}.u` + } +}; + +g.test('invalid_types'). +desc( + ` + Validates that expressions are never accepted for non-scalar and non-vector types. + ` +). +params((u) => +u. +combine('op', keysOf(kOperators)). +combine('type', keysOf(kInvalidTypes)). +combine('control', [true, false]). +beginSubcases() +). +fn((t) => { + const op = kOperators[t.params.op]; + const type = kInvalidTypes[t.params.type]; + const expr = t.params.control ? type.control(type.expr) : type.expr; + const code = ` +@group(0) @binding(0) var t : texture_2d<f32>; +@group(0) @binding(1) var s : sampler; +@group(0) @binding(2) var<storage, read_write> a : atomic<i32>; + +struct S { u : u32 } + +var<private> u : u32; +var<private> m : mat2x2f; +var<private> arr : array<i32, 4>; +var<private> str : S; + +@compute @workgroup_size(1) +fn main() { + let foo = ${expr} ${op.op} ${expr}; +} +`; + + t.expectCompileResult(t.params.control, code); +});
\ No newline at end of file |