nordic-dev.net/wgpu
nordic-dev.net/wgpu
2
0
Fork 0
mirror of https://github.com/gfx-rs/wgpu.git synced 2024-11-30 02:33:38 +00:00
Code Issues Packages Projects Releases Wiki Activity

Tidy up constants IR

Browse Source
This commit is contained in:
Dzmitry Malyshau 2020-12-20 00:49:55 -05:00 committed by Timo de Kort
parent 1e70432691
commit 001336b457
16 changed files with 318 additions and 249 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

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

@ -50,8 +50,9 @@ use crate::{
},
Arena, ArraySize, BinaryOperator, BuiltIn, Bytes, ConservativeDepth, Constant, ConstantInner,
DerivativeAxis, Expression, FastHashMap, Function, GlobalVariable, Handle, ImageClass,
Interpolation, LocalVariable, Module, RelationalFunction, ScalarKind, ShaderStage, Statement,
StorageAccess, StorageClass, StorageFormat, StructMember, Type, TypeInner, UnaryOperator,
Interpolation, LocalVariable, Module, RelationalFunction, ScalarKind, ScalarValue, ShaderStage,
Statement, StorageAccess, StorageClass, StorageFormat, StructMember, Type, TypeInner,
UnaryOperator,
};
use features::FeaturesManager;
use std::{
@ -561,7 +562,10 @@ impl<'a, W: Write> Writer<'a, W> {
match size {
ArraySize::Constant(const_handle) => {
match self.module.constants[const_handle].inner {
ConstantInner::Uint(size) => write!(self.out, "{}", size)?,
ConstantInner::Scalar {
width: _,
value: ScalarValue::Uint(size),
} => write!(self.out, "{}", size)?,
_ => unreachable!(),
}
}
@ -866,22 +870,27 @@ impl<'a, W: Write> Writer<'a, W> {
/// Adds no newlines or leading/trailing whitespaces
fn write_constant(&mut self, constant: &Constant) -> BackendResult {
match constant.inner {
// Signed integers don't need anything special
ConstantInner::Sint(int) => write!(self.out, "{}", int)?,
// Unsigned integers need a `u` at the end
//
// While `core` doesn't necessarily need it, it's allowed and since `es` needs it we
// always write it as the extra branch wouldn't have any benefit in readability
ConstantInner::Uint(int) => write!(self.out, "{}u", int)?,
// Floats are written using `Debug` insted of `Display` because it always appends the
// decimal part even it's zero which is needed for a valid glsl float constant
ConstantInner::Float(float) => write!(self.out, "{:?}", float)?,
// Booleans are either `true` or `false` so nothing special needs to be done
ConstantInner::Bool(boolean) => write!(self.out, "{}", boolean)?,
ConstantInner::Scalar {
width: _,
ref value,
} => match *value {
// Signed integers don't need anything special
ScalarValue::Sint(int) => write!(self.out, "{}", int)?,
// Unsigned integers need a `u` at the end
//
// While `core` doesn't necessarily need it, it's allowed and since `es` needs it we
// always write it as the extra branch wouldn't have any benefit in readability
ScalarValue::Uint(int) => write!(self.out, "{}u", int)?,
// Floats are written using `Debug` insted of `Display` because it always appends the
// decimal part even it's zero which is needed for a valid glsl float constant
ScalarValue::Float(float) => write!(self.out, "{:?}", float)?,
// Booleans are either `true` or `false` so nothing special needs to be done
ScalarValue::Bool(boolean) => write!(self.out, "{}", boolean)?,
},
// Composite constant are created using the same syntax as compose
// `type(components)` where `components` is a comma separated list of constants
ConstantInner::Composite(ref components) => {
self.write_type(constant.ty)?;
ConstantInner::Composite { ty, ref components } => {
self.write_type(ty)?;
write!(self.out, "(")?;
// Write the comma separated constants

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

@ -527,25 +527,30 @@ impl<W: Write> Writer<W> {
) -> Result<(), Error> {
let constant = &module.constants[handle];
match constant.inner {
crate::ConstantInner::Sint(value) => {
write!(self.out, "{}", value)?;
}
crate::ConstantInner::Uint(value) => {
write!(self.out, "{}", value)?;
}
crate::ConstantInner::Float(value) => {
write!(self.out, "{}", value)?;
if value.fract() == 0.0 {
write!(self.out, ".0")?;
crate::ConstantInner::Scalar {
width: _,
ref value,
} => match *value {
crate::ScalarValue::Sint(value) => {
write!(self.out, "{}", value)?;
}
}
crate::ConstantInner::Bool(value) => {
write!(self.out, "{}", value)?;
}
crate::ConstantInner::Composite(ref constituents) => {
let ty_name = &self.names[&NameKey::Type(constant.ty)];
crate::ScalarValue::Uint(value) => {
write!(self.out, "{}", value)?;
}
crate::ScalarValue::Float(value) => {
write!(self.out, "{}", value)?;
if value.fract() == 0.0 {
write!(self.out, ".0")?;
}
}
crate::ScalarValue::Bool(value) => {
write!(self.out, "{}", value)?;
}
},
crate::ConstantInner::Composite { ty, ref components } => {
let ty_name = &self.names[&NameKey::Type(ty)];
write!(self.out, "{}(", ty_name)?;
for (i, &handle) in constituents.iter().enumerate() {
for (i, &handle) in components.iter().enumerate() {
if i != 0 {
write!(self.out, ", ")?;
}

2
src/back/spv/layout.rs
View File

@ -1,4 +1,4 @@
use crate::back::spv::{Instruction, LogicalLayout, PhysicalLayout};
use super::{Instruction, LogicalLayout, PhysicalLayout};
use spirv::*;
use std::iter;

138
src/back/spv/writer.rs
View File

@ -695,76 +695,75 @@ impl Writer {
let arena = &ir_module.types;
let instruction = match constant.inner {
crate::ConstantInner::Sint(val) => {
let ty = &ir_module.types[constant.ty];
let type_id = self.get_type_id(arena, LookupType::Handle(constant.ty))?;
match ty.inner {
crate::TypeInner::Scalar { kind: _, width } => match width {
4 => super::instructions::instruction_constant(type_id, id, &[val as u32]),
8 => {
let (low, high) = ((val >> 32) as u32, val as u32);
super::instructions::instruction_constant(type_id, id, &[low, high])
}
_ => unreachable!(),
},
_ => unreachable!(),
crate::ConstantInner::Scalar { width, ref value } => {
let type_id = self.get_type_id(
arena,
LookupType::Local(LocalType::Scalar {
kind: value.scalar_kind(),
width,
}),
)?;
let (solo, pair);
match *value {
crate::ScalarValue::Sint(val) => {
let words = match width {
4 => {
solo = [val as u32];
&solo[..]
}
8 => {
pair = [(val >> 32) as u32, val as u32];
&pair
}
_ => unreachable!(),
};
super::instructions::instruction_constant(type_id, id, words)
}
crate::ScalarValue::Uint(val) => {
let words = match width {
4 => {
solo = [val as u32];
&solo[..]
}
8 => {
pair = [(val >> 32) as u32, val as u32];
&pair
}
_ => unreachable!(),
};
super::instructions::instruction_constant(type_id, id, words)
}
crate::ScalarValue::Float(val) => {
let words = match width {
4 => {
solo = [(val as f32).to_bits()];
&solo[..]
}
8 => {
let bits = f64::to_bits(val);
pair = [(bits >> 32) as u32, bits as u32];
&pair
}
_ => unreachable!(),
};
super::instructions::instruction_constant(type_id, id, words)
}
crate::ScalarValue::Bool(true) => {
super::instructions::instruction_constant_true(type_id, id)
}
crate::ScalarValue::Bool(false) => {
super::instructions::instruction_constant_false(type_id, id)
}
}
}
crate::ConstantInner::Uint(val) => {
let ty = &ir_module.types[constant.ty];
let type_id = self.get_type_id(arena, LookupType::Handle(constant.ty))?;
match ty.inner {
crate::TypeInner::Scalar { kind: _, width } => match width {
4 => super::instructions::instruction_constant(type_id, id, &[val as u32]),
8 => {
let (low, high) = ((val >> 32) as u32, val as u32);
super::instructions::instruction_constant(type_id, id, &[low, high])
}
_ => unreachable!(),
},
_ => unreachable!(),
}
}
crate::ConstantInner::Float(val) => {
let ty = &ir_module.types[constant.ty];
let type_id = self.get_type_id(arena, LookupType::Handle(constant.ty))?;
match ty.inner {
crate::TypeInner::Scalar { kind: _, width } => match width {
4 => super::instructions::instruction_constant(
type_id,
id,
&[(val as f32).to_bits()],
),
8 => {
let bits = f64::to_bits(val);
let (low, high) = ((bits >> 32) as u32, bits as u32);
super::instructions::instruction_constant(type_id, id, &[low, high])
}
_ => unreachable!(),
},
_ => unreachable!(),
}
}
crate::ConstantInner::Bool(val) => {
let type_id = self.get_type_id(arena, LookupType::Handle(constant.ty))?;
if val {
super::instructions::instruction_constant_true(type_id, id)
} else {
super::instructions::instruction_constant_false(type_id, id)
}
}
crate::ConstantInner::Composite(ref constituents) => {
let mut constituent_ids = Vec::with_capacity(constituents.len());
for constituent in constituents.iter() {
crate::ConstantInner::Composite { ty, ref components } => {
let mut constituent_ids = Vec::with_capacity(components.len());
for constituent in components.iter() {
let constituent_id = self.get_constant_id(*constituent, &ir_module)?;
constituent_ids.push(constituent_id);
}
let type_id = self.get_type_id(arena, LookupType::Handle(constant.ty))?;
let type_id = self.get_type_id(arena, LookupType::Handle(ty))?;
super::instructions::instruction_constant_composite(
type_id,
id,
@ -1115,7 +1114,16 @@ impl Writer {
crate::Expression::Constant(handle) => {
let var = &ir_module.constants[handle];
let id = self.get_constant_id(handle, ir_module)?;
Ok((RawExpression::Value(id), LookupType::Handle(var.ty)))
let lookup_type = match var.inner {
crate::ConstantInner::Scalar { width, ref value } => {
LookupType::Local(LocalType::Scalar {
kind: value.scalar_kind(),
width,
})
}
crate::ConstantInner::Composite { ty, components: _ } => LookupType::Handle(ty),
};
Ok((RawExpression::Value(id), lookup_type))
}
crate::Expression::Compose { ty, ref components } => {
let base_type_id = self.get_type_id(&ir_module.types, LookupType::Handle(ty))?;

41
src/front/glsl/parser.rs
View File

@ -10,7 +10,7 @@ pomelo! {
Arena, BinaryOperator, Binding, Block, Constant,
ConstantInner, EntryPoint, Expression,
Function, GlobalVariable, Handle, Interpolation,
LocalVariable, SampleLevel, ScalarKind,
LocalVariable, SampleLevel, ScalarValue,
Statement, StorageAccess, StorageClass, StructMember,
SwitchCase, Type, TypeInner, UnaryOperator,
};
@ -165,18 +165,13 @@ pomelo! {
primary_expression ::= variable_identifier;
primary_expression ::= IntConstant(i) {
let ty = extra.module.types.fetch_or_append(Type {
name: None,
inner: TypeInner::Scalar {
kind: ScalarKind::Sint,
width: 4,
}
});
let ch = extra.module.constants.fetch_or_append(Constant {
name: None,
specialization: None,
ty,
inner: ConstantInner::Sint(i.1)
inner: ConstantInner::Scalar {
width: 4,
value: ScalarValue::Sint(i.1),
},
});
ExpressionRule::from_expression(
extra.context.expressions.append(Expression::Constant(ch))
@ -184,36 +179,26 @@ pomelo! {
}
// primary_expression ::= UintConstant;
primary_expression ::= FloatConstant(f) {
let ty = extra.module.types.fetch_or_append(Type {
name: None,
inner: TypeInner::Scalar {
kind: ScalarKind::Float,
width: 4,
}
});
let ch = extra.module.constants.fetch_or_append(Constant {
name: None,
specialization: None,
ty,
inner: ConstantInner::Float(f.1 as f64)
inner: ConstantInner::Scalar {
width: 4,
value: ScalarValue::Float(f.1 as f64),
},
});
ExpressionRule::from_expression(
extra.context.expressions.append(Expression::Constant(ch))
)
}
primary_expression ::= BoolConstant(b) {
let ty = extra.module.types.fetch_or_append(Type {
name: None,
inner: TypeInner::Scalar {
kind: ScalarKind::Bool,
width: 4,
}
});
let ch = extra.module.constants.fetch_or_append(Constant {
name: None,
specialization: None,
ty,
inner: ConstantInner::Bool(b.1)
inner: ConstantInner::Scalar {
width: 1,
value: ScalarValue::Bool(b.1)
},
});
ExpressionRule::from_expression(
extra.context.expressions.append(Expression::Constant(ch))

43
src/front/spv/mod.rs
View File

@ -624,8 +624,14 @@ impl<I: Iterator<Item = u32>> Parser<I> {
let index = match expressions[index_expr.handle] {
crate::Expression::Constant(const_handle) => {
match const_arena[const_handle].inner {
crate::ConstantInner::Uint(v) => v as u32,
crate::ConstantInner::Sint(v) => v as u32,
crate::ConstantInner::Scalar {
width: 4,
value: crate::ScalarValue::Uint(v),
} => v as u32,
crate::ConstantInner::Scalar {
width: 4,
value: crate::ScalarValue::Sint(v),
} => v as u32,
_ => {
return Err(Error::InvalidAccess(index_expr.handle))
}
@ -2168,7 +2174,10 @@ impl<I: Iterator<Item = u32>> Parser<I> {
} else {
0
};
crate::ConstantInner::Uint((u64::from(high) << 32) | u64::from(low))
crate::ConstantInner::Scalar {
width,
value: crate::ScalarValue::Uint((u64::from(high) << 32) | u64::from(low)),
}
}
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Sint,
@ -2184,7 +2193,12 @@ impl<I: Iterator<Item = u32>> Parser<I> {
}
Ordering::Equal => 0,
};
crate::ConstantInner::Sint(((u64::from(high) << 32) | u64::from(low)) as i64)
crate::ConstantInner::Scalar {
width,
value: crate::ScalarValue::Sint(
((u64::from(high) << 32) | u64::from(low)) as i64,
),
}
}
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Float,
@ -2200,7 +2214,10 @@ impl<I: Iterator<Item = u32>> Parser<I> {
}
_ => return Err(Error::InvalidTypeWidth(u32::from(width))),
};
crate::ConstantInner::Float(extended)
crate::ConstantInner::Scalar {
width,
value: crate::ScalarValue::Float(extended),
}
}
_ => return Err(Error::UnsupportedType(type_lookup.handle)),
};
@ -2211,7 +2228,6 @@ impl<I: Iterator<Item = u32>> Parser<I> {
name: self.future_decor.remove(&id).and_then(|dec| dec.name),
specialization: None, //TODO
inner,
ty,
}),
type_id,
},
@ -2229,15 +2245,13 @@ impl<I: Iterator<Item = u32>> Parser<I> {
let type_id = self.next()?;
let type_lookup = self.lookup_type.lookup(type_id)?;
let ty = type_lookup.handle;
let id = self.next()?;
let constituents_count = inst.wc - 3;
let mut constituents = Vec::with_capacity(constituents_count as usize);
for _ in 0..constituents_count {
let constituent_id = self.next()?;
let constant = self.lookup_constant.lookup(constituent_id)?;
constituents.push(constant.handle);
let mut components = Vec::with_capacity(inst.wc as usize - 3);
for _ in 0..components.capacity() {
let component_id = self.next()?;
let constant = self.lookup_constant.lookup(component_id)?;
components.push(constant.handle);
}
self.lookup_constant.insert(
@ -2246,8 +2260,7 @@ impl<I: Iterator<Item = u32>> Parser<I> {
handle: module.constants.append(crate::Constant {
name: self.future_decor.remove(&id).and_then(|dec| dec.name),
specialization: None,
inner: crate::ConstantInner::Composite(constituents),
ty,
inner: crate::ConstantInner::Composite { ty, components },
}),
type_id,
},

74
src/front/wgsl/mod.rs
View File

@ -369,16 +369,14 @@ impl Parser {
}
}
fn get_constant_inner(
word: &str,
) -> Result<(crate::ConstantInner, crate::ScalarKind), Error<'_>> {
fn get_scalar_value(word: &str) -> Result<crate::ScalarValue, Error<'_>> {
if word.contains('.') {
word.parse()
.map(|f| (crate::ConstantInner::Float(f), crate::ScalarKind::Float))
.map(crate::ScalarValue::Float)
.map_err(|err| Error::BadFloat(word, err))
} else {
word.parse()
.map(|i| (crate::ConstantInner::Sint(i), crate::ScalarKind::Sint))
.map(crate::ScalarValue::Sint)
.map_err(|err| Error::BadInteger(word, err))
}
}
@ -688,16 +686,24 @@ impl Parser {
let inner = match lexer.peek() {
Token::Word("true") => {
let _ = lexer.next();
crate::ConstantInner::Bool(true)
crate::ConstantInner::Scalar {
width: 1,
value: crate::ScalarValue::Bool(true),
}
}
Token::Word("false") => {
let _ = lexer.next();
crate::ConstantInner::Bool(false)
crate::ConstantInner::Scalar {
width: 1,
value: crate::ScalarValue::Bool(false),
}
}
Token::Number(word) => {
let _ = lexer.next();
let (inner, _) = Self::get_constant_inner(word)?;
inner
crate::ConstantInner::Scalar {
width: 4,
value: Self::get_scalar_value(word)?,
}
}
_ => {
let (composite_ty, _access) =
@ -717,14 +723,16 @@ impl Parser {
self.parse_const_expression(lexer, ty, type_arena, const_arena)?;
components.push(component);
}
crate::ConstantInner::Composite(components)
crate::ConstantInner::Composite {
ty: self_ty,
components,
}
}
};
let handle = const_arena.fetch_or_append(crate::Constant {
name: None,
specialization: None,
inner,
ty: self_ty,
});
self.scopes.pop();
Ok(handle)
@ -747,14 +755,10 @@ impl Parser {
let handle = ctx.constants.fetch_or_append(crate::Constant {
name: None,
specialization: None,
inner: crate::ConstantInner::Bool(true),
ty: ctx.types.fetch_or_append(crate::Type {
name: None,
inner: crate::TypeInner::Scalar {
kind: crate::ScalarKind::Bool,
width: 1,
},
}),
inner: crate::ConstantInner::Scalar {
width: 1,
value: crate::ScalarValue::Bool(true),
},
});
crate::Expression::Constant(handle)
}
@ -762,27 +766,19 @@ impl Parser {
let handle = ctx.constants.fetch_or_append(crate::Constant {
name: None,
specialization: None,
inner: crate::ConstantInner::Bool(false),
ty: ctx.types.fetch_or_append(crate::Type {
name: None,
inner: crate::TypeInner::Scalar {
kind: crate::ScalarKind::Bool,
width: 1,
},
}),
inner: crate::ConstantInner::Scalar {
width: 1,
value: crate::ScalarValue::Bool(false),
},
});
crate::Expression::Constant(handle)
}
Token::Number(word) => {
let (inner, kind) = Self::get_constant_inner(word)?;
let value = Self::get_scalar_value(word)?;
let handle = ctx.constants.fetch_or_append(crate::Constant {
name: None,
specialization: None,
inner,
ty: ctx.types.fetch_or_append(crate::Type {
name: None,
inner: crate::TypeInner::Scalar { kind, width: 4 },
}),
inner: crate::ConstantInner::Scalar { width: 4, value },
});
crate::Expression::Constant(handle)
}
@ -1304,14 +1300,10 @@ impl Parser {
let const_handle = const_arena.fetch_or_append(crate::Constant {
name: None,
specialization: None,
inner: crate::ConstantInner::Uint(value as u64),
ty: type_arena.fetch_or_append(crate::Type {
name: None,
inner: crate::TypeInner::Scalar {
kind: crate::ScalarKind::Uint,
width: 4,
},
}),
inner: crate::ConstantInner::Scalar {
width: 4,
value: crate::ScalarValue::Uint(value as u64),
},
});
crate::ArraySize::Constant(const_handle)
}

25
src/lib.rs
View File

@ -396,7 +396,17 @@ pub struct Constant {
pub name: Option<String>,
pub specialization: Option<u32>,
pub inner: ConstantInner,
pub ty: Handle<Type>,
}
/// A literal scalar value, used in constants.
#[derive(Debug, PartialEq)]
#[cfg_attr(feature = "serialize", derive(Serialize))]
#[cfg_attr(feature = "deserialize", derive(Deserialize))]
pub enum ScalarValue {
Sint(i64),
Uint(u64),
Float(f64),
Bool(bool),
}
/// Additional information, dependendent on the kind of constant.
@ -404,11 +414,14 @@ pub struct Constant {
#[cfg_attr(feature = "serialize", derive(Serialize))]
#[cfg_attr(feature = "deserialize", derive(Deserialize))]
pub enum ConstantInner {
Sint(i64),
Uint(u64),
Float(f64),
Bool(bool),
Composite(Vec<Handle<Constant>>),
Scalar {
width: Bytes,
value: ScalarValue,
},
Composite {
ty: Handle<Type>,
components: Vec<Handle<Constant>>,
},
}
/// Describes how an input/output variable is to be bound.

13
src/proc/mod.rs
View File

@ -15,7 +15,7 @@ pub use interface::{Interface, Visitor};
pub use namer::{EntryPointIndex, NameKey, Namer};
pub use sizer::Sizer;
pub use terminator::ensure_block_returns;
pub use typifier::{check_constant_type, ResolveContext, ResolveError, Typifier};
pub use typifier::{ResolveContext, ResolveError, Typifier};
pub use validator::{ValidationError, Validator};
impl From<super::StorageFormat> for super::ScalarKind {
@ -58,6 +58,17 @@ impl From<super::StorageFormat> for super::ScalarKind {
}
}
impl crate::ScalarValue {
pub fn scalar_kind(&self) -> crate::ScalarKind {
match *self {
Self::Uint(_) => crate::ScalarKind::Uint,
Self::Sint(_) => crate::ScalarKind::Sint,
Self::Float(_) => crate::ScalarKind::Float,
Self::Bool(_) => crate::ScalarKind::Bool,
}
}
}
impl crate::TypeInner {
pub fn scalar_kind(&self) -> Option<super::ScalarKind> {
match *self {

5
src/proc/sizer.rs
View File

@ -28,7 +28,10 @@ impl Sizer {
Ti::Array { base, size, stride } => {
let count = match size {
crate::ArraySize::Constant(handle) => match constants[handle].inner {
crate::ConstantInner::Uint(value) => value as u32,
crate::ConstantInner::Scalar {
width: _,
value: crate::ScalarValue::Uint(value),
} => value as u32,
ref other => unreachable!("Unexpected array size {:?}", other),
},
crate::ArraySize::Dynamic => 1,

45
src/proc/typifier.rs
View File

@ -170,7 +170,15 @@ impl Typifier {
})
}
},
crate::Expression::Constant(h) => Resolution::Handle(ctx.constants[h].ty),
crate::Expression::Constant(h) => match ctx.constants[h].inner {
crate::ConstantInner::Scalar { width, ref value } => {
Resolution::Value(crate::TypeInner::Scalar {
kind: value.scalar_kind(),
width,
})
}
crate::ConstantInner::Composite { ty, components: _ } => Resolution::Handle(ty),
},
crate::Expression::Compose { ty, .. } => Resolution::Handle(ty),
crate::Expression::FunctionArgument(index) => {
Resolution::Handle(ctx.arguments[index as usize].ty)
@ -458,38 +466,3 @@ impl Typifier {
Ok(())
}
}
pub fn check_constant_type(inner: &crate::ConstantInner, type_inner: &crate::TypeInner) -> bool {
match (inner, type_inner) {
(
crate::ConstantInner::Sint(_),
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Sint,
width: _,
},
) => true,
(
crate::ConstantInner::Uint(_),
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Uint,
width: _,
},
) => true,
(
crate::ConstantInner::Float(_),
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Float,
width: _,
},
) => true,
(
crate::ConstantInner::Bool(_),
crate::TypeInner::Scalar {
kind: crate::ScalarKind::Bool,
width: _,
},
) => true,
(crate::ConstantInner::Composite(_inner), _) => true, // TODO recursively check composite types
(_, _) => false,
}
}

70
src/proc/validator.rs
View File

@ -23,6 +23,14 @@ pub enum TypeError {
InvalidArraySizeConstant(Handle<crate::Constant>),
}
#[derive(Clone, Debug, PartialEq, thiserror::Error)]
pub enum ConstantError {
#[error("The type doesn't match the constant")]
InvalidType,
#[error("The component handle {0:?} can not be resolved")]
UnresolvedComponent(Handle<crate::Constant>),
}
#[derive(Clone, Debug, PartialEq, thiserror::Error)]
pub enum GlobalVariableError {
#[error("Usage isn't compatible with the storage class")]
@ -94,6 +102,12 @@ pub enum ValidationError {
name: String,
error: TypeError,
},
#[error("Constant {handle:?} '{name}' is invalid: {error:?}")]
Constant {
handle: Handle<crate::Constant>,
name: String,
error: ConstantError,
},
#[error("Global variable {handle:?} '{name}' is invalid: {error:?}")]
GlobalVariable {
handle: Handle<crate::GlobalVariable>,
@ -227,6 +241,13 @@ impl Validator {
}
}
fn check_width(kind: crate::ScalarKind, width: crate::Bytes) -> bool {
match kind {
crate::ScalarKind::Bool => width == 1,
_ => width == 4,
}
}
fn validate_type(
&self,
ty: &crate::Type,
@ -236,16 +257,12 @@ impl Validator {
use crate::TypeInner as Ti;
match ty.inner {
Ti::Scalar { kind, width } | Ti::Vector { kind, width, .. } => {
let expected = match kind {
crate::ScalarKind::Bool => 1,
_ => 4,
};
if width != expected {
if !Self::check_width(kind, width) {
return Err(TypeError::InvalidWidth(kind, width));
}
}
Ti::Matrix { width, .. } => {
if width != 4 {
if !Self::check_width(crate::ScalarKind::Float, width) {
return Err(TypeError::InvalidWidth(crate::ScalarKind::Float, width));
}
}
@ -261,7 +278,11 @@ impl Validator {
if let crate::ArraySize::Constant(const_handle) = size {
match constants.try_get(const_handle) {
Some(&crate::Constant {
inner: crate::ConstantInner::Uint(_),
inner:
crate::ConstantInner::Scalar {
width: _,
value: crate::ScalarValue::Uint(_),
},
..
}) => {}
_ => {
@ -287,6 +308,32 @@ impl Validator {
Ok(())
}
fn validate_constant(
&self,
handle: Handle<crate::Constant>,
constants: &Arena<crate::Constant>,
_types: &Arena<crate::Type>,
) -> Result<(), ConstantError> {
let con = &constants[handle];
match con.inner {
crate::ConstantInner::Scalar { width, ref value } => {
if !Self::check_width(value.scalar_kind(), width) {
return Err(ConstantError::InvalidType);
}
}
crate::ConstantInner::Composite {
ty: _,
ref components,
} => {
if let Some(&comp) = components.iter().find(|&&comp| handle <= comp) {
return Err(ConstantError::UnresolvedComponent(comp));
}
//TODO
}
}
Ok(())
}
fn validate_global_var(
&self,
var: &crate::GlobalVariable,
@ -522,7 +569,6 @@ impl Validator {
/// Check the given module to be valid.
pub fn validate(&mut self, module: &crate::Module) -> Result<(), ValidationError> {
// check the types
for (handle, ty) in module.types.iter() {
self.validate_type(ty, handle, &module.constants)
.map_err(|error| ValidationError::Type {
@ -532,6 +578,14 @@ impl Validator {
})?;
}
for (handle, constant) in module.constants.iter() {
self.validate_constant(handle, &module.constants, &module.types)
.map_err(|error| ValidationError::Constant {
handle,
name: constant.name.clone().unwrap_or_default(),
error,
})?;
}
for (var_handle, var) in module.global_variables.iter() {
self.validate_global_var(var, &module.types)
.map_err(|error| ValidationError::GlobalVariable {

12
test-data/simple/module.ron
View File

@ -32,14 +32,18 @@
(
name: None,
specialization: None,
inner: Float(1),
ty: 3,
inner: Scalar(
width: 4,
value: Float(1),
),
),
(
name: None,
specialization: None,
inner: Float(0),
ty: 3,
inner: Scalar(
width: 4,
value: Float(0),
),
),
],
global_variables: [

12
tests/snapshots/snapshots__boids.msl.snap
View File

@ -21,12 +21,12 @@ struct SimParams {
type2 rule2Scale;
type2 rule3Scale;
};
typedef uint type3;
typedef Particle type4[5];
typedef Particle type3[5];
struct Particles {
type4 particles;
type3 particles;
};
typedef metal::uint3 type5;
typedef metal::uint3 type4;
typedef uint type5;
typedef int type6;
struct main1Input {
@ -60,7 +60,7 @@ kernel void main3(
constant SimParams& params [[buffer(0)]],
constant Particles& particlesA [[buffer(1)]],
device Particles& particlesB [[buffer(2)]],
type5 gl_GlobalInvocationID [[thread_position_in_grid]]
type4 gl_GlobalInvocationID [[thread_position_in_grid]]
) {
type vPos;
type vVel;
@ -71,7 +71,7 @@ kernel void main3(
type6 cVelCount = 0;
type pos1;
type vel1;
type3 i = 0;
type5 i = 0;
if (gl_GlobalInvocationID.x >= static_cast<uint>(5)) {
}
vPos = particlesA.particles[gl_GlobalInvocationID.x].pos;

1
tests/snapshots/snapshots__simple.msl.snap
View File

@ -6,7 +6,6 @@ expression: msl
#include <simd/simd.h>
typedef metal::float4 type;
typedef int type1;
struct main1Input {
};

4
tests/snapshots/snapshots__skybox.msl.snap
View File

@ -14,8 +14,8 @@ struct Data {
type3 view;
};
typedef int type4;
typedef float type5;
typedef metal::float3x3 type6;
typedef metal::float3x3 type5;
typedef float type6;
typedef metal::texturecube<float, metal::access::sample> type7;
typedef metal::sampler type8;
typedef metal::int3 type9;