encounter
/
dawn-cmake
mirror of https://github.com/encounter/dawn-cmake.git
1
0
Fork 0
Code Issues Packages Projects Releases Wiki Activity

[ir] Remove Jump. 

This CL removes the Jump instruction. This didn't add much value over
the branch instruction and was confusing as to which to use when.

Bug: tint:1718
Change-Id: I69253aa7baf39b00f698e4e8f9608465f6019dcd
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/134221
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 18:45:04 +00:00 committed by Dawn LUCI CQ
parent 24c5ed6b0a
commit e9a4adeff9
15 changed files with 95 additions and 201 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

2
src/tint/BUILD.gn
View File

@ -1242,8 +1242,6 @@ if (tint_build_ir) {
"ir/if.h",
"ir/instruction.cc",
"ir/instruction.h",
"ir/jump.cc",
"ir/jump.h",
"ir/load.cc",
"ir/load.h",
"ir/loop.cc",

2
src/tint/CMakeLists.txt
View File

@ -752,8 +752,6 @@ if(${TINT_BUILD_IR})
ir/if.h
ir/instruction.cc
ir/instruction.h
ir/jump.cc
ir/jump.h
ir/load.cc
ir/load.h
ir/loop.cc

4
src/tint/ir/builder.cc
View File

@ -221,10 +221,6 @@ ir::Branch* Builder::Branch(FlowNode* to, utils::VectorRef<Value*> args) {
return ir.values.Create<ir::Branch>(to, args);
}
ir::Jump* Builder::Jump(FlowNode* to, utils::VectorRef<Value*> args) {
return ir.values.Create<ir::Jump>(to, args);
}
ir::BlockParam* Builder::BlockParam(const type::Type* type) {
return ir.values.Create<ir::BlockParam>(type);
}

7
src/tint/ir/builder.h
View File

@ -30,7 +30,6 @@
#include "src/tint/ir/function_param.h"
#include "src/tint/ir/function_terminator.h"
#include "src/tint/ir/if.h"
#include "src/tint/ir/jump.h"
#include "src/tint/ir/load.h"
#include "src/tint/ir/loop.h"
#include "src/tint/ir/module.h"
@ -405,12 +404,6 @@ class Builder {
/// @returns the instruction
ir::Branch* Branch(FlowNode* to, utils::VectorRef<Value*> args = {});
/// Creates a jump declaration
/// @param to the node being branched too
/// @param args the branch arguments
/// @returns the instruction
ir::Jump* Jump(FlowNode* to, utils::VectorRef<Value*> args = {});
/// Creates a new `BlockParam`
/// @param type the parameter type
/// @returns the value

13
src/tint/ir/disassembler.cc
View File

@ -27,7 +27,6 @@
#include "src/tint/ir/discard.h"
#include "src/tint/ir/function_terminator.h"
#include "src/tint/ir/if.h"
#include "src/tint/ir/jump.h"
#include "src/tint/ir/load.h"
#include "src/tint/ir/loop.h"
#include "src/tint/ir/root_terminator.h"
@ -406,18 +405,8 @@ void Disassembler::EmitSwitch(const Switch* s) {
}
void Disassembler::EmitBranch(const Branch* b) {
if (b->Is<Jump>()) {
out_ << "jmp ";
// Stuff the thing we're jumping too into the front of the walk list so it will be emitted
// next.
walk_list_.push_front(b->To());
} else {
out_ << "br ";
}
std::string suffix = "";
out_ << "%fn" << IdOf(b->To());
out_ << "br %fn" << IdOf(b->To());
if (b->To()->Is<FunctionTerminator>()) {
suffix = "return";
} else if (b->To()->Is<RootTerminator>()) {

22
src/tint/ir/from_program.cc
View File

@ -170,20 +170,6 @@ class Impl {
diagnostics_.add_error(tint::diag::System::IR, err, s);
}
void JumpTo(FlowNode* node, utils::VectorRef<Value*> args = {}) {
TINT_ASSERT(IR, current_flow_block_);
TINT_ASSERT(IR, !current_flow_block_->HasBranchTarget());
current_flow_block_->Instructions().Push(builder_.Jump(node, args));
current_flow_block_ = nullptr;
}
void JumpToIfNeeded(FlowNode* node) {
if (!current_flow_block_ || current_flow_block_->HasBranchTarget()) {
return;
}
JumpTo(node);
}
void BranchTo(FlowNode* node, utils::VectorRef<Value*> args = {}) {
TINT_ASSERT(IR, current_flow_block_);
TINT_ASSERT(IR, !current_flow_block_->HasBranchTarget());
@ -357,7 +343,7 @@ class Impl {
// If the branch target has already been set then a `return` was called. Only set in
// the case where `return` wasn't called.
JumpToIfNeeded(current_function_->EndTarget());
BranchToIfNeeded(current_function_->EndTarget());
}
TINT_ASSERT(IR, control_stack_.IsEmpty());
@ -587,7 +573,7 @@ class Impl {
// The current block didn't `break`, `return` or `continue`, go to the continuing
// block.
JumpToIfNeeded(loop_inst->Continuing());
BranchToIfNeeded(loop_inst->Continuing());
current_flow_block_ = loop_inst->Continuing();
if (stmt->continuing) {
@ -634,7 +620,7 @@ class Impl {
current_flow_block_ = if_inst->Merge();
EmitBlock(stmt->body);
JumpToIfNeeded(loop_inst->Continuing());
BranchToIfNeeded(loop_inst->Continuing());
}
// The while loop always has a path to the Merge().target as the break statement comes
// before anything inside the loop.
@ -678,7 +664,7 @@ class Impl {
}
EmitBlock(stmt->body);
JumpToIfNeeded(loop_inst->Continuing());
BranchToIfNeeded(loop_inst->Continuing());
if (stmt->continuing) {
current_flow_block_ = loop_inst->Continuing();

64
src/tint/ir/from_program_binary_test.cc
View File

@ -44,7 +44,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Add) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = add %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -71,7 +71,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Increment) {
%2:u32 = load %v1
%3:u32 = add %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -98,7 +98,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAdd) {
%2:u32 = load %v1
%3:u32 = add %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -123,7 +123,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Subtract) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = sub %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -150,7 +150,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Decrement) {
%2:i32 = load %v1
%3:i32 = sub %2, 1i
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -177,7 +177,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundSubtract) {
%2:u32 = load %v1
%3:u32 = sub %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -202,7 +202,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Multiply) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = mul %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -229,7 +229,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundMultiply) {
%2:u32 = load %v1
%3:u32 = mul %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -254,7 +254,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Div) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = div %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -281,7 +281,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundDiv) {
%2:u32 = load %v1
%3:u32 = div %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -306,7 +306,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Modulo) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = mod %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -333,7 +333,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundModulo) {
%2:u32 = load %v1
%3:u32 = mod %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -358,7 +358,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_And) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = and %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -385,7 +385,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAnd) {
%2:bool = load %v1
%3:bool = and %2, false
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -410,7 +410,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Or) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = or %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -437,7 +437,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundOr) {
%2:bool = load %v1
%3:bool = or %2, false
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -462,7 +462,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Xor) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = xor %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -489,7 +489,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundXor) {
%2:u32 = load %v1
%3:u32 = xor %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -537,7 +537,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalAnd) {
}
%fn11 = block {
jmp %fn12 # return
br %fn12 # return
}
%fn12 = func_terminator
@ -585,7 +585,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LogicalOr) {
}
%fn11 = block {
jmp %fn12 # return
br %fn12 # return
}
%fn12 = func_terminator
@ -610,7 +610,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Equal) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = eq %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -635,7 +635,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_NotEqual) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = neq %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -660,7 +660,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LessThan) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = lt %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -685,7 +685,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_GreaterThan) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = gt %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -710,7 +710,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_LessThanEqual) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = lte %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -735,7 +735,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_GreaterThanEqual) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:bool = gte %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -760,7 +760,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_ShiftLeft) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = shiftl %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -787,7 +787,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftLeft) {
%2:u32 = load %v1
%3:u32 = shiftl %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -812,7 +812,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_ShiftRight) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = shiftr %1, 4u
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -839,7 +839,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftRight) {
%2:u32 = load %v1
%3:u32 = shiftr %2, 1u
store %v1, %3
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -883,7 +883,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Compound) {
}
%fn8 = block (%tint_symbol:bool) {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
@ -908,7 +908,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_Compound_WithConstEval) {
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5
%fn5 = block {
%tint_symbol:bool = call my_func, false
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator

12
src/tint/ir/from_program_call_test.cc
View File

@ -45,7 +45,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Bitcast) {
%fn5 = block {
%1:f32 = call my_func
%tint_symbol:f32 = bitcast %1
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -65,7 +65,7 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Discard) {
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@fragment] -> %fn2
%fn2 = block {
discard
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator
@ -82,14 +82,14 @@ TEST_F(IR_BuilderImplTest, EmitStatement_UserFunction) {
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func my_func(%p:f32):void -> %fn2
%fn2 = block {
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5
%fn5 = block {
%2:void = call my_func, 6.0f
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -115,7 +115,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Convert) {
%fn4 = block {
%2:i32 = load %i
%tint_symbol:f32 = convert i32, %2
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -158,7 +158,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Construct) {
%fn4 = block {
%2:f32 = load %i
%tint_symbol:vec3<f32> = construct 2.0f, 3.0f, %2
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator

2
src/tint/ir/from_program_store_test.cc
View File

@ -45,7 +45,7 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Assign) {
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4
%fn4 = block {
store %a, 4u
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator

86
src/tint/ir/from_program_test.cc
View File

@ -72,7 +72,7 @@ TEST_F(IR_BuilderImplTest, Func) {
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func f():void -> %fn2
%fn2 = block {
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator
@ -123,7 +123,7 @@ TEST_F(IR_BuilderImplTest, Func_WithMultipleParam) {
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func f(%a:u32, %b:i32, %c:bool):void -> %fn2
%fn2 = block {
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator
@ -173,7 +173,7 @@ TEST_F(IR_BuilderImplTest, IfStatement) {
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -212,7 +212,7 @@ TEST_F(IR_BuilderImplTest, IfStatement_TrueReturns) {
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -251,7 +251,7 @@ TEST_F(IR_BuilderImplTest, IfStatement_FalseReturns) {
br %fn6 # return
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -323,10 +323,8 @@ TEST_F(IR_BuilderImplTest, IfStatement_JumpChainToMerge) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn8
}
@ -339,6 +337,8 @@ TEST_F(IR_BuilderImplTest, IfStatement_JumpChainToMerge) {
br %fn5
}
%fn9 = func_terminator
)");
}
@ -375,7 +375,7 @@ TEST_F(IR_BuilderImplTest, Loop_WithBreak) {
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -421,10 +421,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinue) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn5
}
@ -437,6 +435,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinue) {
br %fn4
}
%fn9 = func_terminator
)");
}
@ -470,7 +470,7 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinuing_BreakIf) {
}
%fn3 = block {
jmp %fn4
br %fn4
}
%fn4 = block {
@ -478,10 +478,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinuing_BreakIf) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn5
}
@ -494,6 +492,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinuing_BreakIf) {
br %fn3
}
%fn9 = func_terminator
)");
}
@ -514,7 +514,7 @@ TEST_F(IR_BuilderImplTest, Loop_Continuing_Body_Scope) {
}
%fn3 = block {
jmp %fn4
br %fn4
}
%fn4 = block {
@ -522,10 +522,8 @@ TEST_F(IR_BuilderImplTest, Loop_Continuing_Body_Scope) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn5
}
@ -538,6 +536,8 @@ TEST_F(IR_BuilderImplTest, Loop_Continuing_Body_Scope) {
br %fn3
}
%fn9 = func_terminator
)");
}
@ -696,7 +696,7 @@ TEST_F(IR_BuilderImplTest, Loop_WithOnlyReturn_ContinuingBreakIf) {
}
%fn12 = block {
jmp %fn6 # return
br %fn6 # return
}
)");
}
@ -739,10 +739,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithIf_BothBranchesBreak) {
}
%fn5 = block {
jmp %fn8 # return
br %fn8 # return
}
%fn8 = func_terminator
%fn6 = block {
br %fn5
}
@ -751,6 +749,8 @@ TEST_F(IR_BuilderImplTest, Loop_WithIf_BothBranchesBreak) {
br %fn5
}
%fn8 = func_terminator
)");
}
@ -786,10 +786,8 @@ TEST_F(IR_BuilderImplTest, Loop_Nested) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
if true [t: %fn10, f: %fn11, m: %fn12]
}
@ -802,6 +800,8 @@ TEST_F(IR_BuilderImplTest, Loop_Nested) {
if true [t: %fn16, f: %fn17, m: %fn18]
}
%fn9 = func_terminator
%fn10 = block {
br %fn8
}
@ -835,7 +835,7 @@ TEST_F(IR_BuilderImplTest, Loop_Nested) {
}
%fn18 = block {
jmp %fn4
br %fn4
}
%fn19 = block {
@ -847,11 +847,11 @@ TEST_F(IR_BuilderImplTest, Loop_Nested) {
}
%fn21 = block {
jmp %fn7
br %fn7
}
%fn22 = block {
jmp %fn23
br %fn23
}
%fn23 = block {
@ -917,10 +917,8 @@ TEST_F(IR_BuilderImplTest, While) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn8
}
@ -930,9 +928,11 @@ TEST_F(IR_BuilderImplTest, While) {
}
%fn8 = block {
jmp %fn4
br %fn4
}
%fn9 = func_terminator
)");
}
@ -976,10 +976,8 @@ TEST_F(IR_BuilderImplTest, While_Return) {
}
%fn5 = block {
jmp %fn9 # return
br %fn9 # return
}
%fn9 = func_terminator
%fn6 = block {
br %fn8
}
@ -991,6 +989,8 @@ TEST_F(IR_BuilderImplTest, While_Return) {
%fn8 = block {
br %fn9 # return
}
%fn9 = func_terminator
)");
}
@ -1068,7 +1068,7 @@ TEST_F(IR_BuilderImplTest, For_NoInitCondOrContinuing) {
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -1132,7 +1132,7 @@ TEST_F(IR_BuilderImplTest, Switch) {
}
%fn6 = block {
jmp %fn7 # return
br %fn7 # return
}
%fn7 = func_terminator
@ -1184,7 +1184,7 @@ TEST_F(IR_BuilderImplTest, Switch_MultiSelector) {
}
%fn4 = block {
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -1224,7 +1224,7 @@ TEST_F(IR_BuilderImplTest, Switch_OnlyDefault) {
}
%fn4 = block {
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -1277,7 +1277,7 @@ TEST_F(IR_BuilderImplTest, Switch_WithBreak) {
}
%fn5 = block {
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -1351,7 +1351,7 @@ TEST_F(IR_BuilderImplTest, Emit_Phony) {
%fn4 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn5
%fn5 = block {
%1:i32 = call b
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator

10
src/tint/ir/from_program_unary_test.cc
View File

@ -44,7 +44,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Not) {
%fn5 = block {
%1:bool = call my_func
%tint_symbol:bool = eq %1, false
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -69,7 +69,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Complement) {
%fn5 = block {
%1:u32 = call my_func
%tint_symbol:u32 = complement %1
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -94,7 +94,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Negation) {
%fn5 = block {
%1:i32 = call my_func
%tint_symbol:i32 = negation %1
jmp %fn6 # return
br %fn6 # return
}
%fn6 = func_terminator
@ -119,7 +119,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_AddressOf) {
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4
%fn4 = block {
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator
@ -147,7 +147,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Indirection) {
%fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn4
%fn4 = block {
store %v3, 42i
jmp %fn5 # return
br %fn5 # return
}
%fn5 = func_terminator

4
src/tint/ir/from_program_var_test.cc
View File

@ -70,7 +70,7 @@ TEST_F(IR_BuilderImplTest, Emit_Var_NoInit) {
R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn2
%fn2 = block {
%a:ptr<function, u32, read_write> = var
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator
@ -89,7 +89,7 @@ TEST_F(IR_BuilderImplTest, Emit_Var_Init) {
R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] -> %fn2
%fn2 = block {
%a:ptr<function, u32, read_write> = var, 2u
jmp %fn3 # return
br %fn3 # return
}
%fn3 = func_terminator

25
src/tint/ir/jump.cc
View File

@ -1,25 +0,0 @@
// Copyright 2023 The Tint Authors.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
#include "src/tint/ir/jump.h"
TINT_INSTANTIATE_TYPEINFO(tint::ir::Jump);
namespace tint::ir {
Jump::Jump(FlowNode* to, utils::VectorRef<Value*> args) : Base(to, args) {}
Jump::~Jump() = default;
} // namespace tint::ir

37
src/tint/ir/jump.h
View File

@ -1,37 +0,0 @@
// Copyright 2023 The Tint Authors.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
#ifndef SRC_TINT_IR_JUMP_H_
#define SRC_TINT_IR_JUMP_H_
#include "src/tint/ir/block.h"
#include "src/tint/ir/branch.h"
#include "src/tint/ir/value.h"
#include "src/tint/utils/castable.h"
namespace tint::ir {
/// A jump instruction. A jump is walk continuing.
class Jump : public utils::Castable<Jump, Branch> {
public:
/// Constructor
/// @param to the block to branch too
/// @param args the branch arguments
explicit Jump(FlowNode* to, utils::VectorRef<Value*> args = {});
~Jump() override;
};
} // namespace tint::ir
#endif // SRC_TINT_IR_JUMP_H_

6
src/tint/ir/to_program.cc
View File

@ -23,7 +23,6 @@
#include "src/tint/ir/function_terminator.h"
#include "src/tint/ir/if.h"
#include "src/tint/ir/instruction.h"
#include "src/tint/ir/jump.h"
#include "src/tint/ir/load.h"
#include "src/tint/ir/module.h"
#include "src/tint/ir/store.h"
@ -137,10 +136,7 @@ class State {
stmts.Push(s);
}
}
if (blk->Branch()->Is<Jump>() && blk->Branch()->To()->Is<Block>()) {
block = blk->Branch()->To()->As<Block>();
return kContinue;
} else if (auto* if_ = blk->Branch()->As<ir::If>()) {
if (auto* if_ = blk->Branch()->As<ir::If>()) {
if (if_->Merge()->HasBranchTarget()) {
block = if_->Merge();
return kContinue;