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

Add support for the saturate function (#2025)

Browse Source 
uses clamp in place of saturate in spv and glsl
This commit is contained in:
Evan Mark Hopkins
2022-08-16 03:43:41 -07:00
committed by GitHub
parent 6f4003ca9b
commit a80967f860
16 changed files with 102 additions and 13 deletions
Download Patch File Download Diff File Expand all files Collapse all files

18
src/back/glsl/mod.rs
View File

@@ -2749,6 +2749,24 @@ impl<'a, W: Write> Writer<'a, W> {
Mf::Min => "min",
Mf::Max => "max",
Mf::Clamp => "clamp",
Mf::Saturate => {
write!(self.out, "clamp(")?;
self.write_expr(arg, ctx)?;
match *ctx.info[arg].ty.inner_with(&self.module.types) {
crate::TypeInner::Vector { size, .. } => write!(
self.out,
", vec{}(0.0), vec{0}(1.0)",
back::vector_size_str(size)
)?,
_ => write!(self.out, ", 0.0, 1.0")?,
}
write!(self.out, ")")?;
return Ok(());
}
// trigonometry
Mf::Cos => "cos",
Mf::Cosh => "cosh",

1
src/back/hlsl/writer.rs
View File

@@ -2433,6 +2433,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> {
Mf::Min => Function::Regular("min"),
Mf::Max => Function::Regular("max"),
Mf::Clamp => Function::Regular("clamp"),
Mf::Saturate => Function::Regular("saturate"),
// trigonometry
Mf::Cos => Function::Regular("cos"),
Mf::Cosh => Function::Regular("cosh"),

1
src/back/msl/writer.rs
View File

@@ -1628,6 +1628,7 @@ impl<W: Write> Writer<W> {
Mf::Min => "min",
Mf::Max => "max",
Mf::Clamp => "clamp",
Mf::Saturate => "saturate",
// trigonometry
Mf::Cos => "cos",
Mf::Cosh => "cosh",

55
src/back/spv/block.rs
View File

@@ -696,6 +696,61 @@ impl<'w> BlockContext<'w> {
Some(crate::ScalarKind::Uint) => spirv::GLOp::UClamp,
other => unimplemented!("Unexpected max({:?})", other),
}),
Mf::Saturate => {
let (maybe_size, width) = match *arg_ty {
crate::TypeInner::Vector { size, width, .. } => (Some(size), width),
crate::TypeInner::Scalar { width, .. } => (None, width),
ref other => unimplemented!("Unexpected saturate({:?})", other),
};
let mut arg1_id = self
.writer
.get_constant_scalar(crate::ScalarValue::Float(0.0), width);
let mut arg2_id = self
.writer
.get_constant_scalar(crate::ScalarValue::Float(1.0), width);
if let Some(size) = maybe_size {
let value = LocalType::Value {
vector_size: Some(size),
kind: crate::ScalarKind::Float,
width,
pointer_space: None,
};
let result_type_id = self.get_type_id(LookupType::Local(value));
self.temp_list.clear();
self.temp_list.resize(size as _, arg1_id);
let id = self.gen_id();
block.body.push(Instruction::composite_construct(
result_type_id,
id,
&self.temp_list,
));
arg1_id = id;
self.temp_list.clear();
self.temp_list.resize(size as _, arg2_id);
let id = self.gen_id();
block.body.push(Instruction::composite_construct(
result_type_id,
id,
&self.temp_list,
));
arg2_id = id;
}
MathOp::Custom(Instruction::ext_inst(
self.writer.gl450_ext_inst_id,
spirv::GLOp::FClamp,
result_type_id,
id,
&[arg0_id, arg1_id, arg2_id],
))
}
// trigonometry
Mf::Sin => MathOp::Ext(spirv::GLOp::Sin),
Mf::Sinh => MathOp::Ext(spirv::GLOp::Sinh),

1
src/back/wgsl/writer.rs
View File

@@ -1511,6 +1511,7 @@ impl<W: Write> Writer<W> {
Mf::Min => Function::Regular("min"),
Mf::Max => Function::Regular("max"),
Mf::Clamp => Function::Regular("clamp"),
Mf::Saturate => Function::Regular("saturate"),
// trigonometry
Mf::Cos => Function::Regular("cos"),
Mf::Cosh => Function::Regular("cosh"),

1
src/front/wgsl/conv.rs
View File

@@ -140,6 +140,7 @@ pub fn map_standard_fun(word: &str) -> Option<crate::MathFunction> {
"min" => Mf::Min,
"max" => Mf::Max,
"clamp" => Mf::Clamp,
"saturate" => Mf::Saturate,
// trigonometry
"cos" => Mf::Cos,
"cosh" => Mf::Cosh,

1
src/lib.rs
View File

@@ -952,6 +952,7 @@ pub enum MathFunction {
Min,
Max,
Clamp,
Saturate,
// trigonometry
Cos,
Cosh,

1
src/proc/mod.rs
View File

@@ -228,6 +228,7 @@ impl super::MathFunction {
Self::Min => 2,
Self::Max => 2,
Self::Clamp => 3,
Self::Saturate => 1,
// trigonometry
Self::Cos => 1,
Self::Cosh => 1,

1
src/proc/typifier.rs
View File

@@ -704,6 +704,7 @@ impl<'a> ResolveContext<'a> {
Mf::Min |
Mf::Max |
Mf::Clamp |
Mf::Saturate |
// trigonometry
Mf::Cos |
Mf::Cosh |

3
src/valid/expression.rs
View File

@@ -1040,7 +1040,8 @@ impl super::Validator {
));
}
}
Mf::Cos
Mf::Saturate
| Mf::Cos
| Mf::Cosh
| Mf::Sin
| Mf::Sinh

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

@@ -6,6 +6,7 @@ fn main() {
let b = radians(f);
let c = degrees(v);
let d = radians(v);
let e = saturate(v);
let const_dot = dot(vec2<i32>(), vec2<i32>());
let first_leading_bit_abs = firstLeadingBit(abs(0u));
}

1
tests/out/glsl/math-functions.main.Vertex.glsl
View File

@@ -10,6 +10,7 @@ void main() {
float b = radians(1.0);
vec4 c = degrees(v);
vec4 d = radians(v);
vec4 e = clamp(v, vec4(0.0), vec4(1.0));
int const_dot = ( + ivec2(0, 0).x * ivec2(0, 0).x + ivec2(0, 0).y * ivec2(0, 0).y);
uint first_leading_bit_abs = uint(findMSB(uint(abs(int(0u)))));
}

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

@@ -6,6 +6,7 @@ void main()
float b = radians(1.0);
float4 c = degrees(v);
float4 d = radians(v);
float4 e = saturate(v);
int const_dot = dot(int2(0, 0), int2(0, 0));
uint first_leading_bit_abs = firstbithigh(abs(0u));
}

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

@@ -13,6 +13,7 @@ vertex void main_(
float b = ((1.0) * 0.017453292519943295474);
metal::float4 c = ((v) * 57.295779513082322865);
metal::float4 d = ((v) * 0.017453292519943295474);
metal::float4 e = metal::saturate(v);
int const_dot = ( + const_type.x * const_type.x + const_type.y * const_type.y);
uint first_leading_bit_abs = (((metal::clz(metal::abs(0u)) + 1) % 33) - 1);
}

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

@@ -1,7 +1,7 @@
; SPIR-V
; Version: 1.1
; Generator: rspirv
; Bound: 33
; Bound: 36
OpCapability Shader
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
@@ -18,7 +18,7 @@ OpEntryPoint Vertex %13 "main"
%11 = OpConstantComposite %10 %6 %6
%14 = OpTypeFunction %2
%16 = OpTypeVector %4 4
%23 = OpConstantNull %7
%26 = OpConstantNull %7
%13 = OpFunction %2 None %14
%12 = OpLabel
OpBranch %15
@@ -28,15 +28,18 @@ OpBranch %15
%19 = OpExtInst %4 %1 Radians %3
%20 = OpExtInst %16 %1 Degrees %17
%21 = OpExtInst %16 %1 Radians %17
%24 = OpCompositeExtract %7 %11 0
%25 = OpCompositeExtract %7 %11 0
%26 = OpIMul %7 %24 %25
%27 = OpIAdd %7 %23 %26
%28 = OpCompositeExtract %7 %11 1
%29 = OpCompositeExtract %7 %11 1
%30 = OpIMul %7 %28 %29
%22 = OpIAdd %7 %27 %30
%31 = OpCopyObject %9 %8
%32 = OpExtInst %9 %1 FindUMsb %31
%23 = OpCompositeConstruct %16 %5 %5 %5 %5
%24 = OpCompositeConstruct %16 %3 %3 %3 %3
%22 = OpExtInst %16 %1 FClamp %17 %23 %24
%27 = OpCompositeExtract %7 %11 0
%28 = OpCompositeExtract %7 %11 0
%29 = OpIMul %7 %27 %28
%30 = OpIAdd %7 %26 %29
%31 = OpCompositeExtract %7 %11 1
%32 = OpCompositeExtract %7 %11 1
%33 = OpIMul %7 %31 %32
%25 = OpIAdd %7 %30 %33
%34 = OpCopyObject %9 %8
%35 = OpExtInst %9 %1 FindUMsb %34
OpReturn
OpFunctionEnd

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

@@ -5,6 +5,7 @@ fn main() {
let b = radians(1.0);
let c = degrees(v);
let d = radians(v);
let e = saturate(v);
let const_dot = dot(vec2<i32>(0, 0), vec2<i32>(0, 0));
let first_leading_bit_abs = firstLeadingBit(abs(0u));
}