From 8519095f554bb18d035d62191d4def05398e8fc8 Mon Sep 17 00:00:00 2001 From: =?UTF-8?q?Jo=C3=A3o=20Capucho?= Date: Fri, 7 May 2021 21:00:16 +0100 Subject: [PATCH] [glsl-in] Handle global declarations parse constant expressions fully --- src/front/glsl/ast.rs | 213 +++++++++++++++++++----------- src/front/glsl/functions.rs | 180 +++++++++++++++---------- src/front/glsl/mod.rs | 4 +- src/front/glsl/parser.rs | 75 +++++++---- src/front/glsl/variables.rs | 257 ++++++++++++++++++++++++++++++++++-- 5 files changed, 543 insertions(+), 186 deletions(-) diff --git a/src/front/glsl/ast.rs b/src/front/glsl/ast.rs index 08c265d311..bfd49d6bc5 100644 --- a/src/front/glsl/ast.rs +++ b/src/front/glsl/ast.rs @@ -1,113 +1,80 @@ use super::{super::Typifier, constants::ConstantSolver, error::ErrorKind, TokenMetadata}; use crate::{ - proc::ResolveContext, Arena, ArraySize, BinaryOperator, Binding, Constant, Expression, + proc::ResolveContext, Arena, ArraySize, BinaryOperator, BuiltIn, Constant, Expression, FastHashMap, Function, FunctionArgument, GlobalVariable, Handle, Interpolation, Module, RelationalFunction, ResourceBinding, Sampling, ShaderStage, Statement, StorageClass, Type, - UnaryOperator, + TypeInner, UnaryOperator, }; +#[derive(Debug)] +pub enum GlobalLookup { + Variable(Handle), + Select(u32), +} + #[derive(Debug)] pub struct Program<'a> { pub version: u16, pub profile: Profile, pub entry_points: &'a FastHashMap, + pub lookup_function: FastHashMap>, pub lookup_type: FastHashMap>, - pub lookup_global_variables: FastHashMap>, + pub lookup_global_variables: FastHashMap, pub lookup_constants: FastHashMap>, + + pub built_ins: Vec<(BuiltIn, Handle)>, + pub entries: Vec<(String, ShaderStage, Handle)>, + + pub input_struct: Handle, + pub output_struct: Handle, + pub module: Module, } impl<'a> Program<'a> { pub fn new(entry_points: &'a FastHashMap) -> Program<'a> { + let mut module = Module::default(); + Program { version: 0, profile: Profile::Core, entry_points, + lookup_function: FastHashMap::default(), lookup_type: FastHashMap::default(), lookup_global_variables: FastHashMap::default(), lookup_constants: FastHashMap::default(), - module: Module::default(), - } - } - pub fn binary_expr( - &mut self, - function: &mut Function, - op: BinaryOperator, - left: &ExpressionRule, - right: &ExpressionRule, - ) -> ExpressionRule { - ExpressionRule::from_expression(function.expressions.append(Expression::Binary { - op, - left: left.expression, - right: right.expression, - })) - } + built_ins: Vec::new(), + entries: Vec::new(), - pub fn unary_expr( - &mut self, - function: &mut Function, - op: UnaryOperator, - tgt: &ExpressionRule, - ) -> ExpressionRule { - ExpressionRule::from_expression(function.expressions.append(Expression::Unary { - op, - expr: tgt.expression, - })) - } - - /// Helper function to insert equality expressions, this handles the special - /// case of `vec1 == vec2` and `vec1 != vec2` since in the IR they are - /// represented as `all(equal(vec1, vec2))` and `any(notEqual(vec1, vec2))` - pub fn equality_expr( - &mut self, - context: &mut FunctionContext, - equals: bool, - left: &ExpressionRule, - right: &ExpressionRule, - ) -> Result { - let left_is_vector = match *self.resolve_type(context, left.expression)? { - crate::TypeInner::Vector { .. } => true, - _ => false, - }; - - let right_is_vector = match *self.resolve_type(context, right.expression)? { - crate::TypeInner::Vector { .. } => true, - _ => false, - }; - - let (op, fun) = match equals { - true => (BinaryOperator::Equal, RelationalFunction::All), - false => (BinaryOperator::NotEqual, RelationalFunction::Any), - }; - - let expr = ExpressionRule::from_expression(context.function.expressions.append( - Expression::Binary { - op, - left: left.expression, - right: right.expression, - }, - )); - - Ok(if left_is_vector && right_is_vector { - ExpressionRule::from_expression(context.function.expressions.append( - Expression::Relational { - fun, - argument: expr.expression, + input_struct: module.types.append(Type { + name: None, + inner: TypeInner::Struct { + level: crate::StructLevel::Root, + members: Vec::new(), + span: 0, }, - )) - } else { - expr - }) + }), + output_struct: module.types.append(Type { + name: None, + inner: TypeInner::Struct { + level: crate::StructLevel::Root, + members: Vec::new(), + span: 0, + }, + }), + + module, + } } pub fn resolve_type<'b>( &'b mut self, context: &'b mut FunctionContext, handle: Handle, - ) -> Result<&'b crate::TypeInner, ErrorKind> { + ) -> Result<&'b TypeInner, ErrorKind> { let resolve_ctx = ResolveContext { constants: &self.module.constants, global_vars: &self.module.global_variables, @@ -226,6 +193,61 @@ impl<'function> FunctionContext<'function> { None } + pub fn lookup_global_var( + &mut self, + program: &mut Program, + name: &str, + ) -> Option { + self.lookup_global_var_exps.get(name).cloned().or_else(|| { + let expr = match *program.lookup_global_variables.get(name)? { + GlobalLookup::Variable(v) => Expression::GlobalVariable(v), + GlobalLookup::Select(index) => { + let base = self + .function + .expressions + .append(Expression::FunctionArgument( + self.function.arguments.len() as u32 - 1, + )); + + Expression::AccessIndex { base, index } + } + }; + + let expr = self.function.expressions.append(expr); + let var = VariableReference { + expr, + load: Some( + self.function + .expressions + .append(Expression::Load { pointer: expr }), + ), + }; + + self.lookup_global_var_exps.insert(name.into(), var.clone()); + + Some(var) + }) + } + + pub fn lookup_constants_var( + &mut self, + program: &mut Program, + name: &str, + ) -> Option { + self.lookup_constant_exps.get(name).cloned().or_else(|| { + let expr = self + .function + .expressions + .append(Expression::Constant(*program.lookup_constants.get(name)?)); + + let var = VariableReference { expr, load: None }; + + self.lookup_constant_exps.insert(name.into(), var.clone()); + + Some(var) + }) + } + #[cfg(feature = "glsl-validate")] pub fn lookup_local_var_current_scope(&self, name: &str) -> Option { if let Some(current) = self.scopes.last() { @@ -316,9 +338,41 @@ impl<'function> FunctionContext<'function> { let left = self.resolve(program, *left, false)?; let right = self.resolve(program, *right, false)?; - self.function - .expressions - .append(Expression::Binary { left, op, right }) + if let BinaryOperator::Equal | BinaryOperator::NotEqual = op { + let equals = op == BinaryOperator::Equal; + let left_is_vector = match *program.resolve_type(self, left)? { + crate::TypeInner::Vector { .. } => true, + _ => false, + }; + + let right_is_vector = match *program.resolve_type(self, right)? { + crate::TypeInner::Vector { .. } => true, + _ => false, + }; + + let (op, fun) = match equals { + true => (BinaryOperator::Equal, RelationalFunction::All), + false => (BinaryOperator::NotEqual, RelationalFunction::Any), + }; + + let expr = + self.function + .expressions + .append(Expression::Binary { op, left, right }); + + if left_is_vector && right_is_vector { + self.function.expressions.append(Expression::Relational { + fun, + argument: expr, + }) + } else { + expr + } + } else { + self.function + .expressions + .append(Expression::Binary { left, op, right }) + } } ExprKind::Unary { op, expr } => { let expr = self.resolve(program, *expr, false)?; @@ -429,7 +483,7 @@ pub enum TypeQualifier { StorageQualifier(StorageQualifier), Interpolation(Interpolation), ResourceBinding(ResourceBinding), - Binding(Binding), + Location(u32), Sampling(Sampling), Layout(StructLayout), EarlyFragmentTests, @@ -454,12 +508,11 @@ pub struct FunctionCall { pub args: Vec, } -#[derive(Debug, Clone, Copy)] +#[derive(Debug, Clone, Copy, PartialEq)] pub enum StorageQualifier { StorageClass(StorageClass), Input, Output, - InOut, Const, } diff --git a/src/front/glsl/functions.rs b/src/front/glsl/functions.rs index bc0cc33676..8b22ad7ab9 100644 --- a/src/front/glsl/functions.rs +++ b/src/front/glsl/functions.rs @@ -1,6 +1,7 @@ use crate::{ - proc::ensure_block_returns, BinaryOperator, EntryPoint, Expression, Function, MathFunction, - RelationalFunction, SampleLevel, TypeInner, + proc::ensure_block_returns, Arena, BinaryOperator, Binding, BuiltIn, EntryPoint, Expression, + Function, FunctionArgument, FunctionResult, MathFunction, RelationalFunction, SampleLevel, + ShaderStage, Statement, TypeInner, }; use super::{ast::*, error::ErrorKind}; @@ -294,69 +295,25 @@ impl Program<'_> { // } // } - pub fn parse_relational_fun( - &mut self, - context: &mut FunctionContext, - name: String, - args: Vec, - fun: RelationalFunction, - ) -> Result { - if args.len() != 1 { - return Err(ErrorKind::WrongNumberArgs(name, 1, args.len())); - } - Ok(ExpressionRule { - expression: context.function.expressions.append(Expression::Relational { - fun, - argument: args[0].expression, - }), - sampler: None, - statements: args.into_iter().flat_map(|a| a.statements).collect(), - }) - } - // TODO: Reenable later - // pub fn add_function_prelude(&mut self, context: &mut FunctionContext) { - // for (var_handle, var) in self.module.global_variables.iter() { - // if let Some(name) = var.name.as_ref() { - // let expr = context - // .function - // .expressions - // .append(Expression::GlobalVariable(var_handle)); - // context.lookup_global_var_exps.insert(name.clone(), expr); - // } else { - // let ty = &self.module.types[var.ty]; - // // anonymous structs - // if let TypeInner::Struct { ref members, .. } = ty.inner { - // let base = context - // .function - // .expressions - // .append(Expression::GlobalVariable(var_handle)); - // for (idx, member) in members.iter().enumerate() { - // if let Some(name) = member.name.as_ref() { - // let exp = - // context - // .function - // .expressions - // .append(Expression::AccessIndex { - // base, - // index: idx as u32, - // }); - // context.lookup_global_var_exps.insert(name.clone(), exp); - // } - // } - // } - // } - // } - - // for (handle, constant) in self.module.constants.iter() { - // if let Some(name) = constant.name.as_ref() { - // let expr = context - // .function - // .expressions - // .append(Expression::Constant(handle)); - // context.lookup_constant_exps.insert(name.clone(), expr); - // } + // pub fn parse_relational_fun( + // &mut self, + // context: &mut FunctionContext, + // name: String, + // args: Vec, + // fun: RelationalFunction, + // ) -> Result { + // if args.len() != 1 { + // return Err(ErrorKind::WrongNumberArgs(name, 1, args.len())); // } + // Ok(ExpressionRule { + // expression: context.function.expressions.append(Expression::Relational { + // fun, + // argument: args[0].expression, + // }), + // sampler: None, + // statements: args.into_iter().flat_map(|a| a.statements).collect(), + // }) // } pub fn add_function(&mut self, mut function: Function) -> Result<(), ErrorKind> { @@ -366,18 +323,103 @@ impl Program<'_> { .clone() .ok_or_else(|| ErrorKind::SemanticError("Unnamed function".into()))?; let stage = self.entry_points.get(&name); + + // Add the input variables as a function argument + function.arguments.push(FunctionArgument { + name: None, + ty: self.input_struct, + binding: None, + }); + + let handle = self.module.functions.append(function); + if let Some(&stage) = stage { + self.entries.push((name, stage, handle)); + } else { + self.lookup_function.insert(name, handle); + } + + Ok(()) + } + + pub fn add_entry_points(&mut self) { + for (name, stage, function) in self.entries.iter().cloned() { + let mut arguments = Vec::new(); + let mut expressions = Arena::new(); + let mut body = Vec::new(); + + for (built_in, handle) in self.built_ins.iter().copied() { + let ty = self.module.global_variables[handle].ty; + let arg = arguments.len() as u32; + + arguments.push(FunctionArgument { + name: None, + ty, + binding: Some(Binding::BuiltIn(built_in)), + }); + + let pointer = expressions.append(Expression::GlobalVariable(handle)); + let value = expressions.append(Expression::FunctionArgument(arg)); + + body.push(Statement::Store { pointer, value }); + } + + let i = arguments.len() as u32; + arguments.push(FunctionArgument { + name: None, + ty: self.input_struct, + binding: None, + }); + + let res = expressions.append(Expression::Call(function)); + + body.push(Statement::Call { + function, + arguments: vec![expressions.append(Expression::FunctionArgument(i))], + result: Some(res), + }); + + for (i, (built_in, handle)) in self.built_ins.iter().copied().enumerate() { + if !should_write(built_in, stage) { + continue; + } + + let value = expressions.append(Expression::GlobalVariable(handle)); + let pointer = expressions.append(Expression::FunctionArgument(i as u32)); + + body.push(Statement::Store { pointer, value }); + } + + body.push(Statement::Return { value: Some(res) }); + self.module.entry_points.push(EntryPoint { name, stage, + // TODO early_depth_test: None, - workgroup_size: [0; 3], //TODO - function, + workgroup_size: [0; 3], + function: Function { + arguments, + expressions, + body, + result: Some(FunctionResult { + ty: self.output_struct, + binding: None, + }), + ..Default::default() + }, }); - } else { - let handle = self.module.functions.append(function); - self.lookup_function.insert(name, handle); } - Ok(()) + } +} + +fn should_write(built_in: BuiltIn, stage: ShaderStage) -> bool { + match (built_in, stage) { + (BuiltIn::Position, ShaderStage::Vertex) + | (BuiltIn::ClipDistance, ShaderStage::Vertex) + | (BuiltIn::PointSize, ShaderStage::Vertex) + | (BuiltIn::FragDepth, ShaderStage::Fragment) + | (BuiltIn::SampleMask, ShaderStage::Fragment) => true, + _ => false, } } diff --git a/src/front/glsl/mod.rs b/src/front/glsl/mod.rs index 17b8a57845..590ddd717e 100644 --- a/src/front/glsl/mod.rs +++ b/src/front/glsl/mod.rs @@ -14,15 +14,13 @@ mod error; pub use error::ParseError; mod constants; // TODO: Remove later -#[allow(dead_code, unused_imports)] +#[allow(unused_imports)] mod functions; mod parser; #[cfg(test)] mod parser_tests; mod token; mod types; -// TODO: Remove later -#[allow(dead_code)] mod variables; pub struct Options { diff --git a/src/front/glsl/parser.rs b/src/front/glsl/parser.rs index 0cc769ccd5..8d6a6f4e1a 100644 --- a/src/front/glsl/parser.rs +++ b/src/front/glsl/parser.rs @@ -8,9 +8,8 @@ use super::{ Program, }; use crate::{ - arena::Handle, ArraySize, BinaryOperator, Binding, Constant, ConstantInner, Function, - FunctionResult, ResourceBinding, ScalarValue, Statement, StorageClass, Type, TypeInner, - UnaryOperator, + arena::Handle, ArraySize, BinaryOperator, Constant, ConstantInner, Function, FunctionResult, + ResourceBinding, ScalarValue, Statement, StorageClass, Type, TypeInner, UnaryOperator, }; use core::convert::TryFrom; use std::iter::Peekable; @@ -73,6 +72,8 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { self.parse_external_declaration()?; } + self.program.add_entry_points(); + Ok(()) } @@ -161,7 +162,6 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { | TokenValue::Const | TokenValue::In | TokenValue::Out - | TokenValue::InOut | TokenValue::Uniform | TokenValue::Layout => true, _ => false, @@ -185,7 +185,6 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { TokenValue::Const => TypeQualifier::StorageQualifier(StorageQualifier::Const), TokenValue::In => TypeQualifier::StorageQualifier(StorageQualifier::Input), TokenValue::Out => TypeQualifier::StorageQualifier(StorageQualifier::Output), - TokenValue::InOut => TypeQualifier::StorageQualifier(StorageQualifier::InOut), TokenValue::Uniform => TypeQualifier::StorageQualifier( StorageQualifier::StorageClass(StorageClass::Uniform), ), @@ -242,6 +241,25 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { Ok(()) } + fn parse_uint_constant(&mut self) -> Result { + let value = self.parse_constant_expression()?; + + // TODO: better errors + match self.program.module.constants[value].inner { + ConstantInner::Scalar { + value: ScalarValue::Uint(int), + .. + } => u32::try_from(int) + .map_err(|_| ErrorKind::SemanticError("int constant overflows".into())), + ConstantInner::Scalar { + value: ScalarValue::Sint(int), + .. + } => u32::try_from(int) + .map_err(|_| ErrorKind::SemanticError("int constant overflows".into())), + _ => Err(ErrorKind::SemanticError("Expected a uint constant".into())), + } + } + fn parse_layout_qualifier_id( &mut self, qualifiers: &mut Vec, @@ -256,18 +274,12 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { match token.value { TokenValue::Identifier(name) => { if self.bump_if(TokenValue::Assign).is_some() { - let value = self.parse_constant_expression()?; + let value = self.parse_uint_constant()?; match name.as_str() { - "location" => qualifiers.push(TypeQualifier::Binding(Binding::Location { - // TODO: use better error here - location: u32::try_from(value).map_err(|_| ErrorKind::InvalidInput)?, - interpolation: None, - sampling: None, - })), - // TODO: produce error when try_from fails - "set" => *group = u32::try_from(value).ok(), - "binding" => *binding = u32::try_from(value).ok(), + "location" => qualifiers.push(TypeQualifier::Location(value)), + "set" => *group = Some(value), + "binding" => *binding = Some(value), _ => return Err(ErrorKind::UnknownLayoutQualifier(token.meta, name)), } } else { @@ -287,13 +299,14 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { } } - // TODO: parse more than just integer literal - fn parse_constant_expression(&mut self) -> Result { - let token = self.bump()?; - match token.value { - TokenValue::IntConstant(int) => Ok(int.value as i64), - _ => Err(ErrorKind::InvalidToken(token)), - } + fn parse_constant_expression(&mut self) -> Result> { + let mut function = Function::default(); + let mut ctx = FunctionContext::new(&mut function); + + let expr = self.parse_conditional(&mut ctx, None)?; + let root = ctx.resolve(self.program, expr, false)?; + + self.program.solve_constant(&function.expressions, root) } fn parse_external_declaration(&mut self) -> Result<()> { @@ -329,8 +342,7 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { // type_qualifier IDENTIFIER identifier_list SEMICOLON // TODO: Handle precision qualifiers if self.peek_type_qualifier() || self.peek_type_name() { - // TODO: Use qualifiers - let _qualifiers = self.parse_type_qualifiers()?; + let qualifiers = self.parse_type_qualifiers()?; if self.peek_type_name() { // Functions and variable declarations @@ -371,9 +383,20 @@ impl<'source, 'program, 'options> Parser<'source, 'program, 'options> { } // Variable Declaration TokenValue::Semicolon => { - // TODO: are we supposed to consume the semicolon here self.bump()?; - // TODO: use parsed type & qualifiers + + if let Some(ty) = ty { + if external { + self.program.add_global_var(qualifiers, ty, name, None)?; + } else { + // TODO: local variables + } + } else { + return Err(ErrorKind::SemanticError( + "Declaration cannot have void type".into(), + )); + } + Ok(true) } TokenValue::Comma => todo!(), diff --git a/src/front/glsl/variables.rs b/src/front/glsl/variables.rs index b7d32f4dec..b27fe14fb7 100644 --- a/src/front/glsl/variables.rs +++ b/src/front/glsl/variables.rs @@ -1,4 +1,7 @@ -use crate::{Expression, Handle, Type, TypeInner, VectorSize}; +use crate::{ + Binding, BuiltIn, Constant, Expression, GlobalVariable, Handle, ScalarKind, StorageAccess, + StorageClass, StructMember, Type, TypeInner, VectorSize, +}; use super::ast::*; use super::error::ErrorKind; @@ -13,21 +16,88 @@ impl Program<'_> { if let Some(local_var) = context.lookup_local_var(name) { return Ok(Some(local_var)); } - if let Some(global_var) = context.lookup_global_var_exps.get(name) { - return Ok(Some(global_var.clone())); + if let Some(global_var) = context.lookup_global_var(self, name) { + return Ok(Some(global_var)); } - if let Some(constant) = context.lookup_constant_exps.get(name) { - return Ok(Some(constant.clone())); + if let Some(constant) = context.lookup_global_var(self, name) { + return Ok(Some(constant)); } match name { "gl_Position" => { - todo!() + let ty = self.module.types.fetch_or_append(Type { + name: None, + inner: TypeInner::Vector { + size: VectorSize::Quad, + kind: ScalarKind::Float, + width: 4, + }, + }); + + let handle = self.module.global_variables.append(GlobalVariable { + name: Some(name.into()), + class: StorageClass::Function, + binding: None, + ty, + init: None, + storage_access: StorageAccess::all(), + }); + + self.built_ins.push((BuiltIn::Position, handle)); + + self.lookup_global_variables + .insert(name.into(), GlobalLookup::Variable(handle)); + + Ok(context.lookup_global_var(self, name)) } "gl_VertexIndex" => { - todo!() + let ty = self.module.types.fetch_or_append(Type { + name: None, + inner: TypeInner::Scalar { + kind: ScalarKind::Sint, + width: 4, + }, + }); + + let handle = self.module.global_variables.append(GlobalVariable { + name: Some(name.into()), + class: StorageClass::Function, + binding: None, + ty, + init: None, + storage_access: StorageAccess::all(), + }); + + self.built_ins.push((BuiltIn::VertexIndex, handle)); + + self.lookup_global_variables + .insert(name.into(), GlobalLookup::Variable(handle)); + + Ok(context.lookup_global_var(self, name)) } "gl_InstanceIndex" => { - todo!() + let ty = self.module.types.fetch_or_append(Type { + name: None, + inner: TypeInner::Scalar { + kind: ScalarKind::Sint, + width: 4, + }, + }); + + let handle = self.module.global_variables.append(GlobalVariable { + name: Some(name.into()), + class: StorageClass::Function, + binding: None, + ty, + init: None, + storage_access: StorageAccess::all(), + }); + + self.built_ins.push((BuiltIn::InstanceIndex, handle)); + + self.lookup_global_variables + .insert(name.into(), GlobalLookup::Variable(handle)); + + Ok(context.lookup_global_var(self, name)) } _ => Ok(None), } @@ -130,4 +200,175 @@ impl Program<'_> { )), } } + + // TODO: constants + pub fn add_global_var( + &mut self, + qualifiers: Vec, + ty: Handle, + name: String, + init: Option>, + ) -> Result<(), ErrorKind> { + let mut storage = StorageQualifier::StorageClass(StorageClass::Function); + let mut interpolation = None; + let mut binding = None; + let mut location = None; + let mut sampling = None; + let mut layout = None; + + for qualifier in qualifiers { + match qualifier { + TypeQualifier::StorageQualifier(s) => { + if StorageQualifier::StorageClass(StorageClass::Function) != storage { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + storage = s; + } + TypeQualifier::Interpolation(i) => { + if interpolation.is_some() { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + interpolation = Some(i); + } + TypeQualifier::ResourceBinding(r) => { + if binding.is_some() { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + binding = Some(r); + } + TypeQualifier::Location(l) => { + if location.is_some() { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + location = Some(l); + } + TypeQualifier::Sampling(s) => { + if sampling.is_some() { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + sampling = Some(s); + } + TypeQualifier::Layout(l) => { + if layout.is_some() { + return Err(ErrorKind::SemanticError( + "Cannot use more than one storage qualifier per declaration".into(), + )); + } + + layout = Some(l); + } + TypeQualifier::EarlyFragmentTests => { + return Err(ErrorKind::SemanticError( + "Cannot set early fragment tests on a declaration".into(), + )); + } + } + } + + if binding.is_some() && storage != StorageQualifier::StorageClass(StorageClass::Uniform) { + return Err(ErrorKind::SemanticError( + "binding requires uniform or buffer storage qualifier".into(), + )); + } + + if (sampling.is_some() || interpolation.is_some()) && location.is_none() { + return Err(ErrorKind::SemanticError( + "Sampling and interpolation qualifiers can only be used in in/out variables".into(), + )); + } + + if let Some(location) = location { + let input = StorageQualifier::Input == storage; + + let index = self.add_member( + input, + Some(name.clone()), + ty, + Some(Binding::Location { + location, + interpolation, + sampling, + }), + ); + + if let Some(i) = index { + self.lookup_global_variables + .insert(name, GlobalLookup::Select(i)); + } + } else if let StorageQualifier::StorageClass(class) = storage { + let handle = self.module.global_variables.append(GlobalVariable { + name: Some(name.clone()), + class, + binding, + ty, + init, + // TODO + storage_access: StorageAccess::all(), + }); + + self.lookup_global_variables + .insert(name, GlobalLookup::Variable(handle)); + } + + Ok(()) + } + + fn add_member( + &mut self, + input: bool, + name: Option, + ty: Handle, + binding: Option, + ) -> Option { + let handle = match input { + true => self.input_struct, + false => self.output_struct, + }; + + let offset = if let TypeInner::Struct { ref members, .. } = self.module.types[handle].inner + { + members + .last() + .map(|member| { + member.offset + + self.module.types[member.ty] + .inner + .span(&self.module.constants) + }) + .unwrap_or(0) + } else { + 0 + }; + + if let TypeInner::Struct { + ref mut members, .. + } = self.module.types.get_mut(handle).inner + { + members.push(StructMember { + name, + ty, + binding, + offset, + }); + + Some(members.len() as u32 - 1) + } else { + None + } + } }