Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[naga] Handle phony statements properly by treating them as named expressions #6328

Merged
merged 5 commits into from
Sep 27, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
6 changes: 4 additions & 2 deletions naga/src/front/wgsl/lower/mod.rs
Original file line number Diff line number Diff line change
Expand Up @@ -1778,12 +1778,14 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {

return Ok(());
}
ast::StatementKind::Ignore(expr) => {
ast::StatementKind::Phony(expr) => {
let mut emitter = Emitter::default();
emitter.start(&ctx.function.expressions);

let _ = self.expression(expr, &mut ctx.as_expression(block, &mut emitter))?;
let value = self.expression(expr, &mut ctx.as_expression(block, &mut emitter))?;
block.extend(emitter.finish(&ctx.function.expressions));
ctx.named_expressions
.insert(value, ("phony".to_string(), stmt.span));
return Ok(());
}
};
Expand Down
2 changes: 1 addition & 1 deletion naga/src/front/wgsl/parse/ast.rs
Original file line number Diff line number Diff line change
Expand Up @@ -284,7 +284,7 @@ pub enum StatementKind<'a> {
},
Increment(Handle<Expression<'a>>),
Decrement(Handle<Expression<'a>>),
Ignore(Handle<Expression<'a>>),
Phony(Handle<Expression<'a>>),
ConstAssert(Handle<Expression<'a>>),
}

Expand Down
2 changes: 1 addition & 1 deletion naga/src/front/wgsl/parse/mod.rs
Original file line number Diff line number Diff line change
Expand Up @@ -1696,7 +1696,7 @@ impl Parser {
let expr = self.general_expression(lexer, ctx)?;
lexer.expect(Token::Separator(';'))?;

ast::StatementKind::Ignore(expr)
ast::StatementKind::Phony(expr)
}
"let" => {
let _ = lexer.next();
Expand Down
2 changes: 2 additions & 0 deletions naga/tests/in/phony_assignment.ron
Original file line number Diff line number Diff line change
@@ -0,0 +1,2 @@
(
)
18 changes: 18 additions & 0 deletions naga/tests/in/phony_assignment.wgsl
Original file line number Diff line number Diff line change
@@ -0,0 +1,18 @@
@group(0) @binding(0) var<uniform> binding: f32;

fn five() -> i32 {
return 5;
}

@compute @workgroup_size(1) fn main(
@builtin(global_invocation_id) id: vec3<u32>
) {
_ = binding;
_ = binding;
let a = 5;
_ = a;
_ = five();
let b = five();
// check for name collision
let phony = binding;
}
23 changes: 23 additions & 0 deletions naga/tests/out/glsl/phony_assignment.main.Compute.glsl
Original file line number Diff line number Diff line change
@@ -0,0 +1,23 @@
#version 310 es

precision highp float;
precision highp int;

layout(local_size_x = 1, local_size_y = 1, local_size_z = 1) in;

uniform type_block_0Compute { float _group_0_binding_0_cs; };


int five() {
return 5;
}

void main() {
uvec3 id = gl_GlobalInvocationID;
float phony = _group_0_binding_0_cs;
float phony_1 = _group_0_binding_0_cs;
int _e6 = five();
int _e7 = five();
float phony_2 = _group_0_binding_0_cs;
}

16 changes: 16 additions & 0 deletions naga/tests/out/hlsl/phony_assignment.hlsl
Original file line number Diff line number Diff line change
@@ -0,0 +1,16 @@
cbuffer binding : register(b0) { float binding; }

int five()
{
return 5;
}

[numthreads(1, 1, 1)]
void main(uint3 id : SV_DispatchThreadID)
{
float phony = binding;
float phony_1 = binding;
const int _e6 = five();
const int _e7 = five();
float phony_2 = binding;
}
12 changes: 12 additions & 0 deletions naga/tests/out/hlsl/phony_assignment.ron
Original file line number Diff line number Diff line change
@@ -0,0 +1,12 @@
(
vertex:[
],
fragment:[
],
compute:[
(
entry_point:"main",
target_profile:"cs_5_1",
),
],
)
24 changes: 24 additions & 0 deletions naga/tests/out/msl/phony_assignment.msl
Original file line number Diff line number Diff line change
@@ -0,0 +1,24 @@
// language: metal1.0
#include <metal_stdlib>
#include <simd/simd.h>

using metal::uint;


int five(
) {
return 5;
}

struct main_Input {
};
kernel void main_(
metal::uint3 id [[thread_position_in_grid]]
, constant float& binding [[user(fake0)]]
) {
float phony = binding;
float phony_1 = binding;
int _e6 = five();
int _e7 = five();
float phony_2 = binding;
}
48 changes: 48 additions & 0 deletions naga/tests/out/spv/phony_assignment.spvasm
Original file line number Diff line number Diff line change
@@ -0,0 +1,48 @@
; SPIR-V
; Version: 1.1
; Generator: rspirv
; Bound: 30
OpCapability Shader
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint GLCompute %19 "main" %16
OpExecutionMode %19 LocalSize 1 1 1
OpDecorate %7 DescriptorSet 0
OpDecorate %7 Binding 0
OpDecorate %8 Block
OpMemberDecorate %8 0 Offset 0
OpDecorate %16 BuiltIn GlobalInvocationId
%2 = OpTypeVoid
%3 = OpTypeFloat 32
%4 = OpTypeInt 32 1
%6 = OpTypeInt 32 0
%5 = OpTypeVector %6 3
%8 = OpTypeStruct %3
%9 = OpTypePointer Uniform %8
%7 = OpVariable %9 Uniform
%12 = OpTypeFunction %4
%13 = OpConstant %4 5
%17 = OpTypePointer Input %5
%16 = OpVariable %17 Input
%20 = OpTypeFunction %2
%21 = OpTypePointer Uniform %3
%22 = OpConstant %6 0
%11 = OpFunction %4 None %12
%10 = OpLabel
OpBranch %14
%14 = OpLabel
OpReturnValue %13
OpFunctionEnd
%19 = OpFunction %2 None %20
%15 = OpLabel
%18 = OpLoad %5 %16
%23 = OpAccessChain %21 %7 %22
OpBranch %24
%24 = OpLabel
%25 = OpLoad %3 %23
%26 = OpLoad %3 %23
%27 = OpFunctionCall %4 %11
%28 = OpFunctionCall %4 %11
%29 = OpLoad %3 %23
OpReturn
OpFunctionEnd
15 changes: 15 additions & 0 deletions naga/tests/out/wgsl/phony_assignment.wgsl
Original file line number Diff line number Diff line change
@@ -0,0 +1,15 @@
@group(0) @binding(0)
var<uniform> binding: f32;

fn five() -> i32 {
return 5i;
}

@compute @workgroup_size(1, 1, 1)
fn main(@builtin(global_invocation_id) id: vec3<u32>) {
let phony = binding;
let phony_1 = binding;
let _e6 = five();
let _e7 = five();
let phony_2 = binding;
}
4 changes: 4 additions & 0 deletions naga/tests/snapshots.rs
Original file line number Diff line number Diff line change
Expand Up @@ -928,6 +928,10 @@ fn convert_wgsl() {
"cross",
Targets::SPIRV | Targets::METAL | Targets::GLSL | Targets::HLSL | Targets::WGSL,
),
(
"phony_assignment",
Targets::SPIRV | Targets::METAL | Targets::GLSL | Targets::HLSL | Targets::WGSL,
),
];

for &(name, targets) in inputs.iter() {
Expand Down