github/wgpu
1
1
Fork 0
mirror of https://github.com/gfx-rs/wgpu.git synced 2026-04-22 03:02:01 -04:00
Code Issues Packages Projects Releases Wiki Activity

Fix validation and glsl parsing of ldexp (#2449)

Browse Source 
Fixes #1908.
This commit is contained in:
Fredrik Fornwall
2023-09-01 18:44:48 +02:00
committed by GitHub
parent 0491d39232
commit f49314dbbd
10 changed files with 160 additions and 95 deletions
Download Patch File Download Diff File Expand all files Collapse all files

14
src/front/glsl/builtins.rs
View File

@@ -1280,14 +1280,14 @@ fn inject_common_builtin(
0b10 => Some(VectorSize::Tri),
_ => Some(VectorSize::Quad),
};
let ty = || match size {
let ty = |kind| match size {
Some(size) => TypeInner::Vector {
size,
kind: Sk::Float,
kind,
width: float_width,
},
None => TypeInner::Scalar {
kind: Sk::Float,
kind,
width: float_width,
},
};
@@ -1300,9 +1300,15 @@ fn inject_common_builtin(
_ => unreachable!(),
};
let second_kind = if fun == MacroCall::MathFunction(MathFunction::Ldexp) {
Sk::Sint
} else {
Sk::Float
};
declaration
.overloads
.push(module.add_builtin(vec![ty(), ty()], fun))
.push(module.add_builtin(vec![ty(Sk::Float), ty(second_kind)], fun))
}
}
"transpose" => {

37
src/valid/expression.rs
View File

@@ -1014,7 +1014,7 @@ impl super::Validator {
));
}
}
Mf::Modf | Mf::Frexp | Mf::Ldexp => {
Mf::Modf | Mf::Frexp => {
let arg1_ty = match (arg1_ty, arg2_ty, arg3_ty) {
(Some(ty1), None, None) => ty1,
_ => return Err(ExpressionError::WrongArgumentCount(fun)),
@@ -1049,6 +1049,41 @@ impl super::Validator {
));
}
}
Mf::Ldexp => {
let arg1_ty = match (arg1_ty, arg2_ty, arg3_ty) {
(Some(ty1), None, None) => ty1,
_ => return Err(ExpressionError::WrongArgumentCount(fun)),
};
let size0 = match *arg_ty {
Ti::Scalar {
kind: Sk::Float, ..
} => None,
Ti::Vector {
kind: Sk::Float,
size,
..
} => Some(size),
_ => {
return Err(ExpressionError::InvalidArgumentType(fun, 0, arg));
}
};
let good = match *arg1_ty {
Ti::Scalar { kind: Sk::Sint, .. } if size0.is_none() => true,
Ti::Vector {
size,
kind: Sk::Sint,
..
} if Some(size) == size0 => true,
_ => false,
};
if !good {
return Err(ExpressionError::InvalidArgumentType(
fun,
1,
arg1.unwrap(),
));
}
}
Mf::Dot => {
let arg1_ty = match (arg1_ty, arg2_ty, arg3_ty) {
(Some(ty1), None, None) => ty1,

2
tests/in/glsl/math-functions.frag
View File

@@ -50,7 +50,7 @@ void main() {
// TODO: support out params in wgsl output
// vec4 modfOut = modf(a, b);
// vec4 frexpOut = frexp(a, b);
// float ldexpOut = ldexp(a.x, i);
float ldexpOut = ldexp(a.x, i);
vec4 rad = radians(a);
float deg = degrees(a.x);
float smoothStepScalar = smoothstep(0.0, 1.0, 0.5);

2
tests/in/math-functions.wgsl
View File

@@ -29,4 +29,6 @@ fn main() {
let clz_b = countLeadingZeros(1u);
let clz_c = countLeadingZeros(vec2(-1));
let clz_d = countLeadingZeros(vec2(1u));
let lde_a = ldexp(1.0, 2);
let lde_b = ldexp(vec2(1.0, 2.0), vec2(3, 4));
}

2
tests/out/glsl/math-functions.main.Fragment.glsl
View File

@@ -34,5 +34,7 @@ void main() {
ivec2 _e58 = ivec2(-1);
ivec2 clz_c = mix(ivec2(31) - findMSB(_e58), ivec2(0), lessThan(_e58, ivec2(0)));
uvec2 clz_d = uvec2(ivec2(31) - findMSB(uvec2(1u)));
float lde_a = ldexp(1.0, 2);
vec2 lde_b = ldexp(vec2(1.0, 2.0), ivec2(3, 4));
}

2
tests/out/hlsl/math-functions.hlsl
View File

@@ -29,4 +29,6 @@ void main()
int2 _expr58 = (-1).xx;
int2 clz_c = (_expr58 < (0).xx ? (0).xx : (31).xx - asint(firstbithigh(_expr58)));
uint2 clz_d = ((31u).xx - firstbithigh((1u).xx));
float lde_a = ldexp(1.0, 2);
float2 lde_b = ldexp(float2(1.0, 2.0), int2(3, 4));
}

2
tests/out/msl/math-functions.msl
View File

@@ -38,4 +38,6 @@ fragment void main_(
uint clz_b = metal::clz(1u);
metal::int2 clz_c = metal::clz(metal::int2(-1));
metal::uint2 clz_d = metal::clz(metal::uint2(1u));
float lde_a = metal::ldexp(1.0, 2);
metal::float2 lde_b = metal::ldexp(metal::float2(1.0, 2.0), metal::int2(3, 4));
}

177
tests/out/spv/math-functions.spvasm
View File

@@ -1,97 +1,106 @@
; SPIR-V
; Version: 1.1
; Generator: rspirv
; Bound: 87
; Bound: 96
OpCapability Shader
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Fragment %8 "main"
OpExecutionMode %8 OriginUpperLeft
OpEntryPoint Fragment %9 "main"
OpExecutionMode %9 OriginUpperLeft
%2 = OpTypeVoid
%4 = OpTypeFloat 32
%3 = OpTypeVector %4 4
%6 = OpTypeInt 32 1
%5 = OpTypeVector %6 2
%9 = OpTypeFunction %2
%10 = OpConstant %4 1.0
%11 = OpConstant %4 0.0
%12 = OpConstantNull %5
%13 = OpTypeInt 32 0
%14 = OpConstant %13 0
%15 = OpConstant %6 -1
%16 = OpConstant %13 1
%17 = OpConstant %6 0
%18 = OpConstant %13 4294967295
%19 = OpConstant %6 1
%27 = OpConstantComposite %3 %11 %11 %11 %11
%28 = OpConstantComposite %3 %10 %10 %10 %10
%31 = OpConstantNull %6
%44 = OpTypeVector %13 2
%54 = OpConstant %13 32
%64 = OpConstantComposite %44 %54 %54
%76 = OpConstant %6 31
%82 = OpConstantComposite %5 %76 %76
%8 = OpFunction %2 None %9
%7 = OpLabel
OpBranch %20
%20 = OpLabel
%21 = OpCompositeConstruct %3 %11 %11 %11 %11
%22 = OpExtInst %4 %1 Degrees %10
%23 = OpExtInst %4 %1 Radians %10
%24 = OpExtInst %3 %1 Degrees %21
%25 = OpExtInst %3 %1 Radians %21
%26 = OpExtInst %3 %1 FClamp %21 %27 %28
%29 = OpExtInst %3 %1 Refract %21 %21 %10
%32 = OpCompositeExtract %6 %12 0
%33 = OpCompositeExtract %6 %12 0
%34 = OpIMul %6 %32 %33
%35 = OpIAdd %6 %31 %34
%36 = OpCompositeExtract %6 %12 1
%37 = OpCompositeExtract %6 %12 1
%38 = OpIMul %6 %36 %37
%30 = OpIAdd %6 %35 %38
%39 = OpCopyObject %13 %14
%40 = OpExtInst %13 %1 FindUMsb %39
%41 = OpExtInst %6 %1 FindSMsb %15
%42 = OpCompositeConstruct %5 %15 %15
%43 = OpExtInst %5 %1 FindSMsb %42
%45 = OpCompositeConstruct %44 %16 %16
%46 = OpExtInst %44 %1 FindUMsb %45
%47 = OpExtInst %6 %1 FindILsb %15
%48 = OpExtInst %13 %1 FindILsb %16
%49 = OpCompositeConstruct %5 %15 %15
%50 = OpExtInst %5 %1 FindILsb %49
%51 = OpCompositeConstruct %44 %16 %16
%52 = OpExtInst %44 %1 FindILsb %51
%55 = OpExtInst %13 %1 FindILsb %14
%53 = OpExtInst %13 %1 UMin %54 %55
%57 = OpExtInst %6 %1 FindILsb %17
%56 = OpExtInst %6 %1 UMin %54 %57
%59 = OpExtInst %13 %1 FindILsb %18
%58 = OpExtInst %13 %1 UMin %54 %59
%61 = OpExtInst %6 %1 FindILsb %15
%60 = OpExtInst %6 %1 UMin %54 %61
%62 = OpCompositeConstruct %44 %14 %14
%65 = OpExtInst %44 %1 FindILsb %62
%63 = OpExtInst %44 %1 UMin %64 %65
%66 = OpCompositeConstruct %5 %17 %17
%68 = OpExtInst %5 %1 FindILsb %66
%67 = OpExtInst %5 %1 UMin %64 %68
%69 = OpCompositeConstruct %44 %16 %16
%71 = OpExtInst %44 %1 FindILsb %69
%70 = OpExtInst %44 %1 UMin %64 %71
%72 = OpCompositeConstruct %5 %19 %19
%74 = OpExtInst %5 %1 FindILsb %72
%73 = OpExtInst %5 %1 UMin %64 %74
%77 = OpExtInst %6 %1 FindUMsb %15
%75 = OpISub %6 %76 %77
%79 = OpExtInst %6 %1 FindUMsb %16
%78 = OpISub %13 %76 %79
%80 = OpCompositeConstruct %5 %15 %15
%83 = OpExtInst %5 %1 FindUMsb %80
%81 = OpISub %5 %82 %83
%84 = OpCompositeConstruct %44 %16 %16
%86 = OpExtInst %5 %1 FindUMsb %84
%85 = OpISub %44 %82 %86
%7 = OpTypeVector %4 2
%10 = OpTypeFunction %2
%11 = OpConstant %4 1.0
%12 = OpConstant %4 0.0
%13 = OpConstantNull %5
%14 = OpTypeInt 32 0
%15 = OpConstant %14 0
%16 = OpConstant %6 -1
%17 = OpConstant %14 1
%18 = OpConstant %6 0
%19 = OpConstant %14 4294967295
%20 = OpConstant %6 1
%21 = OpConstant %6 2
%22 = OpConstant %4 2.0
%23 = OpConstant %6 3
%24 = OpConstant %6 4
%32 = OpConstantComposite %3 %12 %12 %12 %12
%33 = OpConstantComposite %3 %11 %11 %11 %11
%36 = OpConstantNull %6
%49 = OpTypeVector %14 2
%59 = OpConstant %14 32
%69 = OpConstantComposite %49 %59 %59
%81 = OpConstant %6 31
%87 = OpConstantComposite %5 %81 %81
%9 = OpFunction %2 None %10
%8 = OpLabel
OpBranch %25
%25 = OpLabel
%26 = OpCompositeConstruct %3 %12 %12 %12 %12
%27 = OpExtInst %4 %1 Degrees %11
%28 = OpExtInst %4 %1 Radians %11
%29 = OpExtInst %3 %1 Degrees %26
%30 = OpExtInst %3 %1 Radians %26
%31 = OpExtInst %3 %1 FClamp %26 %32 %33
%34 = OpExtInst %3 %1 Refract %26 %26 %11
%37 = OpCompositeExtract %6 %13 0
%38 = OpCompositeExtract %6 %13 0
%39 = OpIMul %6 %37 %38
%40 = OpIAdd %6 %36 %39
%41 = OpCompositeExtract %6 %13 1
%42 = OpCompositeExtract %6 %13 1
%43 = OpIMul %6 %41 %42
%35 = OpIAdd %6 %40 %43
%44 = OpCopyObject %14 %15
%45 = OpExtInst %14 %1 FindUMsb %44
%46 = OpExtInst %6 %1 FindSMsb %16
%47 = OpCompositeConstruct %5 %16 %16
%48 = OpExtInst %5 %1 FindSMsb %47
%50 = OpCompositeConstruct %49 %17 %17
%51 = OpExtInst %49 %1 FindUMsb %50
%52 = OpExtInst %6 %1 FindILsb %16
%53 = OpExtInst %14 %1 FindILsb %17
%54 = OpCompositeConstruct %5 %16 %16
%55 = OpExtInst %5 %1 FindILsb %54
%56 = OpCompositeConstruct %49 %17 %17
%57 = OpExtInst %49 %1 FindILsb %56
%60 = OpExtInst %14 %1 FindILsb %15
%58 = OpExtInst %14 %1 UMin %59 %60
%62 = OpExtInst %6 %1 FindILsb %18
%61 = OpExtInst %6 %1 UMin %59 %62
%64 = OpExtInst %14 %1 FindILsb %19
%63 = OpExtInst %14 %1 UMin %59 %64
%66 = OpExtInst %6 %1 FindILsb %16
%65 = OpExtInst %6 %1 UMin %59 %66
%67 = OpCompositeConstruct %49 %15 %15
%70 = OpExtInst %49 %1 FindILsb %67
%68 = OpExtInst %49 %1 UMin %69 %70
%71 = OpCompositeConstruct %5 %18 %18
%73 = OpExtInst %5 %1 FindILsb %71
%72 = OpExtInst %5 %1 UMin %69 %73
%74 = OpCompositeConstruct %49 %17 %17
%76 = OpExtInst %49 %1 FindILsb %74
%75 = OpExtInst %49 %1 UMin %69 %76
%77 = OpCompositeConstruct %5 %20 %20
%79 = OpExtInst %5 %1 FindILsb %77
%78 = OpExtInst %5 %1 UMin %69 %79
%82 = OpExtInst %6 %1 FindUMsb %16
%80 = OpISub %6 %81 %82
%84 = OpExtInst %6 %1 FindUMsb %17
%83 = OpISub %14 %81 %84
%85 = OpCompositeConstruct %5 %16 %16
%88 = OpExtInst %5 %1 FindUMsb %85
%86 = OpISub %5 %87 %88
%89 = OpCompositeConstruct %49 %17 %17
%91 = OpExtInst %5 %1 FindUMsb %89
%90 = OpISub %49 %87 %91
%92 = OpExtInst %4 %1 Ldexp %11 %21
%93 = OpCompositeConstruct %7 %11 %22
%94 = OpCompositeConstruct %5 %23 %24
%95 = OpExtInst %7 %1 Ldexp %93 %94
OpReturn
OpFunctionEnd

15
tests/out/wgsl/math-functions-frag.wgsl
View File

@@ -42,6 +42,7 @@ fn main_1() {
var outerProductOut: mat4x4<f32>;
var distanceOut: f32;
var stepOut: vec4<f32>;
var ldexpOut: f32;
var rad: vec4<f32>;
var deg: f32;
var smoothStepScalar: f32;
@@ -152,11 +153,15 @@ fn main_1() {
let _e225 = a;
let _e226 = b;
stepOut = step(_e225, _e226);
let _e230 = a;
rad = radians(_e230);
let _e233 = a;
let _e235 = a;
deg = degrees(_e235.x);
let _e229 = a;
let _e232 = a;
let _e234 = i;
ldexpOut = ldexp(_e232.x, _e234);
let _e238 = a;
rad = radians(_e238);
let _e241 = a;
let _e243 = a;
deg = degrees(_e243.x);
smoothStepScalar = smoothstep(0.0, 1.0, 0.5);
smoothStepVector = smoothstep(vec4<f32>(0.0), vec4<f32>(1.0), vec4<f32>(0.5));
smoothStepMixed = smoothstep(vec4<f32>(0.0), vec4<f32>(1.0), vec4<f32>(0.5));

2
tests/out/wgsl/math-functions.wgsl
View File

@@ -28,4 +28,6 @@ fn main() {
let clz_b = countLeadingZeros(1u);
let clz_c = countLeadingZeros(vec2<i32>(-1));
let clz_d = countLeadingZeros(vec2<u32>(1u));
let lde_a = ldexp(1.0, 2);
let lde_b = ldexp(vec2<f32>(1.0, 2.0), vec2<i32>(3, 4));
}