[wgsl-in] Implement increment and decrement statements

src/front/wgsl/lexer.rs

@@ -385,11 +385,20 @@ fn consume_token(mut input: &str, generic: bool) -> (Token<'_>, &str) {
  match chars.next() {
  Some('>') => (Token::Arrow, chars.as_str()),
  Some('0'..='9') | Some('.') => consume_number(input),
+ Some('-') => (Token::DecrementOperation, chars.as_str()),
  Some('=') => (Token::AssignmentOperation(cur), chars.as_str()),
  _ => (Token::Operation(cur), sub_input),
  }
  }
- '+' | '*' | '%' | '^' => {
+ '+' => {
+ input = chars.as_str();
+ match chars.next() {
+ Some('+') => (Token::IncrementOperation, chars.as_str()),
+ Some('=') => (Token::AssignmentOperation(cur), chars.as_str()),
+ _ => (Token::Operation(cur), input),
+ }
+ }
+ '*' | '%' | '^' => {
  input = chars.as_str();
  if chars.next() == Some('=') {
  (Token::AssignmentOperation(cur), chars.as_str())

src/front/wgsl/mod.rs

@@ -67,6 +67,8 @@ pub enum Token<'a> {
  LogicalOperation(char),
  ShiftOperation(char),
  AssignmentOperation(char),
+ IncrementOperation,
+ DecrementOperation,
  Arrow,
  Unknown(char),
  UnterminatedString,
@@ -142,6 +144,7 @@ pub enum Error<'a> {
  kind: crate::ScalarKind,
  width: u8,
  },
+ BadIncrDecrReferenceType(Span),
  InvalidResolve(ResolveError),
  InvalidForInitializer(Span),
  InvalidGatherComponent(Span, i32),
@@ -196,6 +199,8 @@ impl<'a> Error<'a> {
  Token::ShiftOperation(c) => format!("bitshift ('{}{}')", c, c),
  Token::AssignmentOperation(c) if c=='<' || c=='>' => format!("bitshift ('{}{}=')", c, c),
  Token::AssignmentOperation(c) => format!("operation ('{}=')", c),
+ Token::IncrementOperation => "increment operation".to_string(),
+ Token::DecrementOperation => "decrement operation".to_string(),
  Token::Arrow => "->".to_string(),
  Token::Unknown(c) => format!("unknown ('{}')", c),
  Token::UnterminatedString => "unterminated string".to_string(),
@@ -317,6 +322,11 @@ impl<'a> Error<'a> {
  labels: vec![(span.clone(), "must be one of f32, i32 or u32".into())],
  notes: vec![],
  },
+ Error::BadIncrDecrReferenceType(ref span) => ParseError {
+ message: "increment/decrement operation requires reference type to be one of i32 or u32".to_string(),
+ labels: vec![(span.clone(), "must be a reference type of i32 or u32".into())],
+ notes: vec![],
+ },
  Error::BadTexture(ref bad_span) => ParseError {
  message: format!("expected an image, but found '{}' which is not an image", &source[bad_span.clone()]),
  labels: vec![(bad_span.clone(), "not an image".into())],
@@ -3252,7 +3262,7 @@ impl Parser {
  Ok((handle, storage_access))
  }
 
- /// Parse a assignment statement
+ /// Parse an assignment statement (will also parse increment and decrement statements)
  fn parse_assignment_statement<'a, 'out>(
  &mut self,
  lexer: &mut Lexer<'a>,
@@ -3264,11 +3274,11 @@ impl Parser {
  context.emitter.start(context.expressions);
  let reference = self.parse_unary_expression(lexer, context.reborrow())?;
  // The left hand side of an assignment must be a reference.
+ let lhs_span = span_start..lexer.current_byte_offset();
  if !reference.is_reference {
- let span = span_start..lexer.current_byte_offset();
  return Err(Error::NotReference(
  "the left-hand side of an assignment",
- span,
+ lhs_span,
  ));
  }
 
@@ -3295,13 +3305,69 @@ impl Parser {
  crate::Expression::Load {
  pointer: reference.handle,
  },
- NagaSpan::from(span_start..lexer.current_byte_offset()),
+ lhs_span.into(),
  );
  let right = self.parse_general_expression(lexer, context.reborrow())?;
  context
  .expressions
  .append(crate::Expression::Binary { op, left, right }, span.into())
  }
+ token @ (Token::IncrementOperation, _) | token @ (Token::DecrementOperation, _) => {
+ let op = match token.0 {
+ Token::IncrementOperation => Bo::Add,
+ Token::DecrementOperation => Bo::Subtract,
+ _ => unreachable!(),
+ };
+ let op_span = token.1;
+
+ // prepare the typifier, but work around mutable borrowing...
+ let _ = context.resolve_type(reference.handle)?;
+
+ let ty = context.typifier.get(reference.handle, context.types);
+ let constant_inner = match ty.canonical_form(context.types) {
+ Some(crate::TypeInner::ValuePointer {
+ size: None,
+ kind,
+ width,
+ space: _,
+ }) => crate::ConstantInner::Scalar {
+ width,
+ value: match kind {
+ crate::ScalarKind::Sint => crate::ScalarValue::Sint(1),
+ crate::ScalarKind::Uint => crate::ScalarValue::Uint(1),
+ _ => {
+ return Err(Error::BadIncrDecrReferenceType(lhs_span));
+ }
+ },
+ },
+ _ => {
+ return Err(Error::BadIncrDecrReferenceType(lhs_span));
+ }
+ };
+ let constant = context.constants.append(
+ crate::Constant {
+ name: None,
+ specialization: None,
+ inner: constant_inner,
+ },
+ crate::Span::default(),
+ );
+
+ let left = context.expressions.append(
+ crate::Expression::Load {
+ pointer: reference.handle,
+ },
+ lhs_span.into(),
+ );
+ let right = context.interrupt_emitter(
+ crate::Expression::Constant(constant),
+ crate::Span::default(),
+ );
+ context.expressions.append(
+ crate::Expression::Binary { op, left, right },
+ op_span.into(),
+ )
+ }
  other => return Err(Error::Unexpected(other, ExpectedToken::SwitchItem)),
  };
 

tests/in/operators.wgsl

@@ -94,6 +94,8 @@ fn binary_assignment() {
  a %= 1;
  a ^= 0;
  a &= 0;
+ a++;
+ a--;
 }
 
 @stage(compute) @workgroup_size(1)

tests/in/shadow.wgsl

@@ -85,7 +85,7 @@ fn fs_main(in: VertexOutput) -> @location(0) vec4<f32> {
  let normal = normalize(in.world_normal);
  // accumulate color
  var color: vec3<f32> = c_ambient;
- for(var i = 0u; i < min(u_globals.num_lights.x, c_max_lights); i += 1u) {
+ for(var i = 0u; i < min(u_globals.num_lights.x, c_max_lights); i++) {
  let light = s_lights[i];
  // project into the light space
  let shadow = fetch_shadow(i, light.proj * in.world_position);
@@ -104,7 +104,7 @@ fn fs_main(in: VertexOutput) -> @location(0) vec4<f32> {
 fn fs_main_without_storage(in: VertexOutput) -> @location(0) vec4<f32> {
  let normal = normalize(in.world_normal);
  var color: vec3<f32> = c_ambient;
- for(var i = 0u; i < min(u_globals.num_lights.x, c_max_lights); i += 1u) {
+ for(var i = 0u; i < min(u_globals.num_lights.x, c_max_lights); i++) {
  // This line is the only difference from the entrypoint above. It uses the lights
  // uniform instead of the lights storage buffer
  let light = u_lights[i];

tests/out/glsl/operators.main.Compute.glsl

@@ -85,6 +85,10 @@ void binary_assignment() {
  a = (_e21 ^ 0);
  int _e24 = a;
  a = (_e24 & 0);
+ int _e27 = a;
+ a = (_e27 + 1);
+ int _e30 = a;
+ a = (_e30 - 1);
  return;
 }
 

tests/out/hlsl/operators.hlsl

@@ -101,6 +101,10 @@ void binary_assignment()
  a = (_expr21 ^ 0);
  int _expr24 = a;
  a = (_expr24 & 0);
+ int _expr27 = a;
+ a = (_expr27 + 1);
+ int _expr30 = a;
+ a = (_expr30 - 1);
  return;
 }
 

tests/out/msl/operators.msl

@@ -98,6 +98,10 @@ void binary_assignment(
  a = _e21 ^ 0;
  int _e24 = a;
  a = _e24 & 0;
+ int _e27 = a;
+ a = _e27 + 1;
+ int _e30 = a;
+ a = _e30 - 1;
  return;
 }
 

tests/out/spv/operators.spvasm

@@ -1,12 +1,12 @@
 ; SPIR-V
 ; Version: 1.1
 ; Generator: rspirv
-; Bound: 176
+; Bound: 180
 OpCapability Shader
 %1 = OpExtInstImport "GLSL.std.450"
 OpMemoryModel Logical GLSL450
-OpEntryPoint GLCompute %164 "main"
-OpExecutionMode %164 LocalSize 1 1 1
+OpEntryPoint GLCompute %168 "main"
+OpExecutionMode %168 LocalSize 1 1 1
 OpMemberDecorate %23 0 Offset 0
 OpMemberDecorate %23 1 Offset 16
 %2 = OpTypeVoid
@@ -207,21 +207,27 @@ OpStore %142 %160
 %161 = OpLoad %8 %142
 %162 = OpBitwiseAnd %8 %161 %11
 OpStore %142 %162
+%163 = OpLoad %8 %142
+%164 = OpIAdd %8 %163 %7
+OpStore %142 %164
+%165 = OpLoad %8 %142
+%166 = OpISub %8 %165 %7
+OpStore %142 %166
 OpReturn
 OpFunctionEnd
-%164 = OpFunction %2 None %117
-%163 = OpLabel
-OpBranch %165
-%165 = OpLabel
-%166 = OpFunctionCall %19 %32
-%167 = OpFunctionCall %19 %57
-%168 = OpFunctionCall %8 %73
-%169 = OpVectorShuffle %22 %27 %27 0 1 2
-%170 = OpFunctionCall %22 %84 %169
-%171 = OpFunctionCall %4 %96
-%172 = OpFunctionCall %2 %116
-%173 = OpFunctionCall %2 %129
-%174 = OpFunctionCall %2 %138
-%175 = OpFunctionCall %2 %145
+%168 = OpFunction %2 None %117
+%167 = OpLabel
+OpBranch %169
+%169 = OpLabel
+%170 = OpFunctionCall %19 %32
+%171 = OpFunctionCall %19 %57
+%172 = OpFunctionCall %8 %73
+%173 = OpVectorShuffle %22 %27 %27 0 1 2
+%174 = OpFunctionCall %22 %84 %173
+%175 = OpFunctionCall %4 %96
+%176 = OpFunctionCall %2 %116
+%177 = OpFunctionCall %2 %129
+%178 = OpFunctionCall %2 %138
+%179 = OpFunctionCall %2 %145
 OpReturn
 OpFunctionEnd

tests/out/wgsl/operators.wgsl

@@ -84,6 +84,10 @@ fn binary_assignment() {
  a = (_e21 ^ 0);
  let _e24 = a;
  a = (_e24 & 0);
+ let _e27 = a;
+ a = (_e27 + 1);
+ let _e30 = a;
+ a = (_e30 - 1);
  return;
 }