[ir] Convert function to Value.

This CL moves the `Function` node to be a `Value` instead of a `Block`.

Bug: tint:1718
Change-Id: I7e67f87fc0a442e0b390fc59f761287382a4b682
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/134301
Reviewed-by: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Commit-Queue: Dan Sinclair <dsinclair@chromium.org>
Reviewed-by: James Price <jrprice@google.com>
This commit is contained in:
dan sinclair 2023-05-24 22:50:03 +00:00 committed by Dawn LUCI CQ
parent a518707db4
commit 56e5fb57ef
13 changed files with 782 additions and 781 deletions

View File

@ -58,7 +58,7 @@ Function* Builder::CreateFunction(Symbol name,
std::optional<std::array<uint32_t, 3>> wg_size) {
TINT_ASSERT(IR, return_type);
auto* ir_func = ir.flow_nodes.Create<Function>(name, return_type, stage, wg_size);
auto* ir_func = ir.values.Create<Function>(name, return_type, stage, wg_size);
ir_func->SetStartTarget(CreateBlock());
ir_func->SetEndTarget(CreateFunctionTerminator());

View File

@ -78,7 +78,7 @@ std::string Debug::AsDotGraph(const Module* mod) {
out << "digraph G {" << std::endl;
for (const auto* func : mod->functions) {
// Cluster each function to label and draw a box around it.
out << "subgraph cluster_" << name_for(func) << " {" << std::endl;
out << "subgraph cluster_" << func->Name().Name() << " {" << std::endl;
out << R"(label=")" << func->Name().Name() << R"(")" << std::endl;
out << name_for(func->StartTarget()) << R"( [label="start"])" << std::endl;
out << name_for(func->EndTarget()) << R"( [label="end"])" << std::endl;

View File

@ -151,7 +151,7 @@ void Disassembler::Walk(const Block* blk) {
void Disassembler::EmitFunction(const Function* func) {
in_function_ = true;
Indent() << "%fn" << IdOf(func) << " = func " << func->Name().Name() << "(";
Indent() << "%" << IdOf(func) << " = func " << func->Name().Name() << "(";
for (auto* p : func->Params()) {
if (p != func->Params().Front()) {
out_ << ", ";

View File

@ -34,19 +34,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Add) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = add %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = add %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -67,14 +67,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Increment) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = add %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = add %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -95,14 +95,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAdd) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = add %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = add %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -115,19 +115,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Subtract) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = sub %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = sub %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -148,14 +148,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Decrement) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:i32 = load %v1
%3:i32 = sub %2, 1i
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:i32 = load %v1
%4:i32 = sub %3, 1i
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -176,14 +176,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundSubtract) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = sub %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = sub %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -196,19 +196,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Multiply) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = mul %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = mul %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -229,14 +229,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundMultiply) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = mul %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = mul %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -249,19 +249,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Div) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = div %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = div %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -282,14 +282,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundDiv) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = div %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = div %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -302,19 +302,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Modulo) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = mod %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = mod %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -335,14 +335,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundModulo) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = mod %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = mod %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -355,19 +355,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_And) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = and %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = and %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -388,14 +388,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAnd) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:bool = load %v1
%3:bool = and %2, false
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:bool = load %v1
%4:bool = and %3, false
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -408,19 +408,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Or) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = or %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = or %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -441,14 +441,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundOr) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:bool = load %v1
%3:bool = or %2, false
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:bool = load %v1
%4:bool = or %3, false
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -461,19 +461,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Xor) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = xor %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = xor %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -494,14 +494,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundXor) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = xor %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = xor %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -514,42 +514,42 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalAnd) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():bool -> %fn2 {
%fn2 = block {
br %fn3 true # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():bool -> %fn1 {
%fn1 = block {
br %fn2 true # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:bool = call my_func
if %3 [t: %fn4, f: %fn5, m: %fn6]
# True block
%fn4 = block {
br %fn6 false
}
# False block
%fn5 = block {
%1:bool = call my_func
if %1 [t: %fn6, f: %fn7, m: %fn8]
# True block
%fn6 = block {
br %fn8 false
br %fn6 %3
}
# False block
# Merge block
%fn6 = block (%4:bool) {
if %4:bool [t: %fn7, f: %fn8, m: %fn9]
# True block
%fn7 = block {
br %fn8 %1
}
# Merge block
%fn8 = block (%2:bool) {
if %2:bool [t: %fn9, f: %fn10, m: %fn11]
# True block
%fn9 = block {
br %fn11
br %fn9
}
# False block
%fn10 = block {
br %fn11
%fn8 = block {
br %fn9
}
# Merge block
%fn11 = block {
br %fn12 # return
%fn9 = block {
br %fn10 # return
}
}
@ -557,7 +557,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalAnd) {
}
%fn12 = func_terminator
%fn10 = func_terminator
}
)");
}
@ -570,42 +570,42 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalOr) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():bool -> %fn2 {
%fn2 = block {
br %fn3 true # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():bool -> %fn1 {
%fn1 = block {
br %fn2 true # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:bool = call my_func
if %3 [t: %fn4, f: %fn5, m: %fn6]
# True block
%fn4 = block {
br %fn6 %3
}
# False block
%fn5 = block {
%1:bool = call my_func
if %1 [t: %fn6, f: %fn7, m: %fn8]
# True block
%fn6 = block {
br %fn8 %1
br %fn6 true
}
# False block
# Merge block
%fn6 = block (%4:bool) {
if %4:bool [t: %fn7, f: %fn8, m: %fn9]
# True block
%fn7 = block {
br %fn8 true
}
# Merge block
%fn8 = block (%2:bool) {
if %2:bool [t: %fn9, f: %fn10, m: %fn11]
# True block
%fn9 = block {
br %fn11
br %fn9
}
# False block
%fn10 = block {
br %fn11
%fn8 = block {
br %fn9
}
# Merge block
%fn11 = block {
br %fn12 # return
%fn9 = block {
br %fn10 # return
}
}
@ -613,7 +613,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalOr) {
}
%fn12 = func_terminator
%fn10 = func_terminator
}
)");
}
@ -626,19 +626,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Equal) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = eq %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = eq %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -651,19 +651,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_NotEqual) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = neq %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = neq %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -676,19 +676,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LessThan) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = lt %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = lt %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -701,19 +701,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_GreaterThan) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = gt %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = gt %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -726,19 +726,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LessThanEqual) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = lte %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = lte %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -751,19 +751,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_GreaterThanEqual) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = gte %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:bool = gte %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -776,19 +776,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_ShiftLeft) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = shiftl %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = shiftl %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -809,14 +809,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftLeft) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = shiftl %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = shiftl %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -829,19 +829,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_ShiftRight) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 0u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 0u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = shiftr %1, 4u
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = shiftr %3, 4u
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -862,14 +862,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftRight) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:u32 = load %v1
%3:u32 = shiftr %2, 1u
store %v1, %3
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = load %v1
%4:u32 = shiftr %3, 1u
store %v1, %4
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -884,40 +884,40 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Compound) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():f32 -> %fn2 {
%fn2 = block {
br %fn3 0.0f # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():f32 -> %fn1 {
%fn1 = block {
br %fn2 0.0f # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:f32 = call my_func
%2:bool = lt %1, 2.0f
if %2 [t: %fn6, f: %fn7, m: %fn8]
# True block
%fn6 = block {
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:f32 = call my_func
%4:f32 = call my_func
%5:f32 = mul 2.29999995231628417969f, %4
%6:f32 = div %3, %5
%7:bool = gt 2.5f, %6
br %fn8 %7
%4:bool = lt %3, 2.0f
if %4 [t: %fn4, f: %fn5, m: %fn6]
# True block
%fn4 = block {
%5:f32 = call my_func
%6:f32 = call my_func
%7:f32 = mul 2.29999995231628417969f, %6
%8:f32 = div %5, %7
%9:bool = gt 2.5f, %8
br %fn6 %9
}
# False block
%fn7 = block {
br %fn8 %2
%fn5 = block {
br %fn6 %4
}
# Merge block
%fn8 = block (%tint_symbol:bool) {
br %fn9 # return
%fn6 = block (%tint_symbol:bool) {
br %fn7 # return
}
}
%fn9 = func_terminator
%fn7 = func_terminator
}
)");
}
@ -931,18 +931,18 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Compound_WithConstEval) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func(%p:bool):bool -> %fn2 {
%fn2 = block {
br %fn3 true # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func(%p:bool):bool -> %fn1 {
%fn1 = block {
br %fn2 true # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%tint_symbol:bool = call my_func, false
br %fn6 # return
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}

View File

@ -35,19 +35,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Bitcast) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():f32 -> %fn2 {
%fn2 = block {
br %fn3 0.0f # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():f32 -> %fn1 {
%fn1 = block {
br %fn2 0.0f # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:f32 = call my_func
%tint_symbol:f32 = bitcast %1
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:f32 = call my_func
%tint_symbol:f32 = bitcast %3
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -62,12 +62,12 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Discard) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@fragment] -> %fn2 {
%fn2 = block {
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func test_function():void [@fragment] -> %fn1 {
%fn1 = block {
discard
br %fn3 # return
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)");
}
@ -80,18 +80,18 @@ TEST_F(IR_BuilderImplTest, EmitStatement_UserFunction) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func(%p:f32):void -> %fn2 {
%fn2 = block {
br %fn3 # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func(%p:f32):void -> %fn1 {
%fn1 = block {
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%2:void = call my_func, 6.0f
br %fn6 # return
%3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%4:void = call my_func, 6.0f
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -112,13 +112,13 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Convert) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:i32 = load %i
%tint_symbol:f32 = convert i32, %2
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:i32 = load %i
%tint_symbol:f32 = convert i32, %3
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -157,13 +157,13 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Construct) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2:f32 = load %i
%tint_symbol:vec3<f32> = construct 2.0f, 3.0f, %2
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:f32 = load %i
%tint_symbol:vec3<f32> = construct 2.0f, 3.0f, %3
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}

View File

@ -34,11 +34,11 @@ TEST_F(IR_BuilderImplTest, EmitExpression_MaterializedCall) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():f32 -> %fn2 {
%fn2 = block {
br %fn3 2.0f # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func test_function():f32 -> %fn1 {
%fn1 = block {
br %fn2 2.0f # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)");
}

View File

@ -43,12 +43,12 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Assign) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
store %a, 4u
br %fn5 # return
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}

File diff suppressed because it is too large Load Diff

View File

@ -34,19 +34,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Not) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():bool -> %fn2 {
%fn2 = block {
br %fn3 false # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():bool -> %fn1 {
%fn1 = block {
br %fn2 false # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:bool = call my_func
%tint_symbol:bool = eq %1, false
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:bool = call my_func
%tint_symbol:bool = eq %3, false
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -59,19 +59,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Complement) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():u32 -> %fn2 {
%fn2 = block {
br %fn3 1u # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():u32 -> %fn1 {
%fn1 = block {
br %fn2 1u # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = complement %1
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:u32 = call my_func
%tint_symbol:u32 = complement %3
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -84,19 +84,19 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Negation) {
auto m = Build();
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func():i32 -> %fn2 {
%fn2 = block {
br %fn3 1i # return
EXPECT_EQ(Disassemble(m.Get()), R"(%1 = func my_func():i32 -> %fn1 {
%fn1 = block {
br %fn2 1i # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5 {
%fn5 = block {
%1:i32 = call my_func
%tint_symbol:i32 = negation %1
br %fn6 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
%3:i32 = call my_func
%tint_symbol:i32 = negation %3
br %fn4 # return
}
%fn6 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -118,11 +118,11 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_AddressOf) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
br %fn5 # return
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}
@ -146,12 +146,12 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Indirection) {
%fn2 = root_terminator
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4 {
%fn4 = block {
%2 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn3 {
%fn3 = block {
store %v3, 42i
br %fn5 # return
br %fn4 # return
}
%fn5 = func_terminator
%fn4 = func_terminator
}
)");
}

View File

@ -69,12 +69,12 @@ TEST_F(IR_BuilderImplTest, Emit_Var_NoInit) {
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()),
R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn2 {
%fn2 = block {
R"(%1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn1 {
%fn1 = block {
%a:ptr<function, u32, read_write> = var
br %fn3 # return
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)");
}
@ -88,12 +88,12 @@ TEST_F(IR_BuilderImplTest, Emit_Var_Init) {
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()),
R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn2 {
%fn2 = block {
R"(%1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn1 {
%fn1 = block {
%a:ptr<function, u32, read_write> = var, 2u
br %fn3 # return
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)");
}

View File

@ -19,8 +19,8 @@
#include <optional>
#include <utility>
#include "src/tint/ir/flow_node.h"
#include "src/tint/ir/function_param.h"
#include "src/tint/ir/value.h"
#include "src/tint/symbol.h"
#include "src/tint/type/type.h"
@ -33,7 +33,7 @@ class FunctionTerminator;
namespace tint::ir {
/// An IR representation of a function
class Function : public utils::Castable<Function, FlowNode> {
class Function : public utils::Castable<Function, Value> {
public:
/// The pipeline stage for an entry point
enum class PipelineStage {

View File

@ -19,6 +19,7 @@
#include "src/tint/constant/value.h"
#include "src/tint/ir/constant.h"
#include "src/tint/ir/flow_node.h"
#include "src/tint/ir/function.h"
#include "src/tint/ir/instruction.h"
#include "src/tint/ir/value.h"

View File

@ -25,11 +25,11 @@ using IR_AddEmptyEntryPointTest = TransformTest;
TEST_F(IR_AddEmptyEntryPointTest, EmptyModule) {
auto* expect = R"(
%fn1 = func unused_entry_point():void [@compute @workgroup_size(1, 1, 1)] -> %fn2 {
%fn2 = block {
br %fn3 # return
%1 = func unused_entry_point():void [@compute @workgroup_size(1, 1, 1)] -> %fn1 {
%fn1 = block {
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)";
@ -44,11 +44,11 @@ TEST_F(IR_AddEmptyEntryPointTest, ExistingEntryPoint) {
mod.functions.Push(ep);
auto* expect = R"(
%fn1 = func main():void [@fragment] -> %fn2 {
%fn2 = block {
br %fn3 # return
%1 = func main():void [@fragment] -> %fn1 {
%fn1 = block {
br %fn2 # return
}
%fn3 = func_terminator
%fn2 = func_terminator
}
)";