summaryrefslogtreecommitdiffstats
path: root/third_party/rust/naga/src/front
diff options
context:
space:
mode:
Diffstat (limited to 'third_party/rust/naga/src/front')
-rw-r--r--third_party/rust/naga/src/front/glsl/functions.rs2
-rw-r--r--third_party/rust/naga/src/front/glsl/parser/functions.rs2
-rw-r--r--third_party/rust/naga/src/front/spv/function.rs464
-rw-r--r--third_party/rust/naga/src/front/spv/mod.rs68
-rw-r--r--third_party/rust/naga/src/front/wgsl/error.rs2
-rw-r--r--third_party/rust/naga/src/front/wgsl/lower/mod.rs2
-rw-r--r--third_party/rust/naga/src/front/wgsl/parse/conv.rs8
-rw-r--r--third_party/rust/naga/src/front/wgsl/parse/number.rs16
-rw-r--r--third_party/rust/naga/src/front/wgsl/tests.rs1
9 files changed, 301 insertions, 264 deletions
diff --git a/third_party/rust/naga/src/front/glsl/functions.rs b/third_party/rust/naga/src/front/glsl/functions.rs
index df8cc8a30e..01846eb814 100644
--- a/third_party/rust/naga/src/front/glsl/functions.rs
+++ b/third_party/rust/naga/src/front/glsl/functions.rs
@@ -160,7 +160,7 @@ impl Frontend {
} => self.matrix_one_arg(ctx, ty, columns, rows, scalar, (value, expr_meta), meta)?,
TypeInner::Struct { ref members, .. } => {
let scalar_components = members
- .get(0)
+ .first()
.and_then(|member| scalar_components(&ctx.module.types[member.ty].inner));
if let Some(scalar) = scalar_components {
ctx.implicit_conversion(&mut value, expr_meta, scalar)?;
diff --git a/third_party/rust/naga/src/front/glsl/parser/functions.rs b/third_party/rust/naga/src/front/glsl/parser/functions.rs
index 38184eedf7..d428d74761 100644
--- a/third_party/rust/naga/src/front/glsl/parser/functions.rs
+++ b/third_party/rust/naga/src/front/glsl/parser/functions.rs
@@ -435,7 +435,7 @@ impl<'source> ParsingContext<'source> {
if self.bump_if(frontend, TokenValue::Semicolon).is_none() {
if self.peek_type_name(frontend) || self.peek_type_qualifier(frontend) {
- self.parse_declaration(frontend, ctx, false, false)?;
+ self.parse_declaration(frontend, ctx, false, is_inside_loop)?;
} else {
let mut stmt = ctx.stmt_ctx();
let expr = self.parse_expression(frontend, ctx, &mut stmt)?;
diff --git a/third_party/rust/naga/src/front/spv/function.rs b/third_party/rust/naga/src/front/spv/function.rs
index 198d9c52dd..e81ecf5c9b 100644
--- a/third_party/rust/naga/src/front/spv/function.rs
+++ b/third_party/rust/naga/src/front/spv/function.rs
@@ -292,278 +292,286 @@ impl<I: Iterator<Item = u32>> super::Frontend<I> {
);
if let Some(ep) = self.lookup_entry_point.remove(&fun_id) {
- // create a wrapping function
- let mut function = crate::Function {
- name: Some(format!("{}_wrap", ep.name)),
- arguments: Vec::new(),
- result: None,
- local_variables: Arena::new(),
- expressions: Arena::new(),
- named_expressions: crate::NamedExpressions::default(),
- body: crate::Block::new(),
- };
+ self.deferred_entry_points.push((ep, fun_id));
+ }
- // 1. copy the inputs from arguments to privates
- for &v_id in ep.variable_ids.iter() {
- let lvar = self.lookup_variable.lookup(v_id)?;
- if let super::Variable::Input(ref arg) = lvar.inner {
- let span = module.global_variables.get_span(lvar.handle);
- let arg_expr = function.expressions.append(
- crate::Expression::FunctionArgument(function.arguments.len() as u32),
- span,
- );
- let load_expr = if arg.ty == module.global_variables[lvar.handle].ty {
- arg_expr
- } else {
- // The only case where the type is different is if we need to treat
- // unsigned integer as signed.
- let mut emitter = Emitter::default();
- emitter.start(&function.expressions);
- let handle = function.expressions.append(
- crate::Expression::As {
- expr: arg_expr,
- kind: crate::ScalarKind::Sint,
- convert: Some(4),
- },
- span,
- );
- function.body.extend(emitter.finish(&function.expressions));
- handle
- };
- function.body.push(
- crate::Statement::Store {
- pointer: function
- .expressions
- .append(crate::Expression::GlobalVariable(lvar.handle), span),
- value: load_expr,
+ Ok(())
+ }
+
+ pub(super) fn process_entry_point(
+ &mut self,
+ module: &mut crate::Module,
+ ep: super::EntryPoint,
+ fun_id: u32,
+ ) -> Result<(), Error> {
+ // create a wrapping function
+ let mut function = crate::Function {
+ name: Some(format!("{}_wrap", ep.name)),
+ arguments: Vec::new(),
+ result: None,
+ local_variables: Arena::new(),
+ expressions: Arena::new(),
+ named_expressions: crate::NamedExpressions::default(),
+ body: crate::Block::new(),
+ };
+
+ // 1. copy the inputs from arguments to privates
+ for &v_id in ep.variable_ids.iter() {
+ let lvar = self.lookup_variable.lookup(v_id)?;
+ if let super::Variable::Input(ref arg) = lvar.inner {
+ let span = module.global_variables.get_span(lvar.handle);
+ let arg_expr = function.expressions.append(
+ crate::Expression::FunctionArgument(function.arguments.len() as u32),
+ span,
+ );
+ let load_expr = if arg.ty == module.global_variables[lvar.handle].ty {
+ arg_expr
+ } else {
+ // The only case where the type is different is if we need to treat
+ // unsigned integer as signed.
+ let mut emitter = Emitter::default();
+ emitter.start(&function.expressions);
+ let handle = function.expressions.append(
+ crate::Expression::As {
+ expr: arg_expr,
+ kind: crate::ScalarKind::Sint,
+ convert: Some(4),
},
span,
);
+ function.body.extend(emitter.finish(&function.expressions));
+ handle
+ };
+ function.body.push(
+ crate::Statement::Store {
+ pointer: function
+ .expressions
+ .append(crate::Expression::GlobalVariable(lvar.handle), span),
+ value: load_expr,
+ },
+ span,
+ );
- let mut arg = arg.clone();
- if ep.stage == crate::ShaderStage::Fragment {
- if let Some(ref mut binding) = arg.binding {
- binding.apply_default_interpolation(&module.types[arg.ty].inner);
- }
+ let mut arg = arg.clone();
+ if ep.stage == crate::ShaderStage::Fragment {
+ if let Some(ref mut binding) = arg.binding {
+ binding.apply_default_interpolation(&module.types[arg.ty].inner);
}
- function.arguments.push(arg);
}
+ function.arguments.push(arg);
}
- // 2. call the wrapped function
- let fake_id = !(module.entry_points.len() as u32); // doesn't matter, as long as it's not a collision
- let dummy_handle = self.add_call(fake_id, fun_id);
- function.body.push(
- crate::Statement::Call {
- function: dummy_handle,
- arguments: Vec::new(),
- result: None,
- },
- crate::Span::default(),
- );
-
- // 3. copy the outputs from privates to the result
- let mut members = Vec::new();
- let mut components = Vec::new();
- for &v_id in ep.variable_ids.iter() {
- let lvar = self.lookup_variable.lookup(v_id)?;
- if let super::Variable::Output(ref result) = lvar.inner {
- let span = module.global_variables.get_span(lvar.handle);
- let expr_handle = function
- .expressions
- .append(crate::Expression::GlobalVariable(lvar.handle), span);
+ }
+ // 2. call the wrapped function
+ let fake_id = !(module.entry_points.len() as u32); // doesn't matter, as long as it's not a collision
+ let dummy_handle = self.add_call(fake_id, fun_id);
+ function.body.push(
+ crate::Statement::Call {
+ function: dummy_handle,
+ arguments: Vec::new(),
+ result: None,
+ },
+ crate::Span::default(),
+ );
- // Cull problematic builtins of gl_PerVertex.
- // See the docs for `Frontend::gl_per_vertex_builtin_access`.
+ // 3. copy the outputs from privates to the result
+ let mut members = Vec::new();
+ let mut components = Vec::new();
+ for &v_id in ep.variable_ids.iter() {
+ let lvar = self.lookup_variable.lookup(v_id)?;
+ if let super::Variable::Output(ref result) = lvar.inner {
+ let span = module.global_variables.get_span(lvar.handle);
+ let expr_handle = function
+ .expressions
+ .append(crate::Expression::GlobalVariable(lvar.handle), span);
+
+ // Cull problematic builtins of gl_PerVertex.
+ // See the docs for `Frontend::gl_per_vertex_builtin_access`.
+ {
+ let ty = &module.types[result.ty];
+ if let crate::TypeInner::Struct {
+ members: ref original_members,
+ span,
+ } = ty.inner
{
- let ty = &module.types[result.ty];
- match ty.inner {
- crate::TypeInner::Struct {
- members: ref original_members,
- span,
- } if ty.name.as_deref() == Some("gl_PerVertex") => {
- let mut new_members = original_members.clone();
- for member in &mut new_members {
- if let Some(crate::Binding::BuiltIn(built_in)) = member.binding
- {
- if !self.gl_per_vertex_builtin_access.contains(&built_in) {
- member.binding = None
- }
- }
- }
- if &new_members != original_members {
- module.types.replace(
- result.ty,
- crate::Type {
- name: ty.name.clone(),
- inner: crate::TypeInner::Struct {
- members: new_members,
- span,
- },
- },
- );
+ let mut new_members = None;
+ for (idx, member) in original_members.iter().enumerate() {
+ if let Some(crate::Binding::BuiltIn(built_in)) = member.binding {
+ if !self.gl_per_vertex_builtin_access.contains(&built_in) {
+ new_members.get_or_insert_with(|| original_members.clone())
+ [idx]
+ .binding = None;
}
}
- _ => {}
+ }
+ if let Some(new_members) = new_members {
+ module.types.replace(
+ result.ty,
+ crate::Type {
+ name: ty.name.clone(),
+ inner: crate::TypeInner::Struct {
+ members: new_members,
+ span,
+ },
+ },
+ );
}
}
+ }
- match module.types[result.ty].inner {
- crate::TypeInner::Struct {
- members: ref sub_members,
- ..
- } => {
- for (index, sm) in sub_members.iter().enumerate() {
- if sm.binding.is_none() {
- continue;
- }
- let mut sm = sm.clone();
-
- if let Some(ref mut binding) = sm.binding {
- if ep.stage == crate::ShaderStage::Vertex {
- binding.apply_default_interpolation(
- &module.types[sm.ty].inner,
- );
- }
- }
-
- members.push(sm);
-
- components.push(function.expressions.append(
- crate::Expression::AccessIndex {
- base: expr_handle,
- index: index as u32,
- },
- span,
- ));
+ match module.types[result.ty].inner {
+ crate::TypeInner::Struct {
+ members: ref sub_members,
+ ..
+ } => {
+ for (index, sm) in sub_members.iter().enumerate() {
+ if sm.binding.is_none() {
+ continue;
}
- }
- ref inner => {
- let mut binding = result.binding.clone();
- if let Some(ref mut binding) = binding {
+ let mut sm = sm.clone();
+
+ if let Some(ref mut binding) = sm.binding {
if ep.stage == crate::ShaderStage::Vertex {
- binding.apply_default_interpolation(inner);
+ binding.apply_default_interpolation(&module.types[sm.ty].inner);
}
}
- members.push(crate::StructMember {
- name: None,
- ty: result.ty,
- binding,
- offset: 0,
- });
- // populate just the globals first, then do `Load` in a
- // separate step, so that we can get a range.
- components.push(expr_handle);
+ members.push(sm);
+
+ components.push(function.expressions.append(
+ crate::Expression::AccessIndex {
+ base: expr_handle,
+ index: index as u32,
+ },
+ span,
+ ));
}
}
- }
- }
+ ref inner => {
+ let mut binding = result.binding.clone();
+ if let Some(ref mut binding) = binding {
+ if ep.stage == crate::ShaderStage::Vertex {
+ binding.apply_default_interpolation(inner);
+ }
+ }
- for (member_index, member) in members.iter().enumerate() {
- match member.binding {
- Some(crate::Binding::BuiltIn(crate::BuiltIn::Position { .. }))
- if self.options.adjust_coordinate_space =>
- {
- let mut emitter = Emitter::default();
- emitter.start(&function.expressions);
- let global_expr = components[member_index];
- let span = function.expressions.get_span(global_expr);
- let access_expr = function.expressions.append(
- crate::Expression::AccessIndex {
- base: global_expr,
- index: 1,
- },
- span,
- );
- let load_expr = function.expressions.append(
- crate::Expression::Load {
- pointer: access_expr,
- },
- span,
- );
- let neg_expr = function.expressions.append(
- crate::Expression::Unary {
- op: crate::UnaryOperator::Negate,
- expr: load_expr,
- },
- span,
- );
- function.body.extend(emitter.finish(&function.expressions));
- function.body.push(
- crate::Statement::Store {
- pointer: access_expr,
- value: neg_expr,
- },
- span,
- );
+ members.push(crate::StructMember {
+ name: None,
+ ty: result.ty,
+ binding,
+ offset: 0,
+ });
+ // populate just the globals first, then do `Load` in a
+ // separate step, so that we can get a range.
+ components.push(expr_handle);
}
- _ => {}
}
}
+ }
- let mut emitter = Emitter::default();
- emitter.start(&function.expressions);
- for component in components.iter_mut() {
- let load_expr = crate::Expression::Load {
- pointer: *component,
- };
- let span = function.expressions.get_span(*component);
- *component = function.expressions.append(load_expr, span);
- }
-
- match members[..] {
- [] => {}
- [ref member] => {
- function.body.extend(emitter.finish(&function.expressions));
- let span = function.expressions.get_span(components[0]);
- function.body.push(
- crate::Statement::Return {
- value: components.first().cloned(),
+ for (member_index, member) in members.iter().enumerate() {
+ match member.binding {
+ Some(crate::Binding::BuiltIn(crate::BuiltIn::Position { .. }))
+ if self.options.adjust_coordinate_space =>
+ {
+ let mut emitter = Emitter::default();
+ emitter.start(&function.expressions);
+ let global_expr = components[member_index];
+ let span = function.expressions.get_span(global_expr);
+ let access_expr = function.expressions.append(
+ crate::Expression::AccessIndex {
+ base: global_expr,
+ index: 1,
},
span,
);
- function.result = Some(crate::FunctionResult {
- ty: member.ty,
- binding: member.binding.clone(),
- });
- }
- _ => {
- let span = crate::Span::total_span(
- components.iter().map(|h| function.expressions.get_span(*h)),
+ let load_expr = function.expressions.append(
+ crate::Expression::Load {
+ pointer: access_expr,
+ },
+ span,
);
- let ty = module.types.insert(
- crate::Type {
- name: None,
- inner: crate::TypeInner::Struct {
- members,
- span: 0xFFFF, // shouldn't matter
- },
+ let neg_expr = function.expressions.append(
+ crate::Expression::Unary {
+ op: crate::UnaryOperator::Negate,
+ expr: load_expr,
},
span,
);
- let result_expr = function
- .expressions
- .append(crate::Expression::Compose { ty, components }, span);
function.body.extend(emitter.finish(&function.expressions));
function.body.push(
- crate::Statement::Return {
- value: Some(result_expr),
+ crate::Statement::Store {
+ pointer: access_expr,
+ value: neg_expr,
},
span,
);
- function.result = Some(crate::FunctionResult { ty, binding: None });
}
+ _ => {}
}
+ }
- module.entry_points.push(crate::EntryPoint {
- name: ep.name,
- stage: ep.stage,
- early_depth_test: ep.early_depth_test,
- workgroup_size: ep.workgroup_size,
- function,
- });
+ let mut emitter = Emitter::default();
+ emitter.start(&function.expressions);
+ for component in components.iter_mut() {
+ let load_expr = crate::Expression::Load {
+ pointer: *component,
+ };
+ let span = function.expressions.get_span(*component);
+ *component = function.expressions.append(load_expr, span);
}
+ match members[..] {
+ [] => {}
+ [ref member] => {
+ function.body.extend(emitter.finish(&function.expressions));
+ let span = function.expressions.get_span(components[0]);
+ function.body.push(
+ crate::Statement::Return {
+ value: components.first().cloned(),
+ },
+ span,
+ );
+ function.result = Some(crate::FunctionResult {
+ ty: member.ty,
+ binding: member.binding.clone(),
+ });
+ }
+ _ => {
+ let span = crate::Span::total_span(
+ components.iter().map(|h| function.expressions.get_span(*h)),
+ );
+ let ty = module.types.insert(
+ crate::Type {
+ name: None,
+ inner: crate::TypeInner::Struct {
+ members,
+ span: 0xFFFF, // shouldn't matter
+ },
+ },
+ span,
+ );
+ let result_expr = function
+ .expressions
+ .append(crate::Expression::Compose { ty, components }, span);
+ function.body.extend(emitter.finish(&function.expressions));
+ function.body.push(
+ crate::Statement::Return {
+ value: Some(result_expr),
+ },
+ span,
+ );
+ function.result = Some(crate::FunctionResult { ty, binding: None });
+ }
+ }
+
+ module.entry_points.push(crate::EntryPoint {
+ name: ep.name,
+ stage: ep.stage,
+ early_depth_test: ep.early_depth_test,
+ workgroup_size: ep.workgroup_size,
+ function,
+ });
+
Ok(())
}
}
diff --git a/third_party/rust/naga/src/front/spv/mod.rs b/third_party/rust/naga/src/front/spv/mod.rs
index 8b1c854358..b793448597 100644
--- a/third_party/rust/naga/src/front/spv/mod.rs
+++ b/third_party/rust/naga/src/front/spv/mod.rs
@@ -577,6 +577,9 @@ pub struct Frontend<I> {
lookup_function_type: FastHashMap<spirv::Word, LookupFunctionType>,
lookup_function: FastHashMap<spirv::Word, LookupFunction>,
lookup_entry_point: FastHashMap<spirv::Word, EntryPoint>,
+ // When parsing functions, each entry point function gets an entry here so that additional
+ // processing for them can be performed after all function parsing.
+ deferred_entry_points: Vec<(EntryPoint, spirv::Word)>,
//Note: each `OpFunctionCall` gets a single entry here, indexed by the
// dummy `Handle<crate::Function>` of the call site.
deferred_function_calls: Vec<spirv::Word>,
@@ -628,6 +631,7 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
lookup_function_type: FastHashMap::default(),
lookup_function: FastHashMap::default(),
lookup_entry_point: FastHashMap::default(),
+ deferred_entry_points: Vec::default(),
deferred_function_calls: Vec::default(),
dummy_functions: Arena::new(),
function_call_graph: GraphMap::new(),
@@ -1561,12 +1565,10 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
span,
);
- if ty.name.as_deref() == Some("gl_PerVertex") {
- if let Some(crate::Binding::BuiltIn(built_in)) =
- members[index as usize].binding
- {
- self.gl_per_vertex_builtin_access.insert(built_in);
- }
+ if let Some(crate::Binding::BuiltIn(built_in)) =
+ members[index as usize].binding
+ {
+ self.gl_per_vertex_builtin_access.insert(built_in);
}
AccessExpression {
@@ -3956,6 +3958,12 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
}?;
}
+ // Do entry point specific processing after all functions are parsed so that we can
+ // cull unused problematic builtins of gl_PerVertex.
+ for (ep, fun_id) in core::mem::take(&mut self.deferred_entry_points) {
+ self.process_entry_point(&mut module, ep, fun_id)?;
+ }
+
log::info!("Patching...");
{
let mut nodes = petgraph::algo::toposort(&self.function_call_graph, None)
@@ -4868,6 +4876,11 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
let low = self.next()?;
match width {
4 => crate::Literal::U32(low),
+ 8 => {
+ inst.expect(5)?;
+ let high = self.next()?;
+ crate::Literal::U64(u64::from(high) << 32 | u64::from(low))
+ }
_ => return Err(Error::InvalidTypeWidth(width as u32)),
}
}
@@ -5081,7 +5094,7 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
None
};
let span = self.span_from_with_op(start);
- let mut dec = self.future_decor.remove(&id).unwrap_or_default();
+ let dec = self.future_decor.remove(&id).unwrap_or_default();
let original_ty = self.lookup_type.lookup(type_id)?.handle;
let mut ty = original_ty;
@@ -5127,17 +5140,6 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
None => map_storage_class(storage_class)?,
};
- // Fix empty name for gl_PerVertex struct generated by glslang
- if let crate::TypeInner::Pointer { .. } = module.types[original_ty].inner {
- if ext_class == ExtendedClass::Input || ext_class == ExtendedClass::Output {
- if let Some(ref dec_name) = dec.name {
- if dec_name.is_empty() {
- dec.name = Some("perVertexStruct".to_string())
- }
- }
- }
- }
-
let (inner, var) = match ext_class {
ExtendedClass::Global(mut space) => {
if let crate::AddressSpace::Storage { ref mut access } = space {
@@ -5323,6 +5325,21 @@ pub fn parse_u8_slice(data: &[u8], options: &Options) -> Result<crate::Module, E
Frontend::new(words, options).parse()
}
+/// Helper function to check if `child` is in the scope of `parent`
+fn is_parent(mut child: usize, parent: usize, block_ctx: &BlockContext) -> bool {
+ loop {
+ if child == parent {
+ // The child is in the scope parent
+ break true;
+ } else if child == 0 {
+ // Searched finished at the root the child isn't in the parent's body
+ break false;
+ }
+
+ child = block_ctx.bodies[child].parent;
+ }
+}
+
#[cfg(test)]
mod test {
#[test]
@@ -5339,18 +5356,3 @@ mod test {
let _ = super::parse_u8_slice(&bin, &Default::default()).unwrap();
}
}
-
-/// Helper function to check if `child` is in the scope of `parent`
-fn is_parent(mut child: usize, parent: usize, block_ctx: &BlockContext) -> bool {
- loop {
- if child == parent {
- // The child is in the scope parent
- break true;
- } else if child == 0 {
- // Searched finished at the root the child isn't in the parent's body
- break false;
- }
-
- child = block_ctx.bodies[child].parent;
- }
-}
diff --git a/third_party/rust/naga/src/front/wgsl/error.rs b/third_party/rust/naga/src/front/wgsl/error.rs
index 07e68f8dd9..54aa8296b1 100644
--- a/third_party/rust/naga/src/front/wgsl/error.rs
+++ b/third_party/rust/naga/src/front/wgsl/error.rs
@@ -87,7 +87,7 @@ impl ParseError {
/// Returns a [`SourceLocation`] for the first label in the error message.
pub fn location(&self, source: &str) -> Option<SourceLocation> {
- self.labels.get(0).map(|label| label.0.location(source))
+ self.labels.first().map(|label| label.0.location(source))
}
}
diff --git a/third_party/rust/naga/src/front/wgsl/lower/mod.rs b/third_party/rust/naga/src/front/wgsl/lower/mod.rs
index ba9b49e135..2ca6c182b7 100644
--- a/third_party/rust/naga/src/front/wgsl/lower/mod.rs
+++ b/third_party/rust/naga/src/front/wgsl/lower/mod.rs
@@ -1530,6 +1530,8 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
ast::Literal::Number(Number::F32(f)) => crate::Literal::F32(f),
ast::Literal::Number(Number::I32(i)) => crate::Literal::I32(i),
ast::Literal::Number(Number::U32(u)) => crate::Literal::U32(u),
+ ast::Literal::Number(Number::I64(i)) => crate::Literal::I64(i),
+ ast::Literal::Number(Number::U64(u)) => crate::Literal::U64(u),
ast::Literal::Number(Number::F64(f)) => crate::Literal::F64(f),
ast::Literal::Number(Number::AbstractInt(i)) => crate::Literal::AbstractInt(i),
ast::Literal::Number(Number::AbstractFloat(f)) => {
diff --git a/third_party/rust/naga/src/front/wgsl/parse/conv.rs b/third_party/rust/naga/src/front/wgsl/parse/conv.rs
index 08f1e39285..1a4911a3bd 100644
--- a/third_party/rust/naga/src/front/wgsl/parse/conv.rs
+++ b/third_party/rust/naga/src/front/wgsl/parse/conv.rs
@@ -124,6 +124,14 @@ pub fn get_scalar_type(word: &str) -> Option<Scalar> {
kind: Sk::Uint,
width: 4,
}),
+ "i64" => Some(Scalar {
+ kind: Sk::Sint,
+ width: 8,
+ }),
+ "u64" => Some(Scalar {
+ kind: Sk::Uint,
+ width: 8,
+ }),
"bool" => Some(Scalar {
kind: Sk::Bool,
width: crate::BOOL_WIDTH,
diff --git a/third_party/rust/naga/src/front/wgsl/parse/number.rs b/third_party/rust/naga/src/front/wgsl/parse/number.rs
index 7b09ac59bb..ceb2cb336c 100644
--- a/third_party/rust/naga/src/front/wgsl/parse/number.rs
+++ b/third_party/rust/naga/src/front/wgsl/parse/number.rs
@@ -12,6 +12,10 @@ pub enum Number {
I32(i32),
/// Concrete u32
U32(u32),
+ /// Concrete i64
+ I64(i64),
+ /// Concrete u64
+ U64(u64),
/// Concrete f32
F32(f32),
/// Concrete f64
@@ -31,6 +35,8 @@ enum Kind {
enum IntKind {
I32,
U32,
+ I64,
+ U64,
}
#[derive(Debug)]
@@ -270,6 +276,8 @@ fn parse(input: &str) -> (Result<Number, NumberError>, &str) {
let kind = consume_map!(bytes, [
b'i' => Kind::Int(IntKind::I32),
b'u' => Kind::Int(IntKind::U32),
+ b'l', b'i' => Kind::Int(IntKind::I64),
+ b'l', b'u' => Kind::Int(IntKind::U64),
b'h' => Kind::Float(FloatKind::F16),
b'f' => Kind::Float(FloatKind::F32),
b'l', b'f' => Kind::Float(FloatKind::F64),
@@ -416,5 +424,13 @@ fn parse_int(input: &str, kind: Option<IntKind>, radix: u32) -> Result<Number, N
Ok(num) => Ok(Number::U32(num)),
Err(e) => Err(map_err(e)),
},
+ Some(IntKind::I64) => match i64::from_str_radix(input, radix) {
+ Ok(num) => Ok(Number::I64(num)),
+ Err(e) => Err(map_err(e)),
+ },
+ Some(IntKind::U64) => match u64::from_str_radix(input, radix) {
+ Ok(num) => Ok(Number::U64(num)),
+ Err(e) => Err(map_err(e)),
+ },
}
}
diff --git a/third_party/rust/naga/src/front/wgsl/tests.rs b/third_party/rust/naga/src/front/wgsl/tests.rs
index eb2f8a2eb3..cc3d858317 100644
--- a/third_party/rust/naga/src/front/wgsl/tests.rs
+++ b/third_party/rust/naga/src/front/wgsl/tests.rs
@@ -17,6 +17,7 @@ fn parse_comment() {
#[test]
fn parse_types() {
parse_str("const a : i32 = 2;").unwrap();
+ parse_str("const a : u64 = 2lu;").unwrap();
assert!(parse_str("const a : x32 = 2;").is_err());
parse_str("var t: texture_2d<f32>;").unwrap();
parse_str("var t: texture_cube_array<i32>;").unwrap();
generated by cgit v1.2.3 (git 2.39.1) at 2025-01-05 16:27:58 +0000