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

[wgsl-in] implement increment and decrement

Browse Source
This commit is contained in:
teoxoy
2022-03-22 16:07:54 +01:00
committed by Dzmitry Malyshau
parent d62c0838a6
commit 9dddfeec37
9 changed files with 122 additions and 24 deletions
Download Patch File Download Diff File Expand all files Collapse all files

11
src/front/wgsl/lexer.rs
View File

@@ -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())

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

@@ -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,68 @@ 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())
}
(op @ Token::IncrementOperation | op @ Token::DecrementOperation, op_span) => {
let op = match op {
Token::IncrementOperation => Bo::Add,
Token::DecrementOperation => Bo::Subtract,
_ => unreachable!(),
};
// 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)),
};

2
tests/in/operators.wgsl
View File

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

4
tests/in/shadow.wgsl
View File

@@ -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];

4
tests/out/glsl/operators.main.Compute.glsl
View File

@@ -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;
}

4
tests/out/hlsl/operators.hlsl
View File

@@ -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;
}

4
tests/out/msl/operators.msl
View File

@@ -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;
}

40
tests/out/spv/operators.spvasm
View File

@@ -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

4
tests/out/wgsl/operators.wgsl
View File

@@ -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;
}