From 24cb81116d30d0eee37ed5897c5a65df4baf31a0 Mon Sep 17 00:00:00 2001 From: dan sinclair Date: Thu, 18 May 2023 22:16:08 +0000 Subject: [PATCH] [ir] Shift back to accessors. Some of the IR classes require setters in order to update dependant information. In order to keep the IR access symmetrical this CL moves the IR back to using accessors and private fields. Bug: tint:1718 Change-Id: I101edda004671e07c4594bdcae4b1576e5771782 Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/133640 Reviewed-by: Ben Clayton Kokoro: Kokoro Commit-Queue: Dan Sinclair --- src/tint/BUILD.gn | 1 + src/tint/CMakeLists.txt | 1 + src/tint/ir/binary.cc | 4 +- src/tint/ir/binary.h | 19 +- src/tint/ir/binary_test.cc | 107 ++- src/tint/ir/bitcast_test.cc | 16 +- src/tint/ir/block.cc | 9 + src/tint/ir/block.h | 48 +- src/tint/ir/block_param.cc | 2 +- src/tint/ir/block_param.h | 5 +- src/tint/ir/builder.cc | 51 +- src/tint/ir/builder.h | 10 +- src/tint/ir/call.cc | 4 +- src/tint/ir/call.h | 13 +- src/tint/ir/constant.cc | 2 +- src/tint/ir/constant.h | 16 +- src/tint/ir/constant_test.cc | 62 +- src/tint/ir/debug.cc | 74 +- src/tint/ir/disassembler.cc | 148 ++-- src/tint/ir/flow_node.h | 28 +- src/tint/ir/from_program.cc | 210 +++--- src/tint/ir/from_program_literal_test.cc | 81 ++- src/tint/ir/from_program_test.cc | 682 +++++++++--------- src/tint/ir/function.cc | 4 +- src/tint/ir/function.h | 88 ++- src/tint/ir/function_param.cc | 2 +- src/tint/ir/function_param.h | 5 +- src/tint/ir/if.cc | 2 +- src/tint/ir/if.h | 33 +- src/tint/ir/load.cc | 8 +- src/tint/ir/load.h | 11 +- src/tint/ir/load_test.cc | 12 +- src/tint/ir/loop.h | 32 +- src/tint/ir/root_terminator.h | 5 + src/tint/ir/store.cc | 10 +- src/tint/ir/store.h | 13 +- src/tint/ir/store_test.cc | 18 +- src/tint/ir/switch.cc | 2 +- src/tint/ir/switch.h | 31 +- src/tint/ir/test_helper.h | 1 - src/tint/ir/to_program.cc | 73 +- .../ir/transform/add_empty_entry_point.cc | 4 +- .../transform/add_empty_entry_point_test.cc | 2 +- src/tint/ir/unary.cc | 4 +- src/tint/ir/unary.h | 15 +- src/tint/ir/unary_test.cc | 10 +- src/tint/ir/user_call.cc | 2 +- src/tint/ir/user_call.h | 7 +- src/tint/ir/var.cc | 8 +- src/tint/ir/var.h | 14 +- src/tint/transform/manager_test.cc | 21 +- src/tint/writer/spirv/ir/generator_impl_ir.cc | 54 +- .../spirv/ir/generator_impl_ir_binary_test.cc | 69 +- .../ir/generator_impl_ir_constant_test.cc | 14 +- .../ir/generator_impl_ir_function_test.cc | 16 +- .../spirv/ir/generator_impl_ir_if_test.cc | 44 +- .../spirv/ir/generator_impl_ir_var_test.cc | 42 +- 57 files changed, 1219 insertions(+), 1050 deletions(-) diff --git a/src/tint/BUILD.gn b/src/tint/BUILD.gn index 1358b9c2ff..64fde4cea4 100644 --- a/src/tint/BUILD.gn +++ b/src/tint/BUILD.gn @@ -2265,6 +2265,7 @@ if (tint_build_unittests) { "ir/store_test.cc", "ir/test_helper.h", "ir/to_program_roundtrip_test.cc", + "ir/transform/add_empty_entry_point_test.cc", "ir/unary_test.cc", ] diff --git a/src/tint/CMakeLists.txt b/src/tint/CMakeLists.txt index a696399be3..8ec573a376 100644 --- a/src/tint/CMakeLists.txt +++ b/src/tint/CMakeLists.txt @@ -1476,6 +1476,7 @@ if(TINT_BUILD_TESTS) ir/module_test.cc ir/store_test.cc ir/test_helper.h + ir/transform/add_empty_entry_point_test.cc ir/unary_test.cc ) endif() diff --git a/src/tint/ir/binary.cc b/src/tint/ir/binary.cc index af28306007..2b179acc36 100644 --- a/src/tint/ir/binary.cc +++ b/src/tint/ir/binary.cc @@ -19,8 +19,8 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Binary); namespace tint::ir { -Binary::Binary(Kind k, const type::Type* res_ty, Value* lhs, Value* rhs) - : kind(k), result_type(res_ty), lhs_(lhs), rhs_(rhs) { +Binary::Binary(enum Kind kind, const type::Type* res_ty, Value* lhs, Value* rhs) + : kind_(kind), result_type_(res_ty), lhs_(lhs), rhs_(rhs) { TINT_ASSERT(IR, lhs); TINT_ASSERT(IR, rhs); lhs_->AddUsage(this); diff --git a/src/tint/ir/binary.h b/src/tint/ir/binary.h index a6dc722262..7ea6c08de4 100644 --- a/src/tint/ir/binary.h +++ b/src/tint/ir/binary.h @@ -51,7 +51,7 @@ class Binary : public utils::Castable { /// @param type the result type /// @param lhs the lhs of the instruction /// @param rhs the rhs of the instruction - Binary(Kind kind, const type::Type* type, Value* lhs, Value* rhs); + Binary(enum Kind kind, const type::Type* type, Value* lhs, Value* rhs); Binary(const Binary& inst) = delete; Binary(Binary&& inst) = delete; ~Binary() override; @@ -59,8 +59,11 @@ class Binary : public utils::Castable { Binary& operator=(const Binary& inst) = delete; Binary& operator=(Binary&& inst) = delete; + /// @returns the kind of the binary instruction + enum Kind Kind() const { return kind_; } + /// @returns the type of the value - const type::Type* Type() const override { return result_type; } + const type::Type* Type() const override { return result_type_; } /// @returns the left-hand-side value for the instruction const Value* LHS() const { return lhs_; } @@ -68,15 +71,11 @@ class Binary : public utils::Castable { /// @returns the right-hand-side value for the instruction const Value* RHS() const { return rhs_; } - /// the kind of binary instruction - Kind kind = Kind::kAdd; - - /// the result type of the instruction - const type::Type* result_type = nullptr; - private: - Value* lhs_ = nullptr; - Value* rhs_ = nullptr; + enum Kind kind_; + const type::Type* result_type_; + Value* lhs_; + Value* rhs_; }; } // namespace tint::ir diff --git a/src/tint/ir/binary_test.cc b/src/tint/ir/binary_test.cc index 2968d283a3..dc359781c3 100644 --- a/src/tint/ir/binary_test.cc +++ b/src/tint/ir/binary_test.cc @@ -30,17 +30,16 @@ TEST_F(IR_InstructionTest, CreateAnd) { const auto* inst = b.And(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kAnd); - ASSERT_NE(inst->result_type, nullptr); + EXPECT_EQ(inst->Kind(), Binary::Kind::kAnd); ASSERT_NE(inst->Type(), nullptr); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -52,15 +51,15 @@ TEST_F(IR_InstructionTest, CreateOr) { const auto* inst = b.Or(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kOr); + EXPECT_EQ(inst->Kind(), Binary::Kind::kOr); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -72,15 +71,15 @@ TEST_F(IR_InstructionTest, CreateXor) { const auto* inst = b.Xor(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kXor); + EXPECT_EQ(inst->Kind(), Binary::Kind::kXor); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -92,15 +91,15 @@ TEST_F(IR_InstructionTest, CreateEqual) { const auto* inst = b.Equal(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kEqual); + EXPECT_EQ(inst->Kind(), Binary::Kind::kEqual); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -112,15 +111,15 @@ TEST_F(IR_InstructionTest, CreateNotEqual) { const auto* inst = b.NotEqual(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kNotEqual); + EXPECT_EQ(inst->Kind(), Binary::Kind::kNotEqual); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -132,15 +131,15 @@ TEST_F(IR_InstructionTest, CreateLessThan) { const auto* inst = b.LessThan(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kLessThan); + EXPECT_EQ(inst->Kind(), Binary::Kind::kLessThan); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -153,15 +152,15 @@ TEST_F(IR_InstructionTest, CreateGreaterThan) { b.GreaterThan(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kGreaterThan); + EXPECT_EQ(inst->Kind(), Binary::Kind::kGreaterThan); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -174,15 +173,15 @@ TEST_F(IR_InstructionTest, CreateLessThanEqual) { b.LessThanEqual(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kLessThanEqual); + EXPECT_EQ(inst->Kind(), Binary::Kind::kLessThanEqual); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -195,15 +194,15 @@ TEST_F(IR_InstructionTest, CreateGreaterThanEqual) { b.GreaterThanEqual(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kGreaterThanEqual); + EXPECT_EQ(inst->Kind(), Binary::Kind::kGreaterThanEqual); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -214,15 +213,15 @@ TEST_F(IR_InstructionTest, CreateNot) { const auto* inst = b.Not(b.ir.types.Get(), b.Constant(true)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kEqual); + EXPECT_EQ(inst->Kind(), Binary::Kind::kEqual); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_TRUE(lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_FALSE(rhs->As>()->ValueAs()); } @@ -234,15 +233,15 @@ TEST_F(IR_InstructionTest, CreateShiftLeft) { const auto* inst = b.ShiftLeft(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kShiftLeft); + EXPECT_EQ(inst->Kind(), Binary::Kind::kShiftLeft); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -254,15 +253,15 @@ TEST_F(IR_InstructionTest, CreateShiftRight) { const auto* inst = b.ShiftRight(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kShiftRight); + EXPECT_EQ(inst->Kind(), Binary::Kind::kShiftRight); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -274,15 +273,15 @@ TEST_F(IR_InstructionTest, CreateAdd) { const auto* inst = b.Add(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kAdd); + EXPECT_EQ(inst->Kind(), Binary::Kind::kAdd); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -294,15 +293,15 @@ TEST_F(IR_InstructionTest, CreateSubtract) { const auto* inst = b.Subtract(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kSubtract); + EXPECT_EQ(inst->Kind(), Binary::Kind::kSubtract); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -314,15 +313,15 @@ TEST_F(IR_InstructionTest, CreateMultiply) { const auto* inst = b.Multiply(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kMultiply); + EXPECT_EQ(inst->Kind(), Binary::Kind::kMultiply); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -334,15 +333,15 @@ TEST_F(IR_InstructionTest, CreateDivide) { const auto* inst = b.Divide(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kDivide); + EXPECT_EQ(inst->Kind(), Binary::Kind::kDivide); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -354,15 +353,15 @@ TEST_F(IR_InstructionTest, CreateModulo) { const auto* inst = b.Modulo(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Binary::Kind::kModulo); + EXPECT_EQ(inst->Kind(), Binary::Kind::kModulo); ASSERT_TRUE(inst->LHS()->Is()); - auto lhs = inst->LHS()->As()->value; + auto lhs = inst->LHS()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); ASSERT_TRUE(inst->RHS()->Is()); - auto rhs = inst->RHS()->As()->value; + auto rhs = inst->RHS()->As()->Value(); ASSERT_TRUE(rhs->Is>()); EXPECT_EQ(2_i, rhs->As>()->ValueAs()); } @@ -372,7 +371,7 @@ TEST_F(IR_InstructionTest, Binary_Usage) { Builder b{mod}; const auto* inst = b.And(b.ir.types.Get(), b.Constant(4_i), b.Constant(2_i)); - EXPECT_EQ(inst->kind, Binary::Kind::kAnd); + EXPECT_EQ(inst->Kind(), Binary::Kind::kAnd); ASSERT_NE(inst->LHS(), nullptr); ASSERT_EQ(inst->LHS()->Usage().Length(), 1u); @@ -389,7 +388,7 @@ TEST_F(IR_InstructionTest, Binary_Usage_DuplicateValue) { auto val = b.Constant(4_i); const auto* inst = b.And(b.ir.types.Get(), val, val); - EXPECT_EQ(inst->kind, Binary::Kind::kAnd); + EXPECT_EQ(inst->Kind(), Binary::Kind::kAnd); ASSERT_EQ(inst->LHS(), inst->RHS()); ASSERT_NE(inst->LHS(), nullptr); diff --git a/src/tint/ir/bitcast_test.cc b/src/tint/ir/bitcast_test.cc index e78b734795..6eda56216c 100644 --- a/src/tint/ir/bitcast_test.cc +++ b/src/tint/ir/bitcast_test.cc @@ -32,9 +32,10 @@ TEST_F(IR_InstructionTest, Bitcast) { ASSERT_TRUE(inst->Is()); ASSERT_NE(inst->Type(), nullptr); - ASSERT_EQ(inst->args.Length(), 1u); - ASSERT_TRUE(inst->args[0]->Is()); - auto val = inst->args[0]->As()->value; + const auto args = inst->Args(); + ASSERT_EQ(args.Length(), 1u); + ASSERT_TRUE(args[0]->Is()); + auto val = args[0]->As()->Value(); ASSERT_TRUE(val->Is>()); EXPECT_EQ(4_i, val->As>()->ValueAs()); } @@ -44,10 +45,11 @@ TEST_F(IR_InstructionTest, Bitcast_Usage) { Builder b{mod}; const auto* inst = b.Bitcast(b.ir.types.Get(), b.Constant(4_i)); - ASSERT_EQ(inst->args.Length(), 1u); - ASSERT_NE(inst->args[0], nullptr); - ASSERT_EQ(inst->args[0]->Usage().Length(), 1u); - EXPECT_EQ(inst->args[0]->Usage()[0], inst); + const auto args = inst->Args(); + ASSERT_EQ(args.Length(), 1u); + ASSERT_NE(args[0], nullptr); + ASSERT_EQ(args[0]->Usage().Length(), 1u); + EXPECT_EQ(args[0]->Usage()[0], inst); } } // namespace diff --git a/src/tint/ir/block.cc b/src/tint/ir/block.cc index 2030c19c66..f5b5bfb077 100644 --- a/src/tint/ir/block.cc +++ b/src/tint/ir/block.cc @@ -22,4 +22,13 @@ Block::Block() : Base() {} Block::~Block() = default; +void Block::BranchTo(FlowNode* to, utils::VectorRef args) { + TINT_ASSERT(IR, to); + branch_.target = to; + branch_.args = args; + if (to) { + to->AddInboundBranch(this); + } +} + } // namespace tint::ir diff --git a/src/tint/ir/block.h b/src/tint/ir/block.h index 5af32fedc1..2b01fd00b1 100644 --- a/src/tint/ir/block.h +++ b/src/tint/ir/block.h @@ -15,6 +15,8 @@ #ifndef SRC_TINT_IR_BLOCK_H_ #define SRC_TINT_IR_BLOCK_H_ +#include + #include "src/tint/ir/block_param.h" #include "src/tint/ir/branch.h" #include "src/tint/ir/flow_node.h" @@ -30,20 +32,48 @@ class Block : public utils::Castable { public: /// Constructor Block(); + Block(const Block&) = delete; + Block(Block&&) = delete; ~Block() override; - /// @returns true if this is a dead block. This can happen in the case like a loop merge block - /// which is never reached. - bool IsDead() const override { return branch.target == nullptr; } + Block& operator=(const Block&) = delete; + Block& operator=(Block&&) = delete; - /// The node this block branches too. - Branch branch = {}; + /// Sets the blocks branch target to the given node. + /// @param to the node to branch too + /// @param args the branch arguments + void BranchTo(FlowNode* to, utils::VectorRef args = {}); - /// The instructions in the block - utils::Vector instructions; + /// @returns true if this is block has a branch target set + bool HasBranchTarget() const override { return branch_.target != nullptr; } - /// The parameters passed into the block - utils::Vector params; + /// @return the node this block branches too. + const ir::Branch& Branch() const { return branch_; } + + /// Sets the instructions in the block + /// @param instructions the instructions to set + void SetInstructions(utils::VectorRef instructions) { + instructions_ = std::move(instructions); + } + + /// @returns the instructions in the block + utils::VectorRef Instructions() const { return instructions_; } + /// @returns the instructions in the block + utils::Vector& Instructions() { return instructions_; } + + /// Sets the params to the block + /// @param params the params for the block + void SetParams(utils::VectorRef params) { params_ = std::move(params); } + /// @returns the params to the block + utils::Vector& Params() { return params_; } + + /// @return the parameters passed into the block + utils::VectorRef Params() const { return params_; } + + private: + ir::Branch branch_ = {}; + utils::Vector instructions_; + utils::Vector params_; }; } // namespace tint::ir diff --git a/src/tint/ir/block_param.cc b/src/tint/ir/block_param.cc index f014d196ac..a3a0be5d74 100644 --- a/src/tint/ir/block_param.cc +++ b/src/tint/ir/block_param.cc @@ -18,7 +18,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::BlockParam); namespace tint::ir { -BlockParam::BlockParam(const type::Type* ty) : type(ty) {} +BlockParam::BlockParam(const type::Type* ty) : type_(ty) {} BlockParam::~BlockParam() = default; diff --git a/src/tint/ir/block_param.h b/src/tint/ir/block_param.h index 8ba68a75d5..036ddbf22b 100644 --- a/src/tint/ir/block_param.h +++ b/src/tint/ir/block_param.h @@ -34,10 +34,11 @@ class BlockParam : public utils::Castable { BlockParam& operator=(BlockParam&& inst) = delete; /// @returns the type of the var - const type::Type* Type() const override { return type; } + const type::Type* Type() const override { return type_; } + private: /// the result type of the instruction - const type::Type* type = nullptr; + const type::Type* type_; }; } // namespace tint::ir diff --git a/src/tint/ir/builder.cc b/src/tint/ir/builder.cc index af3c93b053..b0b04d676d 100644 --- a/src/tint/ir/builder.cc +++ b/src/tint/ir/builder.cc @@ -32,7 +32,7 @@ ir::Block* Builder::CreateRootBlockIfNeeded() { // Everything in the module scope must have been const-eval's, so everything will go into a // single block. So, we can create the root terminator for the root-block now. - ir.root_block->branch.target = CreateRootTerminator(); + ir.root_block->BranchTo(CreateRootTerminator()); } return ir.root_block; } @@ -56,11 +56,11 @@ Function* Builder::CreateFunction(Symbol name, TINT_ASSERT(IR, return_type); auto* ir_func = ir.flow_nodes.Create(name, return_type, stage, wg_size); - ir_func->start_target = CreateBlock(); - ir_func->end_target = CreateFunctionTerminator(); + ir_func->SetStartTarget(CreateBlock()); + ir_func->SetEndTarget(CreateFunctionTerminator()); - // Function is always branching into the start target - ir_func->start_target->inbound_branches.Push(ir_func); + // Function is always branching into the Start().target + ir_func->StartTarget()->AddInboundBranch(ir_func); return ir_func; } @@ -69,53 +69,48 @@ If* Builder::CreateIf(Value* condition) { TINT_ASSERT(IR, condition); auto* ir_if = ir.flow_nodes.Create(condition); - ir_if->true_.target = CreateBlock(); - ir_if->false_.target = CreateBlock(); - ir_if->merge.target = CreateBlock(); + ir_if->True().target = CreateBlock(); + ir_if->False().target = CreateBlock(); + ir_if->Merge().target = CreateBlock(); // An if always branches to both the true and false block. - ir_if->true_.target->inbound_branches.Push(ir_if); - ir_if->false_.target->inbound_branches.Push(ir_if); + ir_if->True().target->AddInboundBranch(ir_if); + ir_if->False().target->AddInboundBranch(ir_if); return ir_if; } Loop* Builder::CreateLoop() { auto* ir_loop = ir.flow_nodes.Create(); - ir_loop->start.target = CreateBlock(); - ir_loop->continuing.target = CreateBlock(); - ir_loop->merge.target = CreateBlock(); + ir_loop->Start().target = CreateBlock(); + ir_loop->Continuing().target = CreateBlock(); + ir_loop->Merge().target = CreateBlock(); // A loop always branches to the start block. - ir_loop->start.target->inbound_branches.Push(ir_loop); + ir_loop->Start().target->AddInboundBranch(ir_loop); return ir_loop; } Switch* Builder::CreateSwitch(Value* condition) { auto* ir_switch = ir.flow_nodes.Create(condition); - ir_switch->merge.target = CreateBlock(); + ir_switch->Merge().target = CreateBlock(); return ir_switch; } Block* Builder::CreateCase(Switch* s, utils::VectorRef selectors) { - s->cases.Push(Switch::Case{selectors, {CreateBlock(), utils::Empty}}); + s->Cases().Push(Switch::Case{selectors, {CreateBlock(), utils::Empty}}); - Block* b = s->cases.Back().start.target->As(); + Block* b = s->Cases().Back().Start().target->As(); // Switch branches into the case block - b->inbound_branches.Push(s); + b->AddInboundBranch(s); return b; } -void Builder::Branch(Block* from, FlowNode* to, utils::VectorRef args) { - TINT_ASSERT(IR, from); - TINT_ASSERT(IR, to); - from->branch.target = to; - from->branch.args = args; - to->inbound_branches.Push(from); -} - -Binary* Builder::CreateBinary(Binary::Kind kind, const type::Type* type, Value* lhs, Value* rhs) { +Binary* Builder::CreateBinary(enum Binary::Kind kind, + const type::Type* type, + Value* lhs, + Value* rhs) { return ir.values.Create(kind, type, lhs, rhs); } @@ -183,7 +178,7 @@ Binary* Builder::Modulo(const type::Type* type, Value* lhs, Value* rhs) { return CreateBinary(Binary::Kind::kModulo, type, lhs, rhs); } -Unary* Builder::CreateUnary(Unary::Kind kind, const type::Type* type, Value* val) { +Unary* Builder::CreateUnary(enum Unary::Kind kind, const type::Type* type, Value* val) { return ir.values.Create(kind, type, val); } diff --git a/src/tint/ir/builder.h b/src/tint/ir/builder.h index 6de608eda9..ebd7a87280 100644 --- a/src/tint/ir/builder.h +++ b/src/tint/ir/builder.h @@ -98,12 +98,6 @@ class Builder { /// @returns the start block for the case flow node Block* CreateCase(Switch* s, utils::VectorRef selectors); - /// Branches the given block to the given flow node. - /// @param from the block to branch from - /// @param to the node to branch too - /// @param args arguments to the branch - void Branch(Block* from, FlowNode* to, utils::VectorRef args = {}); - /// Creates a constant::Value /// @param args the arguments /// @returns the new constant value @@ -161,7 +155,7 @@ class Builder { /// @param lhs the left-hand-side of the operation /// @param rhs the right-hand-side of the operation /// @returns the operation - Binary* CreateBinary(Binary::Kind kind, const type::Type* type, Value* lhs, Value* rhs); + Binary* CreateBinary(enum Binary::Kind kind, const type::Type* type, Value* lhs, Value* rhs); /// Creates an And operation /// @param type the result type of the expression @@ -280,7 +274,7 @@ class Builder { /// @param type the result type of the binary expression /// @param val the value of the operation /// @returns the operation - Unary* CreateUnary(Unary::Kind kind, const type::Type* type, Value* val); + Unary* CreateUnary(enum Unary::Kind kind, const type::Type* type, Value* val); /// Creates a Complement operation /// @param type the result type of the expression diff --git a/src/tint/ir/call.cc b/src/tint/ir/call.cc index dd503eadaa..909079f6e7 100644 --- a/src/tint/ir/call.cc +++ b/src/tint/ir/call.cc @@ -21,8 +21,8 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Call); namespace tint::ir { Call::Call(const type::Type* res_ty, utils::VectorRef arguments) - : result_type(res_ty), args(std::move(arguments)) { - for (auto* arg : args) { + : result_type_(res_ty), args_(std::move(arguments)) { + for (auto* arg : args_) { arg->AddUsage(this); } } diff --git a/src/tint/ir/call.h b/src/tint/ir/call.h index 5393810af5..f4e12f913b 100644 --- a/src/tint/ir/call.h +++ b/src/tint/ir/call.h @@ -31,13 +31,10 @@ class Call : public utils::Castable { Call& operator=(Call&& inst) = delete; /// @returns the type of the value - const type::Type* Type() const override { return result_type; } + const type::Type* Type() const override { return result_type_; } - /// The instruction type - const type::Type* result_type = nullptr; - - /// The constructor arguments - utils::Vector args; + /// @returns the call arguments + utils::VectorRef Args() const { return args_; } protected: /// Constructor @@ -46,6 +43,10 @@ class Call : public utils::Castable { /// @param result_type the result type /// @param args the constructor arguments Call(const type::Type* result_type, utils::VectorRef args); + + private: + const type::Type* result_type_; + utils::Vector args_; }; } // namespace tint::ir diff --git a/src/tint/ir/constant.cc b/src/tint/ir/constant.cc index 8b5260c667..a49d140241 100644 --- a/src/tint/ir/constant.cc +++ b/src/tint/ir/constant.cc @@ -18,7 +18,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Constant); namespace tint::ir { -Constant::Constant(const constant::Value* val) : value(val) {} +Constant::Constant(const constant::Value* val) : value_(val) {} Constant::~Constant() = default; diff --git a/src/tint/ir/constant.h b/src/tint/ir/constant.h index 68e0dc7b92..dcf3e5b2f8 100644 --- a/src/tint/ir/constant.h +++ b/src/tint/ir/constant.h @@ -26,13 +26,21 @@ class Constant : public utils::Castable { /// Constructor /// @param val the value stored in the constant explicit Constant(const constant::Value* val); + Constant(const Constant&) = delete; + Constant(Constant&&) = delete; ~Constant() override; - /// @returns the type of the constant - const type::Type* Type() const override { return value->Type(); } + Constant& operator=(const Constant&) = delete; + Constant& operator=(Constant&&) = delete; - /// The constants value - const constant::Value* const value; + /// @returns the constants value + const constant::Value* Value() const { return value_; } + + /// @returns the type of the constant + const type::Type* Type() const override { return value_->Type(); } + + private: + const constant::Value* const value_; }; } // namespace tint::ir diff --git a/src/tint/ir/constant_test.cc b/src/tint/ir/constant_test.cc index 7005751913..7c3024a322 100644 --- a/src/tint/ir/constant_test.cc +++ b/src/tint/ir/constant_test.cc @@ -30,13 +30,13 @@ TEST_F(IR_ConstantTest, f32) { utils::StringStream str; auto* c = b.Constant(1.2_f); - EXPECT_EQ(1.2_f, c->value->As>()->ValueAs()); + EXPECT_EQ(1.2_f, c->Value()->As>()->ValueAs()); - EXPECT_TRUE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); + EXPECT_TRUE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); } TEST_F(IR_ConstantTest, f16) { @@ -46,13 +46,13 @@ TEST_F(IR_ConstantTest, f16) { utils::StringStream str; auto* c = b.Constant(1.1_h); - EXPECT_EQ(1.1_h, c->value->As>()->ValueAs()); + EXPECT_EQ(1.1_h, c->Value()->As>()->ValueAs()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_TRUE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_TRUE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); } TEST_F(IR_ConstantTest, i32) { @@ -62,13 +62,13 @@ TEST_F(IR_ConstantTest, i32) { utils::StringStream str; auto* c = b.Constant(1_i); - EXPECT_EQ(1_i, c->value->As>()->ValueAs()); + EXPECT_EQ(1_i, c->Value()->As>()->ValueAs()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_TRUE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_TRUE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); } TEST_F(IR_ConstantTest, u32) { @@ -78,13 +78,13 @@ TEST_F(IR_ConstantTest, u32) { utils::StringStream str; auto* c = b.Constant(2_u); - EXPECT_EQ(2_u, c->value->As>()->ValueAs()); + EXPECT_EQ(2_u, c->Value()->As>()->ValueAs()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_TRUE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_TRUE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); } TEST_F(IR_ConstantTest, bool) { @@ -95,19 +95,19 @@ TEST_F(IR_ConstantTest, bool) { utils::StringStream str; auto* c = b.Constant(false); - EXPECT_FALSE(c->value->As>()->ValueAs()); + EXPECT_FALSE(c->Value()->As>()->ValueAs()); } { utils::StringStream str; auto c = b.Constant(true); - EXPECT_TRUE(c->value->As>()->ValueAs()); + EXPECT_TRUE(c->Value()->As>()->ValueAs()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_FALSE(c->value->Is>()); - EXPECT_TRUE(c->value->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_FALSE(c->Value()->Is>()); + EXPECT_TRUE(c->Value()->Is>()); } } diff --git a/src/tint/ir/debug.cc b/src/tint/ir/debug.cc index 655c45507e..85666ff628 100644 --- a/src/tint/ir/debug.cc +++ b/src/tint/ir/debug.cc @@ -60,81 +60,81 @@ std::string Debug::AsDotGraph(const Module* mod) { if (node_to_name.count(b) == 0) { out << name_for(b) << R"( [label="block"])" << std::endl; } - out << name_for(b) << " -> " << name_for(b->branch.target); + out << name_for(b) << " -> " << name_for(b->Branch().target); // Dashed lines to merge blocks - if (merge_nodes.count(b->branch.target) != 0) { + if (merge_nodes.count(b->Branch().target) != 0) { out << " [style=dashed]"; } out << std::endl; - Graph(b->branch.target); + Graph(b->Branch().target); }, [&](const ir::Switch* s) { out << name_for(s) << R"( [label="switch"])" << std::endl; - out << name_for(s->merge.target) << R"( [label="switch merge"])" << std::endl; - merge_nodes.insert(s->merge.target); + out << name_for(s->Merge().target) << R"( [label="switch merge"])" << std::endl; + merge_nodes.insert(s->Merge().target); size_t i = 0; - for (const auto& c : s->cases) { - out << name_for(c.start.target) + for (const auto& c : s->Cases()) { + out << name_for(c.Start().target) << R"( [label="case )" + std::to_string(i++) + R"("])" << std::endl; } out << name_for(s) << " -> {"; - for (const auto& c : s->cases) { - if (&c != &(s->cases[0])) { + for (const auto& c : s->Cases()) { + if (&c != &(s->Cases().Front())) { out << ", "; } - out << name_for(c.start.target); + out << name_for(c.Start().target); } out << "}" << std::endl; - for (const auto& c : s->cases) { - Graph(c.start.target); + for (const auto& c : s->Cases()) { + Graph(c.Start().target); } - Graph(s->merge.target); + Graph(s->Merge().target); }, [&](const ir::If* i) { out << name_for(i) << R"( [label="if"])" << std::endl; - out << name_for(i->true_.target) << R"( [label="true"])" << std::endl; - out << name_for(i->false_.target) << R"( [label="false"])" << std::endl; - out << name_for(i->merge.target) << R"( [label="if merge"])" << std::endl; - merge_nodes.insert(i->merge.target); + out << name_for(i->True().target) << R"( [label="true"])" << std::endl; + out << name_for(i->False().target) << R"( [label="false"])" << std::endl; + out << name_for(i->Merge().target) << R"( [label="if merge"])" << std::endl; + merge_nodes.insert(i->Merge().target); out << name_for(i) << " -> {"; - out << name_for(i->true_.target) << ", " << name_for(i->false_.target); + out << name_for(i->True().target) << ", " << name_for(i->False().target); out << "}" << std::endl; // Subgraph if true/false branches so they draw on the same line out << "subgraph sub_" << name_for(i) << " {" << std::endl; out << R"(rank="same")" << std::endl; - out << name_for(i->true_.target) << std::endl; - out << name_for(i->false_.target) << std::endl; + out << name_for(i->True().target) << std::endl; + out << name_for(i->False().target) << std::endl; out << "}" << std::endl; - Graph(i->true_.target); - Graph(i->false_.target); - Graph(i->merge.target); + Graph(i->True().target); + Graph(i->False().target); + Graph(i->Merge().target); }, [&](const ir::Loop* l) { out << name_for(l) << R"( [label="loop"])" << std::endl; - out << name_for(l->start.target) << R"( [label="start"])" << std::endl; - out << name_for(l->continuing.target) << R"( [label="continuing"])" << std::endl; - out << name_for(l->merge.target) << R"( [label="loop merge"])" << std::endl; - merge_nodes.insert(l->merge.target); + out << name_for(l->Start().target) << R"( [label="start"])" << std::endl; + out << name_for(l->Continuing().target) << R"( [label="continuing"])" << std::endl; + out << name_for(l->Merge().target) << R"( [label="loop merge"])" << std::endl; + merge_nodes.insert(l->Merge().target); // Subgraph the continuing and merge so they get drawn on the same line out << "subgraph sub_" << name_for(l) << " {" << std::endl; out << R"(rank="same")" << std::endl; - out << name_for(l->continuing.target) << std::endl; - out << name_for(l->merge.target) << std::endl; + out << name_for(l->Continuing().target) << std::endl; + out << name_for(l->Merge().target) << std::endl; out << "}" << std::endl; - out << name_for(l) << " -> " << name_for(l->start.target) << std::endl; + out << name_for(l) << " -> " << name_for(l->Start().target) << std::endl; - Graph(l->start.target); - Graph(l->continuing.target); - Graph(l->merge.target); + Graph(l->Start().target); + Graph(l->Continuing().target); + Graph(l->Merge().target); }, [&](const ir::FunctionTerminator*) { // Already done @@ -145,10 +145,10 @@ std::string Debug::AsDotGraph(const Module* mod) { 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 << R"(label=")" << func->name.Name() << R"(")" << std::endl; - out << name_for(func->start_target) << R"( [label="start"])" << std::endl; - out << name_for(func->end_target) << R"( [label="end"])" << std::endl; - Graph(func->start_target); + 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; + Graph(func->StartTarget()); out << "}" << std::endl; } out << "}"; diff --git a/src/tint/ir/disassembler.cc b/src/tint/ir/disassembler.cc index ef5c20b79a..76af74091b 100644 --- a/src/tint/ir/disassembler.cc +++ b/src/tint/ir/disassembler.cc @@ -81,7 +81,7 @@ utils::StringStream& Disassembler::Indent() { } void Disassembler::EmitBlockInstructions(const Block* b) { - for (const auto* inst : b->instructions) { + for (const auto* inst : b->Instructions()) { Indent(); EmitInstruction(inst); out_ << std::endl; @@ -114,31 +114,31 @@ void Disassembler::Walk(const FlowNode* node) { [&](const ir::Function* f) { TINT_SCOPED_ASSIGNMENT(in_function_, true); - Indent() << "%fn" << IdOf(f) << " = func " << f->name.Name() << "("; - for (auto* p : f->params) { - if (p != f->params.Front()) { + Indent() << "%fn" << IdOf(f) << " = func " << f->Name().Name() << "("; + for (auto* p : f->Params()) { + if (p != f->Params().Front()) { out_ << ", "; } - out_ << "%" << IdOf(p) << ":" << p->type->FriendlyName(); + out_ << "%" << IdOf(p) << ":" << p->Type()->FriendlyName(); } - out_ << "):" << f->return_type->FriendlyName(); + out_ << "):" << f->ReturnType()->FriendlyName(); - if (f->pipeline_stage != Function::PipelineStage::kUndefined) { - out_ << " [@" << f->pipeline_stage; + if (f->Stage() != Function::PipelineStage::kUndefined) { + out_ << " [@" << f->Stage(); - if (f->workgroup_size) { - auto arr = f->workgroup_size.value(); + if (f->WorkgroupSize()) { + auto arr = f->WorkgroupSize().value(); out_ << " @workgroup_size(" << arr[0] << ", " << arr[1] << ", " << arr[2] << ")"; } - if (!f->return_attributes.IsEmpty()) { + if (!f->ReturnAttributes().IsEmpty()) { out_ << " ra:"; - for (auto attr : f->return_attributes) { + for (auto attr : f->ReturnAttributes()) { out_ << " @" << attr; if (attr == Function::ReturnAttribute::kLocation) { - out_ << "(" << f->return_location.value() << ")"; + out_ << "(" << f->ReturnLocation().value() << ")"; } } } @@ -149,23 +149,23 @@ void Disassembler::Walk(const FlowNode* node) { { ScopedIndent func_indent(indent_size_); - ScopedStopNode scope(stop_nodes_, f->end_target); - Walk(f->start_target); + ScopedStopNode scope(stop_nodes_, f->EndTarget()); + Walk(f->StartTarget()); } out_ << "} "; - Walk(f->end_target); + Walk(f->EndTarget()); }, [&](const ir::Block* b) { // If this block is dead, nothing to do - if (b->IsDead()) { + if (!b->HasBranchTarget()) { return; } Indent() << "%fn" << IdOf(b) << " = block"; - if (!b->params.IsEmpty()) { + if (!b->Params().IsEmpty()) { out_ << " ("; - for (auto* p : b->params) { - if (p != b->params.Front()) { + for (const auto* p : b->Params()) { + if (p != b->Params().Front()) { out_ << ", "; } EmitValue(p); @@ -181,20 +181,20 @@ void Disassembler::Walk(const FlowNode* node) { Indent() << "}"; std::string suffix = ""; - if (b->branch.target->Is()) { + if (b->Branch().target->Is()) { out_ << " -> %func_end"; suffix = "return"; - } else if (b->branch.target->Is()) { + } else if (b->Branch().target->Is()) { // Nothing to do } else { out_ << " -> " - << "%fn" << IdOf(b->branch.target); + << "%fn" << IdOf(b->Branch().target); suffix = "branch"; } - if (!b->branch.args.IsEmpty()) { + if (!b->Branch().args.IsEmpty()) { out_ << " "; - for (const auto* v : b->branch.args) { - if (v != b->branch.args.Front()) { + for (const auto* v : b->Branch().args) { + if (v != b->Branch().args.Front()) { out_ << ", "; } EmitValue(v); @@ -205,18 +205,18 @@ void Disassembler::Walk(const FlowNode* node) { } out_ << std::endl; - if (!b->branch.target->Is()) { + if (!b->Branch().target->Is()) { out_ << std::endl; } - Walk(b->branch.target); + Walk(b->Branch().target); }, [&](const ir::Switch* s) { Indent() << "%fn" << IdOf(s) << " = switch "; - EmitValue(s->condition); + EmitValue(s->Condition()); out_ << " ["; - for (const auto& c : s->cases) { - if (&c != &s->cases.Front()) { + for (const auto& c : s->Cases()) { + if (&c != &s->Cases().Front()) { out_ << ", "; } out_ << "c: ("; @@ -231,17 +231,17 @@ void Disassembler::Walk(const FlowNode* node) { EmitValue(selector.val); } } - out_ << ", %fn" << IdOf(c.start.target) << ")"; + out_ << ", %fn" << IdOf(c.Start().target) << ")"; } - if (s->merge.target->IsConnected()) { - out_ << ", m: %fn" << IdOf(s->merge.target); + if (s->Merge().target->IsConnected()) { + out_ << ", m: %fn" << IdOf(s->Merge().target); } out_ << "]" << std::endl; { ScopedIndent switch_indent(indent_size_); - ScopedStopNode scope(stop_nodes_, s->merge.target); - for (const auto& c : s->cases) { + ScopedStopNode scope(stop_nodes_, s->Merge().target); + for (const auto& c : s->Cases()) { Indent() << "# case "; for (const auto& selector : c.selectors) { if (&selector != &c.selectors.Front()) { @@ -255,86 +255,86 @@ void Disassembler::Walk(const FlowNode* node) { } } out_ << std::endl; - Walk(c.start.target); + Walk(c.Start().target); } } - if (s->merge.target->IsConnected()) { + if (s->Merge().target->IsConnected()) { Indent() << "# switch merge" << std::endl; - Walk(s->merge.target); + Walk(s->Merge().target); } }, [&](const ir::If* i) { Indent() << "%fn" << IdOf(i) << " = if "; - EmitValue(i->condition); + EmitValue(i->Condition()); - bool has_true = !i->true_.target->IsDead(); - bool has_false = !i->false_.target->IsDead(); + bool has_true = i->True().target->HasBranchTarget(); + bool has_false = i->False().target->HasBranchTarget(); out_ << " ["; if (has_true) { - out_ << "t: %fn" << IdOf(i->true_.target); + out_ << "t: %fn" << IdOf(i->True().target); } if (has_false) { if (has_true) { out_ << ", "; } - out_ << "f: %fn" << IdOf(i->false_.target); + out_ << "f: %fn" << IdOf(i->False().target); } - if (i->merge.target->IsConnected()) { - out_ << ", m: %fn" << IdOf(i->merge.target); + if (i->Merge().target->IsConnected()) { + out_ << ", m: %fn" << IdOf(i->Merge().target); } out_ << "]" << std::endl; { ScopedIndent if_indent(indent_size_); - ScopedStopNode scope(stop_nodes_, i->merge.target); + ScopedStopNode scope(stop_nodes_, i->Merge().target); if (has_true) { Indent() << "# true branch" << std::endl; - Walk(i->true_.target); + Walk(i->True().target); } if (has_false) { Indent() << "# false branch" << std::endl; - Walk(i->false_.target); + Walk(i->False().target); } } - if (i->merge.target->IsConnected()) { + if (i->Merge().target->IsConnected()) { Indent() << "# if merge" << std::endl; - Walk(i->merge.target); + Walk(i->Merge().target); } }, [&](const ir::Loop* l) { - Indent() << "%fn" << IdOf(l) << " = loop [s: %fn" << IdOf(l->start.target); + Indent() << "%fn" << IdOf(l) << " = loop [s: %fn" << IdOf(l->Start().target); - if (l->continuing.target->IsConnected()) { - out_ << ", c: %fn" << IdOf(l->continuing.target); + if (l->Continuing().target->IsConnected()) { + out_ << ", c: %fn" << IdOf(l->Continuing().target); } - if (l->merge.target->IsConnected()) { - out_ << ", m: %fn" << IdOf(l->merge.target); + if (l->Merge().target->IsConnected()) { + out_ << ", m: %fn" << IdOf(l->Merge().target); } out_ << "]" << std::endl; { - ScopedStopNode loop_scope(stop_nodes_, l->merge.target); + ScopedStopNode loop_scope(stop_nodes_, l->Merge().target); ScopedIndent loop_indent(indent_size_); { - ScopedStopNode inner_scope(stop_nodes_, l->continuing.target); + ScopedStopNode inner_scope(stop_nodes_, l->Continuing().target); Indent() << "# loop start" << std::endl; - Walk(l->start.target); + Walk(l->Start().target); } - if (l->continuing.target->IsConnected()) { + if (l->Continuing().target->IsConnected()) { Indent() << "# loop continuing" << std::endl; - Walk(l->continuing.target); + Walk(l->Continuing().target); } } - if (l->merge.target->IsConnected()) { + if (l->Merge().target->IsConnected()) { Indent() << "# loop merge" << std::endl; - Walk(l->merge.target); + Walk(l->Merge().target); } }, [&](const ir::FunctionTerminator*) { @@ -407,7 +407,7 @@ void Disassembler::EmitValue(const Value* val) { } }); }; - emit(constant->value); + emit(constant->Value()); }, [&](const ir::Instruction* i) { out_ << "%" << IdOf(i); }, [&](const ir::BlockParam* p) { @@ -445,18 +445,18 @@ void Disassembler::EmitInstruction(const Instruction* inst) { [&](const ir::Load* l) { EmitValueWithType(l); out_ << " = load "; - EmitValue(l->from); + EmitValue(l->From()); }, [&](const ir::Store* s) { out_ << "store "; - EmitValue(s->to); + EmitValue(s->To()); out_ << ", "; - EmitValue(s->from); + EmitValue(s->From()); }, [&](const ir::UserCall* uc) { EmitValueWithType(uc); - out_ << " = call " << uc->name.Name(); - if (uc->args.Length() > 0) { + out_ << " = call " << uc->Name().Name(); + if (!uc->Args().IsEmpty()) { out_ << ", "; } EmitArgs(uc); @@ -464,16 +464,16 @@ void Disassembler::EmitInstruction(const Instruction* inst) { [&](const ir::Var* v) { EmitValueWithType(v); out_ << " = var"; - if (v->initializer) { + if (v->Initializer()) { out_ << ", "; - EmitValue(v->initializer); + EmitValue(v->Initializer()); } }); } void Disassembler::EmitArgs(const Call* call) { bool first = true; - for (const auto* arg : call->args) { + for (const auto* arg : call->Args()) { if (!first) { out_ << ", "; } @@ -485,7 +485,7 @@ void Disassembler::EmitArgs(const Call* call) { void Disassembler::EmitBinary(const Binary* b) { EmitValueWithType(b); out_ << " = "; - switch (b->kind) { + switch (b->Kind()) { case Binary::Kind::kAdd: out_ << "add"; break; @@ -544,7 +544,7 @@ void Disassembler::EmitBinary(const Binary* b) { void Disassembler::EmitUnary(const Unary* u) { EmitValueWithType(u); out_ << " = "; - switch (u->kind) { + switch (u->Kind()) { case Unary::Kind::kComplement: out_ << "complement"; break; diff --git a/src/tint/ir/flow_node.h b/src/tint/ir/flow_node.h index 905f077605..289873b1c5 100644 --- a/src/tint/ir/flow_node.h +++ b/src/tint/ir/flow_node.h @@ -25,22 +25,30 @@ class FlowNode : public utils::Castable { public: ~FlowNode() override; - /// The list of flow nodes which branch into this node. This list maybe empty for several - /// reasons: - /// - Node is a start node - /// - Node is a merge target outside control flow (if that returns in both branches) - /// - Node is a continue target outside control flow (loop that returns) - utils::Vector inbound_branches; - /// @returns true if this node has inbound branches and branches out - bool IsConnected() const { return !IsDead() && !inbound_branches.IsEmpty(); } + bool IsConnected() const { return HasBranchTarget() && !inbound_branches_.IsEmpty(); } - /// @returns true if the node does not branch out - virtual bool IsDead() const { return false; } + /// @returns true if the node has a branch target + virtual bool HasBranchTarget() const { return false; } + + /// @returns the inbound branch list for the flow node + utils::VectorRef InboundBranches() const { return inbound_branches_; } + + /// Adds the given node to the inbound branches + /// @param node the node to add + void AddInboundBranch(FlowNode* node) { inbound_branches_.Push(node); } protected: /// Constructor FlowNode(); + + private: + /// The list of flow nodes which branch into this node. This list maybe empty for several + /// reasons: + /// - Node is a start node + /// - Node is a merge target outside control flow (e.g. an if that returns in both branches) + /// - Node is a continue target outside control flow (e.g. a loop that returns) + utils::Vector inbound_branches_; }; } // namespace tint::ir diff --git a/src/tint/ir/from_program.cc b/src/tint/ir/from_program.cc index 11bd4028fb..595db584d7 100644 --- a/src/tint/ir/from_program.cc +++ b/src/tint/ir/from_program.cc @@ -98,17 +98,13 @@ namespace { using ResultType = utils::Result; -bool IsBranched(const Block* b) { - return b->branch.target != nullptr; -} - bool IsConnected(const FlowNode* b) { // Function is always connected as it's the start. if (b->Is()) { return true; } - for (auto* parent : b->inbound_branches) { + for (auto* parent : b->InboundBranches()) { if (IsConnected(parent)) { return true; } @@ -184,14 +180,14 @@ class Impl { void BranchTo(FlowNode* node, utils::VectorRef args = {}) { TINT_ASSERT(IR, current_flow_block_); - TINT_ASSERT(IR, !IsBranched(current_flow_block_)); + TINT_ASSERT(IR, !current_flow_block_->HasBranchTarget()); - builder_.Branch(current_flow_block_, node, args); + current_flow_block_->BranchTo(node, args); current_flow_block_ = nullptr; } void BranchToIfNeeded(FlowNode* node) { - if (!current_flow_block_ || IsBranched(current_flow_block_)) { + if (!current_flow_block_ || current_flow_block_->HasBranchTarget()) { return; } BranchTo(node); @@ -271,20 +267,17 @@ class Impl { if (ast_func->IsEntryPoint()) { switch (ast_func->PipelineStage()) { case ast::PipelineStage::kVertex: - ir_func->pipeline_stage = Function::PipelineStage::kVertex; + ir_func->SetStage(Function::PipelineStage::kVertex); break; case ast::PipelineStage::kFragment: - ir_func->pipeline_stage = Function::PipelineStage::kFragment; + ir_func->SetStage(Function::PipelineStage::kFragment); break; case ast::PipelineStage::kCompute: { - ir_func->pipeline_stage = Function::PipelineStage::kCompute; + ir_func->SetStage(Function::PipelineStage::kCompute); auto wg_size = sem->WorkgroupSize(); - ir_func->workgroup_size = { - wg_size[0].value(), - wg_size[1].value_or(1), - wg_size[2].value_or(1), - }; + ir_func->SetWorkgroupSize(wg_size[0].value(), wg_size[1].value_or(1), + wg_size[2].value_or(1)); break; } default: { @@ -293,14 +286,15 @@ class Impl { } } + utils::Vector return_attributes; for (auto* attr : ast_func->return_type_attributes) { tint::Switch( attr, // [&](const ast::LocationAttribute*) { - ir_func->return_attributes.Push(Function::ReturnAttribute::kLocation); + return_attributes.Push(Function::ReturnAttribute::kLocation); }, [&](const ast::InvariantAttribute*) { - ir_func->return_attributes.Push(Function::ReturnAttribute::kInvariant); + return_attributes.Push(Function::ReturnAttribute::kInvariant); }, [&](const ast::BuiltinAttribute* b) { if (auto* ident_sem = @@ -309,16 +303,13 @@ class Impl { ->As>()) { switch (ident_sem->Value()) { case builtin::BuiltinValue::kPosition: - ir_func->return_attributes.Push( - Function::ReturnAttribute::kPosition); + return_attributes.Push(Function::ReturnAttribute::kPosition); break; case builtin::BuiltinValue::kFragDepth: - ir_func->return_attributes.Push( - Function::ReturnAttribute::kFragDepth); + return_attributes.Push(Function::ReturnAttribute::kFragDepth); break; case builtin::BuiltinValue::kSampleMask: - ir_func->return_attributes.Push( - Function::ReturnAttribute::kSampleMask); + return_attributes.Push(Function::ReturnAttribute::kSampleMask); break; default: TINT_ICE(IR, diagnostics_) @@ -332,8 +323,9 @@ class Impl { } }); } + ir_func->SetReturnAttributes(return_attributes); } - ir_func->return_location = sem->ReturnLocation(); + ir_func->SetReturnLocation(sem->ReturnLocation()); scopes_.Push(); TINT_DEFER(scopes_.Pop()); @@ -348,17 +340,17 @@ class Impl { builder_.ir.SetName(param, p->name->symbol.NameView()); params.Push(param); } - ir_func->params = std::move(params); + ir_func->SetParams(params); { FlowStackScope scope(this, ir_func); - current_flow_block_ = ir_func->start_target; + current_flow_block_ = ir_func->StartTarget(); EmitBlock(ast_func->body); // If the branch target has already been set then a `return` was called. Only set in the // case where `return` wasn't called. - BranchToIfNeeded(current_function_->end_target); + BranchToIfNeeded(current_function_->EndTarget()); } TINT_ASSERT(IR, flow_stack_.IsEmpty()); @@ -372,7 +364,7 @@ class Impl { // If the current flow block has a branch target then the rest of the statements in this // block are dead code. Skip them. - if (!current_flow_block_ || IsBranched(current_flow_block_)) { + if (!current_flow_block_ || current_flow_block_->HasBranchTarget()) { break; } } @@ -427,7 +419,7 @@ class Impl { return; } auto store = builder_.Store(lhs.Get(), rhs.Get()); - current_flow_block_->instructions.Push(store); + current_flow_block_->Instructions().Push(store); } void EmitIncrementDecrement(const ast::IncrementDecrementStatement* stmt) { @@ -438,7 +430,7 @@ class Impl { // Load from the LHS. auto* lhs_value = builder_.Load(lhs.Get()); - current_flow_block_->instructions.Push(lhs_value); + current_flow_block_->Instructions().Push(lhs_value); auto* ty = lhs_value->Type(); @@ -451,10 +443,10 @@ class Impl { } else { inst = builder_.Subtract(ty, lhs_value, rhs); } - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); auto store = builder_.Store(lhs.Get(), inst); - current_flow_block_->instructions.Push(store); + current_flow_block_->Instructions().Push(store); } void EmitCompoundAssignment(const ast::CompoundAssignmentStatement* stmt) { @@ -470,7 +462,7 @@ class Impl { // Load from the LHS. auto* lhs_value = builder_.Load(lhs.Get()); - current_flow_block_->instructions.Push(lhs_value); + current_flow_block_->Instructions().Push(lhs_value); auto* ty = lhs_value->Type(); @@ -520,10 +512,10 @@ class Impl { TINT_ICE(IR, diagnostics_) << "missing binary operand type"; return; } - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); auto store = builder_.Store(lhs.Get(), inst); - current_flow_block_->instructions.Push(store); + current_flow_block_->Instructions().Push(store); } void EmitBlock(const ast::BlockStatement* block) { @@ -551,27 +543,27 @@ class Impl { { FlowStackScope scope(this, if_node); - current_flow_block_ = if_node->true_.target->As(); + current_flow_block_ = if_node->True().target->As(); EmitBlock(stmt->body); - // If the true branch did not execute control flow, then go to the merge target - BranchToIfNeeded(if_node->merge.target); + // If the true branch did not execute control flow, then go to the Merge().target + BranchToIfNeeded(if_node->Merge().target); - current_flow_block_ = if_node->false_.target->As(); + current_flow_block_ = if_node->False().target->As(); if (stmt->else_statement) { EmitStatement(stmt->else_statement); } - // If the false branch did not execute control flow, then go to the merge target - BranchToIfNeeded(if_node->merge.target); + // If the false branch did not execute control flow, then go to the Merge().target + BranchToIfNeeded(if_node->Merge().target); } current_flow_block_ = nullptr; // If both branches went somewhere, then they both returned, continued or broke. So, there // is no need for the if merge-block and there is nothing to branch to the merge block // anyway. - if (IsConnected(if_node->merge.target)) { - current_flow_block_ = if_node->merge.target->As(); + if (IsConnected(if_node->Merge().target)) { + current_flow_block_ = if_node->Merge().target->As(); } } @@ -585,7 +577,7 @@ class Impl { { FlowStackScope scope(this, loop_node); - current_flow_block_ = loop_node->start.target->As(); + current_flow_block_ = loop_node->Start().target->As(); // The loop doesn't use EmitBlock because it needs the scope stack to not get popped // until after the continuing block. @@ -594,21 +586,22 @@ class Impl { EmitStatements(stmt->body->statements); // The current block didn't `break`, `return` or `continue`, go to the continuing block. - BranchToIfNeeded(loop_node->continuing.target); + BranchToIfNeeded(loop_node->Continuing().target); - current_flow_block_ = loop_node->continuing.target->As(); + current_flow_block_ = loop_node->Continuing().target->As(); if (stmt->continuing) { EmitBlock(stmt->continuing); } // Branch back to the start node if the continue target didn't branch out already - BranchToIfNeeded(loop_node->start.target); + BranchToIfNeeded(loop_node->Start().target); } // The loop merge can get disconnected if the loop returns directly, or the continuing - // target branches, eventually, to the merge, but nothing branched to the continuing target. - current_flow_block_ = loop_node->merge.target->As(); - if (!IsConnected(loop_node->merge.target)) { + // target branches, eventually, to the merge, but nothing branched to the + // Continuing().target. + current_flow_block_ = loop_node->Merge().target->As(); + if (!IsConnected(loop_node->Merge().target)) { current_flow_block_ = nullptr; } } @@ -616,9 +609,8 @@ class Impl { void EmitWhile(const ast::WhileStatement* stmt) { auto* loop_node = builder_.CreateLoop(); // Continue is always empty, just go back to the start - TINT_ASSERT(IR, loop_node->continuing.target->Is()); - builder_.Branch(loop_node->continuing.target->As(), loop_node->start.target, - utils::Empty); + TINT_ASSERT(IR, loop_node->Continuing().target->Is()); + loop_node->Continuing().target->As()->BranchTo(loop_node->Start().target); BranchTo(loop_node); @@ -627,9 +619,9 @@ class Impl { { FlowStackScope scope(this, loop_node); - current_flow_block_ = loop_node->start.target->As(); + current_flow_block_ = loop_node->Start().target->As(); - // Emit the while condition into the start target of the loop + // Emit the while condition into the Start().target of the loop auto reg = EmitExpression(stmt->condition); if (!reg) { return; @@ -637,31 +629,24 @@ class Impl { // Create an `if (cond) {} else {break;}` control flow auto* if_node = builder_.CreateIf(reg.Get()); - TINT_ASSERT(IR, if_node->true_.target->Is()); - builder_.Branch(if_node->true_.target->As(), if_node->merge.target, - utils::Empty); - - TINT_ASSERT(IR, if_node->false_.target->Is()); - builder_.Branch(if_node->false_.target->As(), loop_node->merge.target, - utils::Empty); + if_node->True().target->As()->BranchTo(if_node->Merge().target); + if_node->False().target->As()->BranchTo(loop_node->Merge().target); BranchTo(if_node); - current_flow_block_ = if_node->merge.target->As(); + current_flow_block_ = if_node->Merge().target->As(); EmitBlock(stmt->body); - BranchToIfNeeded(loop_node->continuing.target); + BranchToIfNeeded(loop_node->Continuing().target); } - // The while loop always has a path to the merge target as the break statement comes before - // anything inside the loop. - current_flow_block_ = loop_node->merge.target->As(); + // The while loop always has a path to the Merge().target as the break statement comes + // before anything inside the loop. + current_flow_block_ = loop_node->Merge().target->As(); } void EmitForLoop(const ast::ForLoopStatement* stmt) { auto* loop_node = builder_.CreateLoop(); - TINT_ASSERT(IR, loop_node->continuing.target->Is()); - builder_.Branch(loop_node->continuing.target->As(), loop_node->start.target, - utils::Empty); + loop_node->Continuing().target->As()->BranchTo(loop_node->Start().target); // Make sure the initializer ends up in a contained scope scopes_.Push(); @@ -679,7 +664,7 @@ class Impl { { FlowStackScope scope(this, loop_node); - current_flow_block_ = loop_node->start.target->As(); + current_flow_block_ = loop_node->Start().target->As(); if (stmt->condition) { // Emit the condition into the target target of the loop @@ -690,30 +675,25 @@ class Impl { // Create an `if (cond) {} else {break;}` control flow auto* if_node = builder_.CreateIf(reg.Get()); - TINT_ASSERT(IR, if_node->true_.target->Is()); - builder_.Branch(if_node->true_.target->As(), if_node->merge.target, - utils::Empty); - - TINT_ASSERT(IR, if_node->false_.target->Is()); - builder_.Branch(if_node->false_.target->As(), loop_node->merge.target, - utils::Empty); + if_node->True().target->As()->BranchTo(if_node->Merge().target); + if_node->False().target->As()->BranchTo(loop_node->Merge().target); BranchTo(if_node); - current_flow_block_ = if_node->merge.target->As(); + current_flow_block_ = if_node->Merge().target->As(); } EmitBlock(stmt->body); - BranchToIfNeeded(loop_node->continuing.target); + BranchToIfNeeded(loop_node->Continuing().target); if (stmt->continuing) { - current_flow_block_ = loop_node->continuing.target->As(); + current_flow_block_ = loop_node->Continuing().target->As(); EmitStatement(stmt->continuing); } } - // The while loop always has a path to the merge target as the break statement comes before - // anything inside the loop. - current_flow_block_ = loop_node->merge.target->As(); + // The while loop always has a path to the Merge().target as the break statement comes + // before anything inside the loop. + current_flow_block_ = loop_node->Merge().target->As(); } void EmitSwitch(const ast::SwitchStatement* stmt) { @@ -745,13 +725,13 @@ class Impl { current_flow_block_ = builder_.CreateCase(switch_node, selectors); EmitBlock(c->Body()->Declaration()); - BranchToIfNeeded(switch_node->merge.target); + BranchToIfNeeded(switch_node->Merge().target); } } current_flow_block_ = nullptr; - if (IsConnected(switch_node->merge.target)) { - current_flow_block_ = switch_node->merge.target->As(); + if (IsConnected(switch_node->Merge().target)) { + current_flow_block_ = switch_node->Merge().target->As(); } } @@ -765,7 +745,7 @@ class Impl { ret_value.Push(ret.Get()); } - BranchTo(current_function_->end_target, std::move(ret_value)); + BranchTo(current_function_->EndTarget(), std::move(ret_value)); } void EmitBreak(const ast::BreakStatement*) { @@ -773,9 +753,9 @@ class Impl { TINT_ASSERT(IR, current_control); if (auto* c = current_control->As()) { - BranchTo(c->merge.target); + BranchTo(c->Merge().target); } else if (auto* s = current_control->As()) { - BranchTo(s->merge.target); + BranchTo(s->Merge().target); } else { TINT_UNREACHABLE(IR, diagnostics_); } @@ -786,7 +766,7 @@ class Impl { TINT_ASSERT(IR, current_control); if (auto* c = current_control->As()) { - BranchTo(c->continuing.target); + BranchTo(c->Continuing().target); } else { TINT_UNREACHABLE(IR, diagnostics_); } @@ -798,7 +778,7 @@ class Impl { // figuring out the multi-level exit that is triggered. void EmitDiscard(const ast::DiscardStatement*) { auto* inst = builder_.Discard(); - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); } void EmitBreakIf(const ast::BreakIfStatement* stmt) { @@ -819,17 +799,17 @@ class Impl { auto* loop = current_control->As(); - current_flow_block_ = if_node->true_.target->As(); - BranchTo(loop->merge.target); + current_flow_block_ = if_node->True().target->As(); + BranchTo(loop->Merge().target); - current_flow_block_ = if_node->false_.target->As(); - BranchTo(if_node->merge.target); + current_flow_block_ = if_node->False().target->As(); + BranchTo(if_node->Merge().target); - current_flow_block_ = if_node->merge.target->As(); + current_flow_block_ = if_node->Merge().target->As(); // The `break-if` has to be the last item in the continuing block. The false branch of the // `break-if` will always take us back to the start of the loop. - BranchTo(loop->start.target); + BranchTo(loop->Start().target); } utils::Result EmitExpression(const ast::Expression* expr) { @@ -876,7 +856,7 @@ class Impl { // If this expression maps to sem::Load, insert a load instruction to get the result. if (result && sem->Is()) { auto* load = builder_.Load(result.Get()); - current_flow_block_->instructions.Push(load); + current_flow_block_->Instructions().Push(load); return load; } @@ -895,14 +875,14 @@ class Impl { ref->Access()); auto* val = builder_.Declare(ty); - current_flow_block_->instructions.Push(val); + current_flow_block_->Instructions().Push(val); if (v->initializer) { auto init = EmitExpression(v->initializer); if (!init) { return; } - val->initializer = init.Get(); + val->SetInitializer(init.Get()); } // Store the declaration so we can get the instruction to store too scopes_.Set(v->name->symbol, val); @@ -969,7 +949,7 @@ class Impl { break; } - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); return inst; } @@ -996,7 +976,7 @@ class Impl { BranchTo(if_node); auto* result = builder_.BlockParam(builder_.ir.types.Get()); - if_node->merge.target->As()->params.Push(result); + if_node->Merge().target->As()->SetParams(utils::Vector{result}); utils::Result rhs; { @@ -1010,17 +990,17 @@ class Impl { if (expr->op == ast::BinaryOp::kLogicalAnd) { // If the lhs is false, then that is the result we want to pass to the merge block // as our argument - current_flow_block_ = if_node->false_.target->As(); - BranchTo(if_node->merge.target, std::move(alt_args)); + current_flow_block_ = if_node->False().target->As(); + BranchTo(if_node->Merge().target, std::move(alt_args)); - current_flow_block_ = if_node->true_.target->As(); + current_flow_block_ = if_node->True().target->As(); } else { // If the lhs is true, then that is the result we want to pass to the merge block // as our argument - current_flow_block_ = if_node->true_.target->As(); - BranchTo(if_node->merge.target, std::move(alt_args)); + current_flow_block_ = if_node->True().target->As(); + BranchTo(if_node->Merge().target, std::move(alt_args)); - current_flow_block_ = if_node->false_.target->As(); + current_flow_block_ = if_node->False().target->As(); } rhs = EmitExpression(expr->rhs); @@ -1030,9 +1010,9 @@ class Impl { utils::Vector args; args.Push(rhs.Get()); - BranchTo(if_node->merge.target, std::move(args)); + BranchTo(if_node->Merge().target, std::move(args)); } - current_flow_block_ = if_node->merge.target->As(); + current_flow_block_ = if_node->Merge().target->As(); return result; } @@ -1114,7 +1094,7 @@ class Impl { return utils::Failure; } - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); return inst; } @@ -1128,7 +1108,7 @@ class Impl { auto* ty = sem->Type()->Clone(clone_ctx_.type_ctx); auto* inst = builder_.Bitcast(ty, val.Get()); - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); return inst; } @@ -1191,7 +1171,7 @@ class Impl { if (inst == nullptr) { return utils::Failure; } - current_flow_block_->instructions.Push(inst); + current_flow_block_->Instructions().Push(inst); return inst; } diff --git a/src/tint/ir/from_program_literal_test.cc b/src/tint/ir/from_program_literal_test.cc index 39b6052207..f746bc00be 100644 --- a/src/tint/ir/from_program_literal_test.cc +++ b/src/tint/ir/from_program_literal_test.cc @@ -25,17 +25,19 @@ namespace tint::ir { namespace { -Value* GlobalVarInitializer(const Module& m) { - if (m.root_block->instructions.Length() == 0u) { +const Value* GlobalVarInitializer(const Module& m) { + const auto instr = m.root_block->Instructions(); + + if (instr.Length() == 0u) { ADD_FAILURE() << "m.root_block has no instruction"; return nullptr; } - auto* var = m.root_block->instructions[0]->As(); + auto* var = instr[0]->As(); if (!var) { ADD_FAILURE() << "m.root_block.instructions[0] was not a var"; return nullptr; } - return var->initializer; + return var->Initializer(); } using namespace tint::number_suffixes; // NOLINT @@ -51,7 +53,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_Bool_True) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_TRUE(val->As>()->ValueAs()); } @@ -65,7 +67,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_Bool_False) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_FALSE(val->As>()->ValueAs()); } @@ -79,18 +81,19 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_Bool_Deduped) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - auto* var_a = m.Get().root_block->instructions[0]->As(); + auto instr = m.Get().root_block->Instructions(); + auto* var_a = instr[0]->As(); ASSERT_NE(var_a, nullptr); - auto* var_b = m.Get().root_block->instructions[1]->As(); + auto* var_b = instr[1]->As(); ASSERT_NE(var_b, nullptr); - auto* var_c = m.Get().root_block->instructions[2]->As(); + auto* var_c = instr[2]->As(); ASSERT_NE(var_c, nullptr); - auto* var_d = m.Get().root_block->instructions[3]->As(); + auto* var_d = instr[3]->As(); ASSERT_NE(var_d, nullptr); - ASSERT_EQ(var_a->initializer, var_c->initializer); - ASSERT_EQ(var_b->initializer, var_d->initializer); - ASSERT_NE(var_a->initializer, var_b->initializer); + ASSERT_EQ(var_a->Initializer(), var_c->Initializer()); + ASSERT_EQ(var_b->Initializer(), var_d->Initializer()); + ASSERT_NE(var_a->Initializer(), var_b->Initializer()); } TEST_F(IR_BuilderImplTest, EmitLiteral_F32) { @@ -102,7 +105,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_F32) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_EQ(1.2_f, val->As>()->ValueAs()); } @@ -115,15 +118,16 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_F32_Deduped) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - auto* var_a = m.Get().root_block->instructions[0]->As(); + auto instr = m.Get().root_block->Instructions(); + auto* var_a = instr[0]->As(); ASSERT_NE(var_a, nullptr); - auto* var_b = m.Get().root_block->instructions[1]->As(); + auto* var_b = instr[1]->As(); ASSERT_NE(var_b, nullptr); - auto* var_c = m.Get().root_block->instructions[2]->As(); + auto* var_c = instr[2]->As(); ASSERT_NE(var_c, nullptr); - ASSERT_EQ(var_a->initializer, var_c->initializer); - ASSERT_NE(var_a->initializer, var_b->initializer); + ASSERT_EQ(var_a->Initializer(), var_c->Initializer()); + ASSERT_NE(var_a->Initializer(), var_b->Initializer()); } TEST_F(IR_BuilderImplTest, EmitLiteral_F16) { @@ -136,7 +140,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_F16) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_EQ(1.2_h, val->As>()->ValueAs()); } @@ -150,15 +154,16 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_F16_Deduped) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - auto* var_a = m.Get().root_block->instructions[0]->As(); + auto instr = m.Get().root_block->Instructions(); + auto* var_a = instr[0]->As(); ASSERT_NE(var_a, nullptr); - auto* var_b = m.Get().root_block->instructions[1]->As(); + auto* var_b = instr[1]->As(); ASSERT_NE(var_b, nullptr); - auto* var_c = m.Get().root_block->instructions[2]->As(); + auto* var_c = instr[2]->As(); ASSERT_NE(var_c, nullptr); - ASSERT_EQ(var_a->initializer, var_c->initializer); - ASSERT_NE(var_a->initializer, var_b->initializer); + ASSERT_EQ(var_a->Initializer(), var_c->Initializer()); + ASSERT_NE(var_a->Initializer(), var_b->Initializer()); } TEST_F(IR_BuilderImplTest, EmitLiteral_I32) { @@ -170,7 +175,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_I32) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_EQ(-2_i, val->As>()->ValueAs()); } @@ -183,15 +188,16 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_I32_Deduped) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - auto* var_a = m.Get().root_block->instructions[0]->As(); + auto instr = m.Get().root_block->Instructions(); + auto* var_a = instr[0]->As(); ASSERT_NE(var_a, nullptr); - auto* var_b = m.Get().root_block->instructions[1]->As(); + auto* var_b = instr[1]->As(); ASSERT_NE(var_b, nullptr); - auto* var_c = m.Get().root_block->instructions[2]->As(); + auto* var_c = instr[2]->As(); ASSERT_NE(var_c, nullptr); - ASSERT_EQ(var_a->initializer, var_c->initializer); - ASSERT_NE(var_a->initializer, var_b->initializer); + ASSERT_EQ(var_a->Initializer(), var_c->Initializer()); + ASSERT_NE(var_a->Initializer(), var_b->Initializer()); } TEST_F(IR_BuilderImplTest, EmitLiteral_U32) { @@ -203,7 +209,7 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_U32) { auto* init = GlobalVarInitializer(m.Get()); ASSERT_TRUE(Is(init)); - auto* val = init->As()->value; + auto* val = init->As()->Value(); EXPECT_TRUE(val->Is>()); EXPECT_EQ(2_u, val->As>()->ValueAs()); } @@ -216,15 +222,16 @@ TEST_F(IR_BuilderImplTest, EmitLiteral_U32_Deduped) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - auto* var_a = m.Get().root_block->instructions[0]->As(); + auto instr = m.Get().root_block->Instructions(); + auto* var_a = instr[0]->As(); ASSERT_NE(var_a, nullptr); - auto* var_b = m.Get().root_block->instructions[1]->As(); + auto* var_b = instr[1]->As(); ASSERT_NE(var_b, nullptr); - auto* var_c = m.Get().root_block->instructions[2]->As(); + auto* var_c = instr[2]->As(); ASSERT_NE(var_c, nullptr); - ASSERT_EQ(var_a->initializer, var_c->initializer); - ASSERT_NE(var_a->initializer, var_b->initializer); + ASSERT_EQ(var_a->Initializer(), var_c->Initializer()); + ASSERT_NE(var_a->Initializer(), var_b->Initializer()); } } // namespace diff --git a/src/tint/ir/from_program_test.cc b/src/tint/ir/from_program_test.cc index 773fde6642..1049386c3e 100644 --- a/src/tint/ir/from_program_test.cc +++ b/src/tint/ir/from_program_test.cc @@ -63,13 +63,13 @@ TEST_F(IR_BuilderImplTest, Func) { ASSERT_EQ(1u, m->functions.Length()); auto* f = m->functions[0]; - ASSERT_NE(f->start_target, nullptr); - ASSERT_NE(f->end_target, nullptr); + ASSERT_NE(f->StartTarget(), nullptr); + ASSERT_NE(f->EndTarget(), nullptr); - EXPECT_EQ(1u, f->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, f->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, f->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, f->EndTarget()->InboundBranches().Length()); - EXPECT_EQ(m->functions[0]->pipeline_stage, Function::PipelineStage::kUndefined); + EXPECT_EQ(m->functions[0]->Stage(), Function::PipelineStage::kUndefined); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func f():void { %fn2 = block { @@ -88,13 +88,13 @@ TEST_F(IR_BuilderImplTest, Func_WithParam) { ASSERT_EQ(1u, m->functions.Length()); auto* f = m->functions[0]; - ASSERT_NE(f->start_target, nullptr); - ASSERT_NE(f->end_target, nullptr); + ASSERT_NE(f->StartTarget(), nullptr); + ASSERT_NE(f->EndTarget(), nullptr); - EXPECT_EQ(1u, f->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, f->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, f->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, f->EndTarget()->InboundBranches().Length()); - EXPECT_EQ(m->functions[0]->pipeline_stage, Function::PipelineStage::kUndefined); + EXPECT_EQ(m->functions[0]->Stage(), Function::PipelineStage::kUndefined); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func f(%a:u32):u32 { %fn2 = block { @@ -114,13 +114,13 @@ TEST_F(IR_BuilderImplTest, Func_WithMultipleParam) { ASSERT_EQ(1u, m->functions.Length()); auto* f = m->functions[0]; - ASSERT_NE(f->start_target, nullptr); - ASSERT_NE(f->end_target, nullptr); + ASSERT_NE(f->StartTarget(), nullptr); + ASSERT_NE(f->EndTarget(), nullptr); - EXPECT_EQ(1u, f->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, f->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, f->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, f->EndTarget()->InboundBranches().Length()); - EXPECT_EQ(m->functions[0]->pipeline_stage, Function::PipelineStage::kUndefined); + EXPECT_EQ(m->functions[0]->Stage(), Function::PipelineStage::kUndefined); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func f(%a:u32, %b:i32, %c:bool):void { %fn2 = block { @@ -137,7 +137,7 @@ TEST_F(IR_BuilderImplTest, EntryPoint) { auto m = Build(); ASSERT_TRUE(m) << (!m ? m.Failure() : ""); - EXPECT_EQ(m->functions[0]->pipeline_stage, Function::PipelineStage::kFragment); + EXPECT_EQ(m->functions[0]->Stage(), Function::PipelineStage::kFragment); } TEST_F(IR_BuilderImplTest, IfStatement) { @@ -148,19 +148,19 @@ TEST_F(IR_BuilderImplTest, IfStatement) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->true_.target, nullptr); - ASSERT_NE(flow->false_.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->True().target, nullptr); + ASSERT_NE(flow->False().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(2u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->False().target->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -192,19 +192,19 @@ TEST_F(IR_BuilderImplTest, IfStatement_TrueReturns) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->true_.target, nullptr); - ASSERT_NE(flow->false_.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->True().target, nullptr); + ASSERT_NE(flow->False().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(2u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(2u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -235,19 +235,19 @@ TEST_F(IR_BuilderImplTest, IfStatement_FalseReturns) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->true_.target, nullptr); - ASSERT_NE(flow->false_.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->True().target, nullptr); + ASSERT_NE(flow->False().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(2u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(2u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -278,19 +278,19 @@ TEST_F(IR_BuilderImplTest, IfStatement_BothReturn) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->true_.target, nullptr); - ASSERT_NE(flow->false_.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->True().target, nullptr); + ASSERT_NE(flow->False().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(0u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(2u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->False().target->InboundBranches().Length()); + EXPECT_EQ(0u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(2u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -318,15 +318,15 @@ TEST_F(IR_BuilderImplTest, IfStatement_JumpChainToMerge) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); auto* loop_flow = FindSingleFlowNode(m.Get()); ASSERT_NE(loop_flow, nullptr); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -367,19 +367,19 @@ TEST_F(IR_BuilderImplTest, Loop_WithBreak) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->start.target, nullptr); - ASSERT_NE(flow->continuing.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target, nullptr); + ASSERT_NE(flow->Continuing().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(2u, flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -408,28 +408,28 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinue) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); auto* if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -475,28 +475,28 @@ TEST_F(IR_BuilderImplTest, Loop_WithContinuing_BreakIf) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); auto* break_if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(break_if_flow->true_.target, nullptr); - ASSERT_NE(break_if_flow->false_.target, nullptr); - ASSERT_NE(break_if_flow->merge.target, nullptr); + ASSERT_NE(break_if_flow->True().target, nullptr); + ASSERT_NE(break_if_flow->False().target, nullptr); + ASSERT_NE(break_if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, break_if_flow->inbound_branches.Length()); - EXPECT_EQ(1u, break_if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, break_if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, break_if_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, break_if_flow->InboundBranches().Length()); + EXPECT_EQ(1u, break_if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, break_if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, break_if_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -586,28 +586,28 @@ TEST_F(IR_BuilderImplTest, Loop_WithReturn) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); auto* if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -648,19 +648,19 @@ TEST_F(IR_BuilderImplTest, Loop_WithOnlyReturn) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -693,25 +693,25 @@ TEST_F(IR_BuilderImplTest, Loop_WithOnlyReturn_ContinuingBreakIf) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); auto* break_if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(break_if_flow->true_.target, nullptr); - ASSERT_NE(break_if_flow->false_.target, nullptr); - ASSERT_NE(break_if_flow->merge.target, nullptr); + ASSERT_NE(break_if_flow->True().target, nullptr); + ASSERT_NE(break_if_flow->False().target, nullptr); + ASSERT_NE(break_if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); // This is 1 because only the loop branch happens. The subsequent if return is dead code. - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -736,28 +736,28 @@ TEST_F(IR_BuilderImplTest, Loop_WithIf_BothBranchesBreak) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* loop_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(loop_flow->start.target, nullptr); - ASSERT_NE(loop_flow->continuing.target, nullptr); - ASSERT_NE(loop_flow->merge.target, nullptr); + ASSERT_NE(loop_flow->Start().target, nullptr); + ASSERT_NE(loop_flow->Continuing().target, nullptr); + ASSERT_NE(loop_flow->Merge().target, nullptr); auto* if_flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(0u, if_flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(0u, if_flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -807,96 +807,96 @@ TEST_F(IR_BuilderImplTest, Loop_Nested) { auto block_exit = [&](const ir::FlowNode* node) -> const ir::FlowNode* { if (auto* block = As(node)) { - return block->branch.target; + return block->Branch().target; } return nullptr; }; - auto* loop_flow_a = As(m->functions[0]->start_target->branch.target); + auto* loop_flow_a = As(m->functions[0]->StartTarget()->Branch().target); ASSERT_NE(loop_flow_a, nullptr); - ASSERT_NE(loop_flow_a->start.target, nullptr); - ASSERT_NE(loop_flow_a->continuing.target, nullptr); - ASSERT_NE(loop_flow_a->merge.target, nullptr); + ASSERT_NE(loop_flow_a->Start().target, nullptr); + ASSERT_NE(loop_flow_a->Continuing().target, nullptr); + ASSERT_NE(loop_flow_a->Merge().target, nullptr); - auto* loop_flow_b = As(block_exit(loop_flow_a->start.target)); + auto* loop_flow_b = As(block_exit(loop_flow_a->Start().target)); ASSERT_NE(loop_flow_b, nullptr); - ASSERT_NE(loop_flow_b->start.target, nullptr); - ASSERT_NE(loop_flow_b->continuing.target, nullptr); - ASSERT_NE(loop_flow_b->merge.target, nullptr); + ASSERT_NE(loop_flow_b->Start().target, nullptr); + ASSERT_NE(loop_flow_b->Continuing().target, nullptr); + ASSERT_NE(loop_flow_b->Merge().target, nullptr); - auto* if_flow_a = As(block_exit(loop_flow_b->start.target)); + auto* if_flow_a = As(block_exit(loop_flow_b->Start().target)); ASSERT_NE(if_flow_a, nullptr); - ASSERT_NE(if_flow_a->true_.target, nullptr); - ASSERT_NE(if_flow_a->false_.target, nullptr); - ASSERT_NE(if_flow_a->merge.target, nullptr); + ASSERT_NE(if_flow_a->True().target, nullptr); + ASSERT_NE(if_flow_a->False().target, nullptr); + ASSERT_NE(if_flow_a->Merge().target, nullptr); - auto* if_flow_b = As(block_exit(if_flow_a->merge.target)); + auto* if_flow_b = As(block_exit(if_flow_a->Merge().target)); ASSERT_NE(if_flow_b, nullptr); - ASSERT_NE(if_flow_b->true_.target, nullptr); - ASSERT_NE(if_flow_b->false_.target, nullptr); - ASSERT_NE(if_flow_b->merge.target, nullptr); + ASSERT_NE(if_flow_b->True().target, nullptr); + ASSERT_NE(if_flow_b->False().target, nullptr); + ASSERT_NE(if_flow_b->Merge().target, nullptr); - auto* loop_flow_c = As(block_exit(loop_flow_b->continuing.target)); + auto* loop_flow_c = As(block_exit(loop_flow_b->Continuing().target)); ASSERT_NE(loop_flow_c, nullptr); - ASSERT_NE(loop_flow_c->start.target, nullptr); - ASSERT_NE(loop_flow_c->continuing.target, nullptr); - ASSERT_NE(loop_flow_c->merge.target, nullptr); + ASSERT_NE(loop_flow_c->Start().target, nullptr); + ASSERT_NE(loop_flow_c->Continuing().target, nullptr); + ASSERT_NE(loop_flow_c->Merge().target, nullptr); - auto* loop_flow_d = As(block_exit(loop_flow_c->merge.target)); + auto* loop_flow_d = As(block_exit(loop_flow_c->Merge().target)); ASSERT_NE(loop_flow_d, nullptr); - ASSERT_NE(loop_flow_d->start.target, nullptr); - ASSERT_NE(loop_flow_d->continuing.target, nullptr); - ASSERT_NE(loop_flow_d->merge.target, nullptr); + ASSERT_NE(loop_flow_d->Start().target, nullptr); + ASSERT_NE(loop_flow_d->Continuing().target, nullptr); + ASSERT_NE(loop_flow_d->Merge().target, nullptr); - auto* if_flow_c = As(block_exit(loop_flow_d->continuing.target)); + auto* if_flow_c = As(block_exit(loop_flow_d->Continuing().target)); ASSERT_NE(if_flow_c, nullptr); - ASSERT_NE(if_flow_c->true_.target, nullptr); - ASSERT_NE(if_flow_c->false_.target, nullptr); - ASSERT_NE(if_flow_c->merge.target, nullptr); + ASSERT_NE(if_flow_c->True().target, nullptr); + ASSERT_NE(if_flow_c->False().target, nullptr); + ASSERT_NE(if_flow_c->Merge().target, nullptr); - auto* if_flow_d = As(block_exit(loop_flow_b->merge.target)); + auto* if_flow_d = As(block_exit(loop_flow_b->Merge().target)); ASSERT_NE(if_flow_d, nullptr); - ASSERT_NE(if_flow_d->true_.target, nullptr); - ASSERT_NE(if_flow_d->false_.target, nullptr); - ASSERT_NE(if_flow_d->merge.target, nullptr); + ASSERT_NE(if_flow_d->True().target, nullptr); + ASSERT_NE(if_flow_d->False().target, nullptr); + ASSERT_NE(if_flow_d->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, loop_flow_a->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow_a->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_a->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_a->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_b->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow_b->start.target->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow_b->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_b->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_c->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow_c->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, loop_flow_c->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_c->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_d->inbound_branches.Length()); - EXPECT_EQ(2u, loop_flow_d->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_d->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, loop_flow_d->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_a->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_a->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_a->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_a->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_b->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_b->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_b->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_b->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_c->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_c->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_c->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_c->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_d->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_d->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_d->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow_d->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->start_target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, loop_flow_a->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow_a->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_a->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_a->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_b->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow_b->Start().target->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow_b->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_b->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_c->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow_c->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, loop_flow_c->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_c->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_d->InboundBranches().Length()); + EXPECT_EQ(2u, loop_flow_d->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_d->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, loop_flow_d->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_a->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_a->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_a->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_a->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_b->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_b->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_b->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_b->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_c->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_c->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_c->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_c->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_d->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_d->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_d->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow_d->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->StartTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1015,28 +1015,28 @@ TEST_F(IR_BuilderImplTest, While) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->start.target, nullptr); - ASSERT_NE(flow->continuing.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target, nullptr); + ASSERT_NE(flow->Continuing().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); - ASSERT_NE(flow->start.target->As()->branch.target, nullptr); - ASSERT_TRUE(flow->start.target->As()->branch.target->Is()); - auto* if_flow = flow->start.target->As()->branch.target->As(); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target->As()->Branch().target, nullptr); + ASSERT_TRUE(flow->Start().target->As()->Branch().target->Is()); + auto* if_flow = flow->Start().target->As()->Branch().target->As(); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(2u, flow->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->merge.target->inbound_branches.Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->Merge().target->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1081,28 +1081,28 @@ TEST_F(IR_BuilderImplTest, While_Return) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->start.target, nullptr); - ASSERT_NE(flow->continuing.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target, nullptr); + ASSERT_NE(flow->Continuing().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); - ASSERT_NE(flow->start.target->As()->branch.target, nullptr); - ASSERT_TRUE(flow->start.target->As()->branch.target->Is()); - auto* if_flow = flow->start.target->As()->branch.target->As(); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target->As()->Branch().target, nullptr); + ASSERT_TRUE(flow->Start().target->As()->Branch().target->Is()); + auto* if_flow = flow->Start().target->As()->Branch().target->As(); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(2u, func->end_target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(2u, flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->merge.target->inbound_branches.Length()); + EXPECT_EQ(2u, func->EndTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->Merge().target->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1155,28 +1155,28 @@ TEST_F(IR_BuilderImplTest, DISABLED_For) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->start.target, nullptr); - ASSERT_NE(flow->continuing.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target, nullptr); + ASSERT_NE(flow->Continuing().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); - ASSERT_NE(flow->start.target->As()->branch.target, nullptr); - ASSERT_TRUE(flow->start.target->As()->branch.target->Is()); - auto* if_flow = flow->start.target->As()->branch.target->As(); - ASSERT_NE(if_flow->true_.target, nullptr); - ASSERT_NE(if_flow->false_.target, nullptr); - ASSERT_NE(if_flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target->As()->Branch().target, nullptr); + ASSERT_TRUE(flow->Start().target->As()->Branch().target->Is()); + auto* if_flow = flow->Start().target->As()->Branch().target->As(); + ASSERT_NE(if_flow->True().target, nullptr); + ASSERT_NE(if_flow->False().target, nullptr); + ASSERT_NE(if_flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(2u, flow->start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->true_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->false_.target->inbound_branches.Length()); - EXPECT_EQ(1u, if_flow->merge.target->inbound_branches.Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->True().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->False().target->InboundBranches().Length()); + EXPECT_EQ(1u, if_flow->Merge().target->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"()"); } @@ -1189,18 +1189,18 @@ TEST_F(IR_BuilderImplTest, For_NoInitCondOrContinuing) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->start.target, nullptr); - ASSERT_NE(flow->continuing.target, nullptr); - ASSERT_NE(flow->merge.target, nullptr); + ASSERT_NE(flow->Start().target, nullptr); + ASSERT_NE(flow->Continuing().target, nullptr); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(2u, flow->start.target->inbound_branches.Length()); - EXPECT_EQ(0u, flow->continuing.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, flow->Continuing().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1231,31 +1231,33 @@ TEST_F(IR_BuilderImplTest, Switch) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->merge.target, nullptr); - ASSERT_EQ(3u, flow->cases.Length()); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - ASSERT_EQ(1u, flow->cases[0].selectors.Length()); - ASSERT_TRUE(flow->cases[0].selectors[0].val->value->Is>()); + auto cases = flow->Cases(); + ASSERT_EQ(3u, cases.Length()); + + ASSERT_EQ(1u, cases[0].selectors.Length()); + ASSERT_TRUE(cases[0].selectors[0].val->Value()->Is>()); EXPECT_EQ(0_i, - flow->cases[0].selectors[0].val->value->As>()->ValueOf()); + cases[0].selectors[0].val->Value()->As>()->ValueOf()); - ASSERT_EQ(1u, flow->cases[1].selectors.Length()); - ASSERT_TRUE(flow->cases[1].selectors[0].val->value->Is>()); + ASSERT_EQ(1u, cases[1].selectors.Length()); + ASSERT_TRUE(cases[1].selectors[0].val->Value()->Is>()); EXPECT_EQ(1_i, - flow->cases[1].selectors[0].val->value->As>()->ValueOf()); + cases[1].selectors[0].val->Value()->As>()->ValueOf()); - ASSERT_EQ(1u, flow->cases[2].selectors.Length()); - EXPECT_TRUE(flow->cases[2].selectors[0].IsDefault()); + ASSERT_EQ(1u, cases[2].selectors.Length()); + EXPECT_TRUE(cases[2].selectors[0].IsDefault()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[0].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[1].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[2].start.target->inbound_branches.Length()); - EXPECT_EQ(3u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, cases[0].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, cases[1].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, cases[2].Start().target->InboundBranches().Length()); + EXPECT_EQ(3u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1295,27 +1297,28 @@ TEST_F(IR_BuilderImplTest, Switch_MultiSelector) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->merge.target, nullptr); - ASSERT_EQ(1u, flow->cases.Length()); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - ASSERT_EQ(3u, flow->cases[0].selectors.Length()); - ASSERT_TRUE(flow->cases[0].selectors[0].val->value->Is>()); + auto cases = flow->Cases(); + ASSERT_EQ(1u, cases.Length()); + ASSERT_EQ(3u, cases[0].selectors.Length()); + ASSERT_TRUE(cases[0].selectors[0].val->Value()->Is>()); EXPECT_EQ(0_i, - flow->cases[0].selectors[0].val->value->As>()->ValueOf()); + cases[0].selectors[0].val->Value()->As>()->ValueOf()); - ASSERT_TRUE(flow->cases[0].selectors[1].val->value->Is>()); + ASSERT_TRUE(cases[0].selectors[1].val->Value()->Is>()); EXPECT_EQ(1_i, - flow->cases[0].selectors[1].val->value->As>()->ValueOf()); + cases[0].selectors[1].val->Value()->As>()->ValueOf()); - EXPECT_TRUE(flow->cases[0].selectors[2].IsDefault()); + EXPECT_TRUE(cases[0].selectors[2].IsDefault()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[0].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, cases[0].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1343,19 +1346,20 @@ TEST_F(IR_BuilderImplTest, Switch_OnlyDefault) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->merge.target, nullptr); - ASSERT_EQ(1u, flow->cases.Length()); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - ASSERT_EQ(1u, flow->cases[0].selectors.Length()); - EXPECT_TRUE(flow->cases[0].selectors[0].IsDefault()); + auto cases = flow->Cases(); + ASSERT_EQ(1u, cases.Length()); + ASSERT_EQ(1u, cases[0].selectors.Length()); + EXPECT_TRUE(cases[0].selectors[0].IsDefault()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[0].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, cases[0].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1385,26 +1389,27 @@ TEST_F(IR_BuilderImplTest, Switch_WithBreak) { ASSERT_TRUE(m) << (!m ? m.Failure() : ""); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->merge.target, nullptr); - ASSERT_EQ(2u, flow->cases.Length()); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - ASSERT_EQ(1u, flow->cases[0].selectors.Length()); - ASSERT_TRUE(flow->cases[0].selectors[0].val->value->Is>()); + auto cases = flow->Cases(); + ASSERT_EQ(2u, cases.Length()); + ASSERT_EQ(1u, cases[0].selectors.Length()); + ASSERT_TRUE(cases[0].selectors[0].val->Value()->Is>()); EXPECT_EQ(0_i, - flow->cases[0].selectors[0].val->value->As>()->ValueOf()); + cases[0].selectors[0].val->Value()->As>()->ValueOf()); - ASSERT_EQ(1u, flow->cases[1].selectors.Length()); - EXPECT_TRUE(flow->cases[1].selectors[0].IsDefault()); + ASSERT_EQ(1u, cases[1].selectors.Length()); + EXPECT_TRUE(cases[1].selectors[0].IsDefault()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[0].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[1].start.target->inbound_branches.Length()); - EXPECT_EQ(2u, flow->merge.target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, cases[0].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, cases[1].Start().target->InboundBranches().Length()); + EXPECT_EQ(2u, flow->Merge().target->InboundBranches().Length()); // This is 1 because the if is dead-code eliminated and the return doesn't happen. - EXPECT_EQ(1u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { @@ -1441,25 +1446,26 @@ TEST_F(IR_BuilderImplTest, Switch_AllReturn) { ASSERT_EQ(FindSingleFlowNode(m.Get()), nullptr); auto* flow = FindSingleFlowNode(m.Get()); - ASSERT_NE(flow->merge.target, nullptr); - ASSERT_EQ(2u, flow->cases.Length()); + ASSERT_NE(flow->Merge().target, nullptr); ASSERT_EQ(1u, m->functions.Length()); auto* func = m->functions[0]; - ASSERT_EQ(1u, flow->cases[0].selectors.Length()); - ASSERT_TRUE(flow->cases[0].selectors[0].val->value->Is>()); + auto cases = flow->Cases(); + ASSERT_EQ(2u, cases.Length()); + ASSERT_EQ(1u, cases[0].selectors.Length()); + ASSERT_TRUE(cases[0].selectors[0].val->Value()->Is>()); EXPECT_EQ(0_i, - flow->cases[0].selectors[0].val->value->As>()->ValueOf()); + cases[0].selectors[0].val->Value()->As>()->ValueOf()); - ASSERT_EQ(1u, flow->cases[1].selectors.Length()); - EXPECT_TRUE(flow->cases[1].selectors[0].IsDefault()); + ASSERT_EQ(1u, cases[1].selectors.Length()); + EXPECT_TRUE(cases[1].selectors[0].IsDefault()); - EXPECT_EQ(1u, flow->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[0].start.target->inbound_branches.Length()); - EXPECT_EQ(1u, flow->cases[1].start.target->inbound_branches.Length()); - EXPECT_EQ(0u, flow->merge.target->inbound_branches.Length()); - EXPECT_EQ(2u, func->end_target->inbound_branches.Length()); + EXPECT_EQ(1u, flow->InboundBranches().Length()); + EXPECT_EQ(1u, cases[0].Start().target->InboundBranches().Length()); + EXPECT_EQ(1u, cases[1].Start().target->InboundBranches().Length()); + EXPECT_EQ(0u, flow->Merge().target->InboundBranches().Length()); + EXPECT_EQ(2u, func->EndTarget()->InboundBranches().Length()); EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] { diff --git a/src/tint/ir/function.cc b/src/tint/ir/function.cc index 2a5f8c3801..d6cd6fc17c 100644 --- a/src/tint/ir/function.cc +++ b/src/tint/ir/function.cc @@ -18,11 +18,11 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Function); namespace tint::ir { -Function::Function(Symbol n, +Function::Function(Symbol name, type::Type* rt, PipelineStage stage, std::optional> wg_size) - : Base(), name(n), pipeline_stage(stage), workgroup_size(wg_size), return_type(rt) {} + : Base(), name_(name), return_type_(rt), pipeline_stage_(stage), workgroup_size_(wg_size) {} Function::~Function() = default; diff --git a/src/tint/ir/function.h b/src/tint/ir/function.h index 1fc651c117..86465aaf6c 100644 --- a/src/tint/ir/function.h +++ b/src/tint/ir/function.h @@ -17,6 +17,7 @@ #include #include +#include #include "src/tint/ir/flow_node.h" #include "src/tint/ir/function_param.h" @@ -71,32 +72,81 @@ class Function : public utils::Castable { type::Type* rt, PipelineStage stage = PipelineStage::kUndefined, std::optional> wg_size = {}); + Function(Function&&) = delete; + Function(const Function&) = delete; ~Function() override; - /// The function name - Symbol name; + Function& operator=(Function&&) = delete; + Function& operator=(const Function&) = delete; - /// The pipeline stage for the function, `kUndefined` if the function is not an entry point - PipelineStage pipeline_stage = PipelineStage::kUndefined; + /// @returns the function name + Symbol Name() const { return name_; } - /// If this is a `compute` entry point, holds the workgroup size information - std::optional> workgroup_size; + /// Sets the function stage + /// @param stage the stage to set + void SetStage(PipelineStage stage) { pipeline_stage_ = stage; } - /// The function return type - const type::Type* return_type = nullptr; - /// The function return attributes if any - utils::Vector return_attributes; - /// If the return attribute is `kLocation` this stores the location value. - std::optional return_location; + /// @returns the function pipeline stage + PipelineStage Stage() const { return pipeline_stage_; } - /// The parameters to the function - utils::Vector params; + /// Sets the workgroup size + /// @param x the x size + /// @param y the y size + /// @param z the z size + void SetWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { workgroup_size_ = {x, y, z}; } - /// The start target is the first block in a function. - Block* start_target = nullptr; - /// The end target is the end of the function. It is used as the branch target if a return is - /// encountered in the function. - FunctionTerminator* end_target = nullptr; + /// @returns the workgroup size information + std::optional> WorkgroupSize() const { return workgroup_size_; } + + /// @returns the return type for the function + const type::Type* ReturnType() const { return return_type_; } + + /// Sets the return attributes + /// @param attrs the attributes to set + void SetReturnAttributes(utils::VectorRef attrs) { + return_attributes_ = std::move(attrs); + } + /// @returns the return attributes + utils::VectorRef ReturnAttributes() const { return return_attributes_; } + + /// Sets the return location + /// @param loc the location to set + void SetReturnLocation(std::optional loc) { return_location_ = loc; } + /// @returns the return location + std::optional ReturnLocation() const { return return_location_; } + + /// Sets the function parameters + /// @param params the function paramters + void SetParams(utils::VectorRef params) { params_ = std::move(params); } + + /// @returns the function parameters + utils::VectorRef Params() const { return params_; } + + /// Sets the start target for the function + /// @param target the start target + void SetStartTarget(Block* target) { start_target_ = target; } + /// @returns the function start target + Block* StartTarget() const { return start_target_; } + + /// Sets the end target for the function + /// @param target the end target + void SetEndTarget(FunctionTerminator* target) { end_target_ = target; } + /// @returns the function end target + FunctionTerminator* EndTarget() const { return end_target_; } + + private: + Symbol name_; + const type::Type* return_type_; + PipelineStage pipeline_stage_; + std::optional> workgroup_size_; + + utils::Vector return_attributes_; + std::optional return_location_; + + utils::Vector params_; + + Block* start_target_ = nullptr; + FunctionTerminator* end_target_ = nullptr; }; utils::StringStream& operator<<(utils::StringStream& out, Function::PipelineStage value); diff --git a/src/tint/ir/function_param.cc b/src/tint/ir/function_param.cc index 8a05d6a797..a9570faa8d 100644 --- a/src/tint/ir/function_param.cc +++ b/src/tint/ir/function_param.cc @@ -18,7 +18,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::FunctionParam); namespace tint::ir { -FunctionParam::FunctionParam(const type::Type* ty) : type(ty) {} +FunctionParam::FunctionParam(const type::Type* ty) : type_(ty) {} FunctionParam::~FunctionParam() = default; diff --git a/src/tint/ir/function_param.h b/src/tint/ir/function_param.h index 46a3d5ee66..1bbb812943 100644 --- a/src/tint/ir/function_param.h +++ b/src/tint/ir/function_param.h @@ -34,10 +34,11 @@ class FunctionParam : public utils::Castable { FunctionParam& operator=(FunctionParam&& inst) = delete; /// @returns the type of the var - const type::Type* Type() const override { return type; } + const type::Type* Type() const override { return type_; } + private: /// The type of the parameter - const type::Type* type; + const type::Type* type_; }; } // namespace tint::ir diff --git a/src/tint/ir/if.cc b/src/tint/ir/if.cc index b59d87f56d..d235b5efd3 100644 --- a/src/tint/ir/if.cc +++ b/src/tint/ir/if.cc @@ -18,7 +18,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::If); namespace tint::ir { -If::If(Value* cond) : Base(), condition(cond) {} +If::If(Value* cond) : Base(), condition_(cond) {} If::~If() = default; diff --git a/src/tint/ir/if.h b/src/tint/ir/if.h index aadc5c9912..d02c32a425 100644 --- a/src/tint/ir/if.h +++ b/src/tint/ir/if.h @@ -32,17 +32,36 @@ class If : public utils::Castable { /// Constructor /// @param cond the if condition explicit If(Value* cond); + If(const If&) = delete; + If(If&&) = delete; ~If() override; - /// The true branch block + If& operator=(const If&) = delete; + If& operator=(If&&) = delete; + + /// @returns the if condition + const Value* Condition() const { return condition_; } + + /// @returns the true branch block + const Branch& True() const { return true_; } + /// @returns the true branch block + Branch& True() { return true_; } + + /// @returns the false branch block + const Branch& False() const { return false_; } + /// @returns the false branch block + Branch& False() { return false_; } + + /// @returns the merge branch block + const Branch& Merge() const { return merge_; } + /// @returns the merge branch block + Branch& Merge() { return merge_; } + + private: Branch true_ = {}; - /// The false branch block Branch false_ = {}; - /// An block to converge the true/false branches. The block always exists, but there maybe no - /// branches into it. (e.g. if both branches `return`) - Branch merge = {}; - /// Value holding the condition result - const Value* condition = nullptr; + Branch merge_ = {}; + Value* condition_; }; } // namespace tint::ir diff --git a/src/tint/ir/load.cc b/src/tint/ir/load.cc index 1fe55c01bb..3b64cdd752 100644 --- a/src/tint/ir/load.cc +++ b/src/tint/ir/load.cc @@ -19,10 +19,10 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Load); namespace tint::ir { -Load::Load(const type::Type* type, Value* f) : Base(), result_type(type), from(f) { - TINT_ASSERT(IR, result_type); - TINT_ASSERT(IR, from); - from->AddUsage(this); +Load::Load(const type::Type* type, Value* f) : Base(), result_type_(type), from_(f) { + TINT_ASSERT(IR, result_type_); + TINT_ASSERT(IR, from_); + from_->AddUsage(this); } Load::~Load() = default; diff --git a/src/tint/ir/load.h b/src/tint/ir/load.h index b15ecedc2e..e1a365fcdf 100644 --- a/src/tint/ir/load.h +++ b/src/tint/ir/load.h @@ -35,13 +35,14 @@ class Load : public utils::Castable { Load& operator=(Load&& inst) = delete; /// @returns the type of the value - const type::Type* Type() const override { return result_type; } + const type::Type* Type() const override { return result_type_; } - /// the result type of the instruction - const type::Type* result_type = nullptr; + /// @returns the avlue being loaded from + Value* From() const { return from_; } - /// the value being loaded - Value* from = nullptr; + private: + const type::Type* result_type_; + Value* from_; }; } // namespace tint::ir diff --git a/src/tint/ir/load_test.cc b/src/tint/ir/load_test.cc index 2c6e5c1402..9881e98a2c 100644 --- a/src/tint/ir/load_test.cc +++ b/src/tint/ir/load_test.cc @@ -33,12 +33,12 @@ TEST_F(IR_InstructionTest, CreateLoad) { const auto* inst = b.Load(var); ASSERT_TRUE(inst->Is()); - ASSERT_EQ(inst->from, var); + ASSERT_EQ(inst->From(), var); EXPECT_EQ(inst->Type(), store_type); - ASSERT_TRUE(inst->from->Is()); - EXPECT_EQ(inst->from, var); + ASSERT_TRUE(inst->From()->Is()); + EXPECT_EQ(inst->From(), var); } TEST_F(IR_InstructionTest, Load_Usage) { @@ -50,9 +50,9 @@ TEST_F(IR_InstructionTest, Load_Usage) { store_type, builtin::AddressSpace::kFunction, builtin::Access::kReadWrite)); const auto* inst = b.Load(var); - ASSERT_NE(inst->from, nullptr); - ASSERT_EQ(inst->from->Usage().Length(), 1u); - EXPECT_EQ(inst->from->Usage()[0], inst); + ASSERT_NE(inst->From(), nullptr); + ASSERT_EQ(inst->From()->Usage().Length(), 1u); + EXPECT_EQ(inst->From()->Usage()[0], inst); } } // namespace diff --git a/src/tint/ir/loop.h b/src/tint/ir/loop.h index e0066f4465..590f49c567 100644 --- a/src/tint/ir/loop.h +++ b/src/tint/ir/loop.h @@ -26,16 +26,32 @@ class Loop : public utils::Castable { public: /// Constructor Loop(); + Loop(const Loop&) = delete; + Loop(Loop&&) = delete; ~Loop() override; - /// The start block is the first block in a loop. - Branch start = {}; - /// The continue target of the block. - Branch continuing = {}; - /// The loop merge target. If the `loop` does a `return` then this block may not actually - /// end up in the control flow. We need it if the loop does a `break` we know where to break - /// too. - Branch merge = {}; + Loop& operator=(const Loop&) = delete; + Loop& operator=(Loop&&) = delete; + + /// @returns the switch start branch + const Branch& Start() const { return start_; } + /// @returns the switch start branch + Branch& Start() { return start_; } + + /// @returns the switch continuing branch + const Branch& Continuing() const { return continuing_; } + /// @returns the switch continuing branch + Branch& Continuing() { return continuing_; } + + /// @returns the switch merge branch + const Branch& Merge() const { return merge_; } + /// @returns the switch merge branch + Branch& Merge() { return merge_; } + + private: + Branch start_ = {}; + Branch continuing_ = {}; + Branch merge_ = {}; }; } // namespace tint::ir diff --git a/src/tint/ir/root_terminator.h b/src/tint/ir/root_terminator.h index 361aa6d464..c6338ba4b4 100644 --- a/src/tint/ir/root_terminator.h +++ b/src/tint/ir/root_terminator.h @@ -25,7 +25,12 @@ class RootTerminator : public utils::Castable { public: /// Constructor RootTerminator(); + RootTerminator(const RootTerminator&) = delete; + RootTerminator(RootTerminator&&) = delete; ~RootTerminator() override; + + RootTerminator& operator=(const RootTerminator&) = delete; + RootTerminator& operator=(RootTerminator&&) = delete; }; } // namespace tint::ir diff --git a/src/tint/ir/store.cc b/src/tint/ir/store.cc index 87f3620da5..8b8de4681b 100644 --- a/src/tint/ir/store.cc +++ b/src/tint/ir/store.cc @@ -19,11 +19,11 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Store); namespace tint::ir { -Store::Store(Value* t, Value* f) : Base(), to(t), from(f) { - TINT_ASSERT(IR, to); - TINT_ASSERT(IR, from); - to->AddUsage(this); - from->AddUsage(this); +Store::Store(Value* to, Value* from) : Base(), to_(to), from_(from) { + TINT_ASSERT(IR, to_); + TINT_ASSERT(IR, from_); + to_->AddUsage(this); + from_->AddUsage(this); } Store::~Store() = default; diff --git a/src/tint/ir/store.h b/src/tint/ir/store.h index 9095c41224..af5377df55 100644 --- a/src/tint/ir/store.h +++ b/src/tint/ir/store.h @@ -34,10 +34,15 @@ class Store : public utils::Castable { Store& operator=(const Store& inst) = delete; Store& operator=(Store&& inst) = delete; - /// the value being stored to - Value* to = nullptr; - /// the value being stored - Value* from = nullptr; + /// @returns the value being stored too + Value* To() const { return to_; } + + /// @returns the value being stored + Value* From() const { return from_; } + + private: + Value* to_; + Value* from_; }; } // namespace tint::ir diff --git a/src/tint/ir/store_test.cc b/src/tint/ir/store_test.cc index 902ca955d9..1906cc6ff1 100644 --- a/src/tint/ir/store_test.cc +++ b/src/tint/ir/store_test.cc @@ -33,10 +33,10 @@ TEST_F(IR_InstructionTest, CreateStore) { const auto* inst = b.Store(to, b.Constant(4_i)); ASSERT_TRUE(inst->Is()); - ASSERT_EQ(inst->to, to); + ASSERT_EQ(inst->To(), to); - ASSERT_TRUE(inst->from->Is()); - auto lhs = inst->from->As()->value; + ASSERT_TRUE(inst->From()->Is()); + auto lhs = inst->From()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); } @@ -48,13 +48,13 @@ TEST_F(IR_InstructionTest, Store_Usage) { auto* to = b.Discard(); const auto* inst = b.Store(to, b.Constant(4_i)); - ASSERT_NE(inst->to, nullptr); - ASSERT_EQ(inst->to->Usage().Length(), 1u); - EXPECT_EQ(inst->to->Usage()[0], inst); + ASSERT_NE(inst->To(), nullptr); + ASSERT_EQ(inst->To()->Usage().Length(), 1u); + EXPECT_EQ(inst->To()->Usage()[0], inst); - ASSERT_NE(inst->from, nullptr); - ASSERT_EQ(inst->from->Usage().Length(), 1u); - EXPECT_EQ(inst->from->Usage()[0], inst); + ASSERT_NE(inst->From(), nullptr); + ASSERT_EQ(inst->From()->Usage().Length(), 1u); + EXPECT_EQ(inst->From()->Usage()[0], inst); } } // namespace diff --git a/src/tint/ir/switch.cc b/src/tint/ir/switch.cc index ad6a1453f7..3bccc83677 100644 --- a/src/tint/ir/switch.cc +++ b/src/tint/ir/switch.cc @@ -18,7 +18,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Switch); namespace tint::ir { -Switch::Switch(Value* cond) : Base(), condition(cond) {} +Switch::Switch(Value* cond) : Base(), condition_(cond) {} Switch::~Switch() = default; diff --git a/src/tint/ir/switch.h b/src/tint/ir/switch.h index 6be7b62efc..2e977eced7 100644 --- a/src/tint/ir/switch.h +++ b/src/tint/ir/switch.h @@ -41,21 +41,40 @@ class Switch : public utils::Castable { utils::Vector selectors; /// The start block for the case block. Branch start = {}; + + /// @returns the case start target + const Branch& Start() const { return start; } + /// @returns the case start target + Branch& Start() { return start; } }; /// Constructor /// @param cond the condition explicit Switch(Value* cond); + Switch(const Switch&) = delete; + Switch(Switch&&) = delete; ~Switch() override; - /// The switch merge target - Branch merge = {}; + Switch& operator=(const Switch&) = delete; + Switch& operator=(Switch&&) = delete; - /// The switch case statements - utils::Vector cases; + /// @returns the switch merge branch + const Branch& Merge() const { return merge_; } + /// @returns the switch merge branch + Branch& Merge() { return merge_; } - /// Value holding the condition result - const Value* condition = nullptr; + /// @returns the switch cases + utils::VectorRef Cases() const { return cases_; } + /// @returns the switch cases + utils::Vector& Cases() { return cases_; } + + /// @returns the condition + const Value* Condition() const { return condition_; } + + private: + Branch merge_ = {}; + utils::Vector cases_; + Value* condition_; }; } // namespace tint::ir diff --git a/src/tint/ir/test_helper.h b/src/tint/ir/test_helper.h index d9055cee86..b7279f3147 100644 --- a/src/tint/ir/test_helper.h +++ b/src/tint/ir/test_helper.h @@ -33,7 +33,6 @@ template class TestHelperBase : public BASE, public ProgramBuilder { public: TestHelperBase() = default; - ~TestHelperBase() override = default; /// Build the module, cleaning up the program before returning. diff --git a/src/tint/ir/to_program.cc b/src/tint/ir/to_program.cc index b1aa67fd62..9be189c216 100644 --- a/src/tint/ir/to_program.cc +++ b/src/tint/ir/to_program.cc @@ -91,14 +91,14 @@ class State { const ast::Function* Fn(const Function* fn) { SCOPED_NESTING(); - auto name = Sym(fn->name); + auto name = Sym(fn->Name()); // TODO(crbug.com/tint/1915): Properly implement this when we've fleshed out Function utils::Vector params{}; - auto ret_ty = Type(fn->return_type); + auto ret_ty = Type(fn->ReturnType()); if (!ret_ty) { return nullptr; } - auto* body = FlowNodeGraph(fn->start_target); + auto* body = FlowNodeGraph(fn->StartTarget()); if (!body) { return nullptr; } @@ -126,7 +126,7 @@ class State { branch->target, [&](const ir::Block* block) { - for (auto* inst : block->instructions) { + for (const auto* inst : block->Instructions()) { auto stmt = Stmt(inst); if (TINT_UNLIKELY(!stmt)) { return kError; @@ -135,7 +135,7 @@ class State { stmts.Push(s); } } - branch = &block->branch; + branch = &block->Branch(); return kContinue; }, @@ -145,8 +145,8 @@ class State { return kError; } stmts.Push(stmt); - branch = &if_->merge; - return branch->target->inbound_branches.IsEmpty() ? kStop : kContinue; + branch = &if_->Merge(); + return branch->target->InboundBranches().IsEmpty() ? kStop : kContinue; }, [&](const ir::Switch* switch_) { @@ -155,8 +155,8 @@ class State { return kError; } stmts.Push(stmt); - branch = &switch_->merge; - return branch->target->inbound_branches.IsEmpty() ? kStop : kContinue; + branch = &switch_->Merge(); + return branch->target->InboundBranches().IsEmpty() ? kStop : kContinue; }, [&](const ir::FunctionTerminator*) { @@ -189,25 +189,25 @@ class State { const ast::IfStatement* If(const ir::If* i) { SCOPED_NESTING(); - auto* cond = Expr(i->condition); - auto* t = FlowNodeGraph(i->true_.target, i->merge.target); + auto* cond = Expr(i->Condition()); + auto* t = FlowNodeGraph(i->True().target, i->Merge().target); if (TINT_UNLIKELY(!t)) { return nullptr; } - if (!IsEmpty(i->false_.target, i->merge.target)) { - // If the else target is an if flow node with the same merge target as this if, then + if (!IsEmpty(i->False().target, i->Merge().target)) { + // If the else target is an if flow node with the same Merge().target as this if, then // emit an 'else if' instead of a block statement for the else. - if (auto* else_if = As(NextNonEmptyNode(i->false_.target)); + if (auto* else_if = As(NextNonEmptyNode(i->False().target)); else_if && - NextNonEmptyNode(i->merge.target) == NextNonEmptyNode(else_if->merge.target)) { + NextNonEmptyNode(i->Merge().target) == NextNonEmptyNode(else_if->Merge().target)) { auto* f = If(else_if); if (!f) { return nullptr; } return b.If(cond, t, b.Else(f)); } else { - auto* f = FlowNodeGraph(i->false_.target, i->merge.target); + auto* f = FlowNodeGraph(i->False().target, i->Merge().target); if (!f) { return nullptr; } @@ -221,16 +221,16 @@ class State { const ast::SwitchStatement* Switch(const ir::Switch* s) { SCOPED_NESTING(); - auto* cond = Expr(s->condition); + auto* cond = Expr(s->Condition()); if (!cond) { return nullptr; } - auto cases = utils::Transform( - s->cases, // + auto cases = utils::Transform<1>( + s->Cases(), // [&](const ir::Switch::Case& c) -> const tint::ast::CaseStatement* { SCOPED_NESTING(); - auto* body = FlowNodeGraph(c.start.target, s->merge.target); + auto* body = FlowNodeGraph(c.start.target, s->Merge().target); if (!body) { return nullptr; } @@ -292,10 +292,10 @@ class State { bool IsEmpty(const ir::FlowNode* node, const ir::FlowNode* stop_at) { while (node != stop_at) { if (auto* block = node->As()) { - if (block->instructions.Length() > 0) { + if (!block->Instructions().IsEmpty()) { return false; } - node = block->branch.target; + node = block->Branch().target; } else { return false; } @@ -307,13 +307,13 @@ class State { const ir::FlowNode* NextNonEmptyNode(const ir::FlowNode* node) { while (node) { if (auto* block = node->As()) { - for (auto* inst : block->instructions) { + for (const auto* inst : block->Instructions()) { // Load instructions will be inlined, so ignore them. if (!inst->Is()) { return node; } } - node = block->branch.target; + node = block->Branch().target; } else { return node; } @@ -351,8 +351,8 @@ class State { } auto ty = Type(ptr->StoreType()); const ast::Expression* init = nullptr; - if (var->initializer) { - init = Expr(var->initializer); + if (var->Initializer()) { + init = Expr(var->Initializer()); if (!init) { return nullptr; } @@ -368,18 +368,19 @@ class State { } const ast::AssignmentStatement* Store(const ir::Store* store) { - auto* expr = Expr(store->from); - return b.Assign(NameOf(store->to), expr); + auto* expr = Expr(store->From()); + return b.Assign(NameOf(store->To()), expr); } const ast::CallExpression* Call(const ir::Call* call) { - auto args = utils::Transform(call->args, [&](const ir::Value* arg) { return Expr(arg); }); + auto args = + utils::Transform<2>(call->Args(), [&](const ir::Value* arg) { return Expr(arg); }); if (args.Any(utils::IsNull)) { return nullptr; } return tint::Switch( call, // - [&](const ir::UserCall* c) { return b.Call(Sym(c->name), std::move(args)); }, + [&](const ir::UserCall* c) { return b.Call(Sym(c->Name()), std::move(args)); }, [&](Default) { UNHANDLED_CASE(call); return nullptr; @@ -401,18 +402,18 @@ class State { const ast::Expression* ConstExpr(const ir::Constant* c) { return tint::Switch( c->Type(), // - [&](const type::I32*) { return b.Expr(c->value->ValueAs()); }, - [&](const type::U32*) { return b.Expr(c->value->ValueAs()); }, - [&](const type::F32*) { return b.Expr(c->value->ValueAs()); }, - [&](const type::F16*) { return b.Expr(c->value->ValueAs()); }, - [&](const type::Bool*) { return b.Expr(c->value->ValueAs()); }, + [&](const type::I32*) { return b.Expr(c->Value()->ValueAs()); }, + [&](const type::U32*) { return b.Expr(c->Value()->ValueAs()); }, + [&](const type::F32*) { return b.Expr(c->Value()->ValueAs()); }, + [&](const type::F16*) { return b.Expr(c->Value()->ValueAs()); }, + [&](const type::Bool*) { return b.Expr(c->Value()->ValueAs()); }, [&](Default) { UNHANDLED_CASE(c); return nullptr; }); } - const ast::Expression* LoadExpr(const ir::Load* l) { return Expr(l->from); } + const ast::Expression* LoadExpr(const ir::Load* l) { return Expr(l->From()); } const ast::Expression* VarExpr(const ir::Var* v) { return b.Expr(NameOf(v)); } diff --git a/src/tint/ir/transform/add_empty_entry_point.cc b/src/tint/ir/transform/add_empty_entry_point.cc index 809a6ad773..6788d7c375 100644 --- a/src/tint/ir/transform/add_empty_entry_point.cc +++ b/src/tint/ir/transform/add_empty_entry_point.cc @@ -29,7 +29,7 @@ AddEmptyEntryPoint::~AddEmptyEntryPoint() = default; void AddEmptyEntryPoint::Run(ir::Module* ir, const DataMap&, DataMap&) const { for (auto* func : ir->functions) { - if (func->pipeline_stage != Function::PipelineStage::kUndefined) { + if (func->Stage() != Function::PipelineStage::kUndefined) { return; } } @@ -38,7 +38,7 @@ void AddEmptyEntryPoint::Run(ir::Module* ir, const DataMap&, DataMap&) const { auto* ep = builder.CreateFunction(ir->symbols.New("unused_entry_point"), ir->types.Get(), Function::PipelineStage::kCompute, std::array{1u, 1u, 1u}); - builder.Branch(ep->start_target, ep->end_target); + ep->StartTarget()->BranchTo(ep->EndTarget()); ir->functions.Push(ep); } diff --git a/src/tint/ir/transform/add_empty_entry_point_test.cc b/src/tint/ir/transform/add_empty_entry_point_test.cc index baba8e0901..a363d83640 100644 --- a/src/tint/ir/transform/add_empty_entry_point_test.cc +++ b/src/tint/ir/transform/add_empty_entry_point_test.cc @@ -40,7 +40,7 @@ TEST_F(IR_AddEmptyEntryPointTest, EmptyModule) { TEST_F(IR_AddEmptyEntryPointTest, ExistingEntryPoint) { auto* ep = b.CreateFunction(mod.symbols.New("main"), mod.types.Get(), Function::PipelineStage::kFragment); - b.Branch(ep->start_target, ep->end_target); + ep->StartTarget()->BranchTo(ep->EndTarget()); mod.functions.Push(ep); auto* expect = R"( diff --git a/src/tint/ir/unary.cc b/src/tint/ir/unary.cc index db13d5d579..2a0ee944e8 100644 --- a/src/tint/ir/unary.cc +++ b/src/tint/ir/unary.cc @@ -19,8 +19,8 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Unary); namespace tint::ir { -Unary::Unary(Kind k, const type::Type* res_ty, Value* val) - : kind(k), result_type(res_ty), val_(val) { +Unary::Unary(enum Kind k, const type::Type* res_ty, Value* val) + : kind_(k), result_type_(res_ty), val_(val) { TINT_ASSERT(IR, val_); val_->AddUsage(this); } diff --git a/src/tint/ir/unary.h b/src/tint/ir/unary.h index 46edd450e9..98ad9b8b2c 100644 --- a/src/tint/ir/unary.h +++ b/src/tint/ir/unary.h @@ -33,7 +33,7 @@ class Unary : public utils::Castable { /// @param kind the kind of unary instruction /// @param result_type the result type /// @param val the input value for the instruction - Unary(Kind kind, const type::Type* result_type, Value* val); + Unary(enum Kind kind, const type::Type* result_type, Value* val); Unary(const Unary& inst) = delete; Unary(Unary&& inst) = delete; ~Unary() override; @@ -42,19 +42,18 @@ class Unary : public utils::Castable { Unary& operator=(Unary&& inst) = delete; /// @returns the type of the value - const type::Type* Type() const override { return result_type; } + const type::Type* Type() const override { return result_type_; } /// @returns the value for the instruction const Value* Val() const { return val_; } - /// the kind of unary instruction - Kind kind = Kind::kNegation; - - /// the result type of the instruction - const type::Type* result_type = nullptr; + /// @returns the kind of unary instruction + enum Kind Kind() const { return kind_; } private: - Value* val_ = nullptr; + enum Kind kind_; + const type::Type* result_type_; + Value* val_; }; } // namespace tint::ir diff --git a/src/tint/ir/unary_test.cc b/src/tint/ir/unary_test.cc index 280c7de831..6cefed38a1 100644 --- a/src/tint/ir/unary_test.cc +++ b/src/tint/ir/unary_test.cc @@ -29,10 +29,10 @@ TEST_F(IR_InstructionTest, CreateComplement) { const auto* inst = b.Complement(b.ir.types.Get(), b.Constant(4_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Unary::Kind::kComplement); + EXPECT_EQ(inst->Kind(), Unary::Kind::kComplement); ASSERT_TRUE(inst->Val()->Is()); - auto lhs = inst->Val()->As()->value; + auto lhs = inst->Val()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); } @@ -43,10 +43,10 @@ TEST_F(IR_InstructionTest, CreateNegation) { const auto* inst = b.Negation(b.ir.types.Get(), b.Constant(4_i)); ASSERT_TRUE(inst->Is()); - EXPECT_EQ(inst->kind, Unary::Kind::kNegation); + EXPECT_EQ(inst->Kind(), Unary::Kind::kNegation); ASSERT_TRUE(inst->Val()->Is()); - auto lhs = inst->Val()->As()->value; + auto lhs = inst->Val()->As()->Value(); ASSERT_TRUE(lhs->Is>()); EXPECT_EQ(4_i, lhs->As>()->ValueAs()); } @@ -56,7 +56,7 @@ TEST_F(IR_InstructionTest, Unary_Usage) { Builder b{mod}; const auto* inst = b.Negation(b.ir.types.Get(), b.Constant(4_i)); - EXPECT_EQ(inst->kind, Unary::Kind::kNegation); + EXPECT_EQ(inst->Kind(), Unary::Kind::kNegation); ASSERT_NE(inst->Val(), nullptr); ASSERT_EQ(inst->Val()->Usage().Length(), 1u); diff --git a/src/tint/ir/user_call.cc b/src/tint/ir/user_call.cc index f718284ee5..e44e28574b 100644 --- a/src/tint/ir/user_call.cc +++ b/src/tint/ir/user_call.cc @@ -23,7 +23,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::UserCall); namespace tint::ir { UserCall::UserCall(const type::Type* ty, Symbol n, utils::VectorRef arguments) - : Base(ty, std::move(arguments)), name(n) {} + : Base(ty, std::move(arguments)), name_(n) {} UserCall::~UserCall() = default; diff --git a/src/tint/ir/user_call.h b/src/tint/ir/user_call.h index 5ada8f8ddf..ba52e20d9b 100644 --- a/src/tint/ir/user_call.h +++ b/src/tint/ir/user_call.h @@ -36,8 +36,11 @@ class UserCall : public utils::Castable { UserCall& operator=(const UserCall& inst) = delete; UserCall& operator=(UserCall&& inst) = delete; - /// The function name - Symbol name; + /// @returns the called function name + Symbol Name() const { return name_; } + + private: + Symbol name_; }; } // namespace tint::ir diff --git a/src/tint/ir/var.cc b/src/tint/ir/var.cc index 9bf4b8da12..9b43329cd5 100644 --- a/src/tint/ir/var.cc +++ b/src/tint/ir/var.cc @@ -19,8 +19,14 @@ TINT_INSTANTIATE_TYPEINFO(tint::ir::Var); namespace tint::ir { -Var::Var(const type::Type* ty) : type(ty) {} +Var::Var(const type::Type* ty) : type_(ty) {} Var::~Var() = default; +void Var::SetInitializer(Value* initializer) { + initializer_ = initializer; + initializer_->AddUsage(this); + // TODO(dsinclair): Probably should do a RemoveUsage on an existing initializer if set +} + } // namespace tint::ir diff --git a/src/tint/ir/var.h b/src/tint/ir/var.h index c874a6235f..67ccf0d3d8 100644 --- a/src/tint/ir/var.h +++ b/src/tint/ir/var.h @@ -36,13 +36,17 @@ class Var : public utils::Castable { Var& operator=(Var&& inst) = delete; /// @returns the type of the var - const type::Type* Type() const override { return type; } + const type::Type* Type() const override { return type_; } - /// the result type of the instruction - const type::Type* type = nullptr; + /// Sets the var initializer + /// @param initializer the initializer + void SetInitializer(Value* initializer); + /// @returns the initializer + const Value* Initializer() const { return initializer_; } - /// The optional initializer - Value* initializer = nullptr; + private: + const type::Type* type_; + Value* initializer_ = nullptr; }; } // namespace tint::ir diff --git a/src/tint/transform/manager_test.cc b/src/tint/transform/manager_test.cc index 3b0ee25f19..a81f7bf07c 100644 --- a/src/tint/transform/manager_test.cc +++ b/src/tint/transform/manager_test.cc @@ -52,7 +52,7 @@ class IR_AddFunction final : public ir::transform::Transform { ir::Builder builder(*mod); auto* func = builder.CreateFunction(mod->symbols.New("ir_func"), mod->types.Get()); - builder.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); mod->functions.Push(func); } }; @@ -70,7 +70,7 @@ ir::Module MakeIR() { ir::Builder builder(mod); auto* func = builder.CreateFunction(builder.ir.symbols.New("main"), builder.ir.types.Get()); - builder.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); builder.ir.functions.Push(func); return mod; } @@ -102,9 +102,10 @@ TEST_F(TransformManagerTest, IR_MutateInPlace) { manager.Add(); manager.Run(&ir, {}, outputs); + ASSERT_EQ(ir.functions.Length(), 2u); - EXPECT_EQ(ir.functions[0]->name.Name(), "main"); - EXPECT_EQ(ir.functions[1]->name.Name(), "ir_func"); + EXPECT_EQ(ir.functions[0]->Name().Name(), "main"); + EXPECT_EQ(ir.functions[1]->Name().Name(), "ir_func"); } TEST_F(TransformManagerTest, AST_MixedTransforms_AST_Before_IR) { @@ -149,9 +150,9 @@ TEST_F(TransformManagerTest, IR_MixedTransforms_AST_Before_IR) { manager.Run(&ir, {}, outputs); ASSERT_EQ(ir.functions.Length(), 3u); - EXPECT_EQ(ir.functions[0]->name.Name(), "ast_func"); - EXPECT_EQ(ir.functions[1]->name.Name(), "main"); - EXPECT_EQ(ir.functions[2]->name.Name(), "ir_func"); + EXPECT_EQ(ir.functions[0]->Name().Name(), "ast_func"); + EXPECT_EQ(ir.functions[1]->Name().Name(), "main"); + EXPECT_EQ(ir.functions[2]->Name().Name(), "ir_func"); } TEST_F(TransformManagerTest, IR_MixedTransforms_IR_Before_AST) { @@ -164,9 +165,9 @@ TEST_F(TransformManagerTest, IR_MixedTransforms_IR_Before_AST) { manager.Run(&ir, {}, outputs); ASSERT_EQ(ir.functions.Length(), 3u); - EXPECT_EQ(ir.functions[0]->name.Name(), "ast_func"); - EXPECT_EQ(ir.functions[1]->name.Name(), "main"); - EXPECT_EQ(ir.functions[2]->name.Name(), "ir_func"); + EXPECT_EQ(ir.functions[0]->Name().Name(), "ast_func"); + EXPECT_EQ(ir.functions[1]->Name().Name(), "main"); + EXPECT_EQ(ir.functions[2]->Name().Name(), "ir_func"); } #endif // TINT_BUILD_IR diff --git a/src/tint/writer/spirv/ir/generator_impl_ir.cc b/src/tint/writer/spirv/ir/generator_impl_ir.cc index 059e5273f3..bd9a735bc5 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir.cc @@ -110,7 +110,7 @@ bool GeneratorImplIr::Generate() { } uint32_t GeneratorImplIr::Constant(const ir::Constant* constant) { - return Constant(constant->value); + return Constant(constant->Value()); } uint32_t GeneratorImplIr::Constant(const constant::Value* constant) { @@ -214,15 +214,15 @@ void GeneratorImplIr::EmitFunction(const ir::Function* func) { auto id = module_.NextId(); // Emit the function name. - module_.PushDebug(spv::Op::OpName, {id, Operand(func->name.Name())}); + module_.PushDebug(spv::Op::OpName, {id, Operand(func->Name().Name())}); // Emit OpEntryPoint and OpExecutionMode declarations if needed. - if (func->pipeline_stage != ir::Function::PipelineStage::kUndefined) { + if (func->Stage() != ir::Function::PipelineStage::kUndefined) { EmitEntryPoint(func, id); } // Get the ID for the return type. - auto return_type_id = Type(func->return_type); + auto return_type_id = Type(func->ReturnType()); // Get the ID for the function type (creating it if needed). // TODO(jrprice): Add the parameter types when they are supported in the IR. @@ -248,7 +248,7 @@ void GeneratorImplIr::EmitFunction(const ir::Function* func) { TINT_DEFER(current_function_ = Function()); // Emit the body of the function. - EmitBlock(func->start_target); + EmitBlock(func->StartTarget()); // Add the function to the module. module_.PushFunction(current_function_); @@ -256,13 +256,13 @@ void GeneratorImplIr::EmitFunction(const ir::Function* func) { void GeneratorImplIr::EmitEntryPoint(const ir::Function* func, uint32_t id) { SpvExecutionModel stage = SpvExecutionModelMax; - switch (func->pipeline_stage) { + switch (func->Stage()) { case ir::Function::PipelineStage::kCompute: { stage = SpvExecutionModelGLCompute; module_.PushExecutionMode( spv::Op::OpExecutionMode, - {id, U32Operand(SpvExecutionModeLocalSize), func->workgroup_size->at(0), - func->workgroup_size->at(1), func->workgroup_size->at(2)}); + {id, U32Operand(SpvExecutionModeLocalSize), func->WorkgroupSize()->at(0), + func->WorkgroupSize()->at(1), func->WorkgroupSize()->at(2)}); break; } case ir::Function::PipelineStage::kFragment: { @@ -282,7 +282,7 @@ void GeneratorImplIr::EmitEntryPoint(const ir::Function* func, uint32_t id) { } // TODO(jrprice): Add the interface list of all referenced global variables. - module_.PushEntryPoint(spv::Op::OpEntryPoint, {U32Operand(stage), id, func->name.Name()}); + module_.PushEntryPoint(spv::Op::OpEntryPoint, {U32Operand(stage), id, func->Name().Name()}); } void GeneratorImplIr::EmitBlock(const ir::Block* block) { @@ -293,7 +293,7 @@ void GeneratorImplIr::EmitBlock(const ir::Block* block) { } // Emit the instructions. - for (auto* inst : block->instructions) { + for (const auto* inst : block->Instructions()) { auto result = Switch( inst, // [&](const ir::Binary* b) { return EmitBinary(b); }, @@ -313,43 +313,43 @@ void GeneratorImplIr::EmitBlock(const ir::Block* block) { // Handle the branch at the end of the block. Switch( - block->branch.target, + block->Branch().target, [&](const ir::Block* b) { current_function_.push_inst(spv::Op::OpBranch, {Label(b)}); }, [&](const ir::If* i) { EmitIf(i); }, [&](const ir::FunctionTerminator*) { // TODO(jrprice): Handle the return value, which will be a branch argument. - if (!block->branch.args.IsEmpty()) { + if (!block->Branch().args.IsEmpty()) { TINT_ICE(Writer, diagnostics_) << "unimplemented return value"; } current_function_.push_inst(spv::Op::OpReturn, {}); }, [&](Default) { - if (!block->branch.target) { + if (!block->Branch().target) { // A block may not have an outward branch (e.g. an unreachable merge block). current_function_.push_inst(spv::Op::OpUnreachable, {}); } else { TINT_ICE(Writer, diagnostics_) - << "unimplemented branch target: " << block->branch.target->TypeInfo().name; + << "unimplemented branch target: " << block->Branch().target->TypeInfo().name; } }); } void GeneratorImplIr::EmitIf(const ir::If* i) { - auto* merge_block = i->merge.target->As(); - auto* true_block = i->true_.target->As(); - auto* false_block = i->false_.target->As(); + auto* merge_block = i->Merge().target->As(); + auto* true_block = i->True().target->As(); + auto* false_block = i->False().target->As(); // Generate labels for the blocks. We emit the true or false block if it: // 1. contains instructions, or - // 2. branches somewhere other then the merge target. + // 2. branches somewhere other then the Merge().target. // Otherwise we skip them and branch straight to the merge block. uint32_t merge_label = Label(merge_block); uint32_t true_label = merge_label; uint32_t false_label = merge_label; - if (!true_block->instructions.IsEmpty() || true_block->branch.target != merge_block) { + if (!true_block->Instructions().IsEmpty() || true_block->Branch().target != merge_block) { true_label = Label(true_block); } - if (!false_block->instructions.IsEmpty() || false_block->branch.target != merge_block) { + if (!false_block->Instructions().IsEmpty() || false_block->Branch().target != merge_block) { false_label = Label(false_block); } @@ -357,7 +357,7 @@ void GeneratorImplIr::EmitIf(const ir::If* i) { current_function_.push_inst(spv::Op::OpSelectionMerge, {merge_label, U32Operand(SpvSelectionControlMaskNone)}); current_function_.push_inst(spv::Op::OpBranchConditional, - {Value(i->condition), true_label, false_label}); + {Value(i->Condition()), true_label, false_label}); // Emit the `true` and `false` blocks, if they're not being skipped. if (true_label != merge_label) { @@ -376,7 +376,7 @@ uint32_t GeneratorImplIr::EmitBinary(const ir::Binary* binary) { // Determine the opcode. spv::Op op = spv::Op::Max; - switch (binary->kind) { + switch (binary->Kind()) { case ir::Binary::Kind::kAdd: { op = binary->Type()->is_integer_scalar_or_vector() ? spv::Op::OpIAdd : spv::Op::OpFAdd; break; @@ -387,7 +387,7 @@ uint32_t GeneratorImplIr::EmitBinary(const ir::Binary* binary) { } default: { TINT_ICE(Writer, diagnostics_) - << "unimplemented binary instruction: " << static_cast(binary->kind); + << "unimplemented binary instruction: " << static_cast(binary->Kind()); } } @@ -400,12 +400,12 @@ uint32_t GeneratorImplIr::EmitBinary(const ir::Binary* binary) { uint32_t GeneratorImplIr::EmitLoad(const ir::Load* load) { auto id = module_.NextId(); - current_function_.push_inst(spv::Op::OpLoad, {Type(load->Type()), id, Value(load->from)}); + current_function_.push_inst(spv::Op::OpLoad, {Type(load->Type()), id, Value(load->From())}); return id; } void GeneratorImplIr::EmitStore(const ir::Store* store) { - current_function_.push_inst(spv::Op::OpStore, {Value(store->to), Value(store->from)}); + current_function_.push_inst(spv::Op::OpStore, {Value(store->To()), Value(store->From())}); } uint32_t GeneratorImplIr::EmitVar(const ir::Var* var) { @@ -417,8 +417,8 @@ uint32_t GeneratorImplIr::EmitVar(const ir::Var* var) { if (ptr->AddressSpace() == builtin::AddressSpace::kFunction) { TINT_ASSERT(Writer, current_function_); current_function_.push_var({ty, id, U32Operand(SpvStorageClassFunction)}); - if (var->initializer) { - current_function_.push_inst(spv::Op::OpStore, {id, Value(var->initializer)}); + if (var->Initializer()) { + current_function_.push_inst(spv::Op::OpStore, {id, Value(var->Initializer())}); } } else { TINT_ICE(Writer, diagnostics_) diff --git a/src/tint/writer/spirv/ir/generator_impl_ir_binary_test.cc b/src/tint/writer/spirv/ir/generator_impl_ir_binary_test.cc index b893496c0e..e9231f32b4 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir_binary_test.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir_binary_test.cc @@ -21,10 +21,10 @@ namespace { TEST_F(SpvGeneratorImplTest, Binary_Add_I32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(2_i))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(2_i))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -43,10 +43,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Add_U32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Add(mod.types.Get(), b.Constant(1_u), b.Constant(2_u))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Add(mod.types.Get(), b.Constant(1_u), b.Constant(2_u))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -65,10 +65,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Add_F32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Add(mod.types.Get(), b.Constant(1_f), b.Constant(2_f))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Add(mod.types.Get(), b.Constant(1_f), b.Constant(2_f))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -87,10 +87,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Sub_I32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Subtract(mod.types.Get(), b.Constant(1_i), b.Constant(2_i))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Subtract(mod.types.Get(), b.Constant(1_i), b.Constant(2_i))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -109,10 +109,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Sub_U32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Subtract(mod.types.Get(), b.Constant(1_u), b.Constant(2_u))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Subtract(mod.types.Get(), b.Constant(1_u), b.Constant(2_u))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -131,10 +131,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Sub_F32) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); - func->start_target->instructions.Push( - b.Subtract(mod.types.Get(), b.Constant(1_f), b.Constant(2_f))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Subtract(mod.types.Get(), b.Constant(1_f), b.Constant(2_f))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -153,17 +153,17 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Sub_Vec2i) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* lhs = mod.constants_arena.Create( mod.types.Get(mod.types.Get(), 2u), - utils::Vector{b.Constant(42_i)->value, b.Constant(-1_i)->value}, false, false); + utils::Vector{b.Constant(42_i)->Value(), b.Constant(-1_i)->Value()}, false, false); auto* rhs = mod.constants_arena.Create( mod.types.Get(mod.types.Get(), 2u), - utils::Vector{b.Constant(0_i)->value, b.Constant(-43_i)->value}, false, false); - func->start_target->instructions.Push( - b.Subtract(mod.types.Get(mod.types.Get(), 2u), b.Constant(lhs), - b.Constant(rhs))); + utils::Vector{b.Constant(0_i)->Value(), b.Constant(-43_i)->Value()}, false, false); + func->StartTarget()->SetInstructions( + utils::Vector{b.Subtract(mod.types.Get(mod.types.Get(), 2u), + b.Constant(lhs), b.Constant(rhs))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -187,21 +187,21 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Sub_Vec4f) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* lhs = mod.constants_arena.Create( mod.types.Get(mod.types.Get(), 4u), - utils::Vector{b.Constant(42_f)->value, b.Constant(-1_f)->value, b.Constant(0_f)->value, - b.Constant(1.25_f)->value}, + utils::Vector{b.Constant(42_f)->Value(), b.Constant(-1_f)->Value(), + b.Constant(0_f)->Value(), b.Constant(1.25_f)->Value()}, false, false); auto* rhs = mod.constants_arena.Create( mod.types.Get(mod.types.Get(), 4u), - utils::Vector{b.Constant(0_f)->value, b.Constant(1.25_f)->value, b.Constant(-42_f)->value, - b.Constant(1_f)->value}, + utils::Vector{b.Constant(0_f)->Value(), b.Constant(1.25_f)->Value(), + b.Constant(-42_f)->Value(), b.Constant(1_f)->Value()}, false, false); - func->start_target->instructions.Push( - b.Subtract(mod.types.Get(mod.types.Get(), 4u), b.Constant(lhs), - b.Constant(rhs))); + func->StartTarget()->SetInstructions( + utils::Vector{b.Subtract(mod.types.Get(mod.types.Get(), 4u), + b.Constant(lhs), b.Constant(rhs))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -227,11 +227,10 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Binary_Chain) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* a = b.Subtract(mod.types.Get(), b.Constant(1_i), b.Constant(2_i)); - func->start_target->instructions.Push(a); - func->start_target->instructions.Push(b.Add(mod.types.Get(), a, a)); + func->StartTarget()->SetInstructions(utils::Vector{a, b.Add(mod.types.Get(), a, a)}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" diff --git a/src/tint/writer/spirv/ir/generator_impl_ir_constant_test.cc b/src/tint/writer/spirv/ir/generator_impl_ir_constant_test.cc index c58859a863..6ab48aa2a2 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir_constant_test.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir_constant_test.cc @@ -67,7 +67,7 @@ TEST_F(SpvGeneratorImplTest, Constant_Vec4Bool) { auto* f = b.Constant(false); auto* v = mod.constants_arena.Create( mod.types.Get(mod.types.Get(), 4u), - utils::Vector{t->value, f->value, f->value, t->value}, false, true); + utils::Vector{t->Value(), f->Value(), f->Value(), t->Value()}, false, true); generator_.Constant(b.Constant(v)); EXPECT_EQ(DumpTypes(), R"(%3 = OpTypeBool %2 = OpTypeVector %3 4 @@ -82,7 +82,8 @@ TEST_F(SpvGeneratorImplTest, Constant_Vec2i) { auto* i_42 = b.Constant(i32(42)); auto* i_n1 = b.Constant(i32(-1)); auto* v = mod.constants_arena.Create( - mod.types.Get(i, 2u), utils::Vector{i_42->value, i_n1->value}, false, false); + mod.types.Get(i, 2u), utils::Vector{i_42->Value(), i_n1->Value()}, false, + false); generator_.Constant(b.Constant(v)); EXPECT_EQ(DumpTypes(), R"(%3 = OpTypeInt 32 1 %2 = OpTypeVector %3 2 @@ -98,8 +99,8 @@ TEST_F(SpvGeneratorImplTest, Constant_Vec3u) { auto* u_0 = b.Constant(u32(0)); auto* u_4b = b.Constant(u32(4000000000)); auto* v = mod.constants_arena.Create( - mod.types.Get(u, 3u), utils::Vector{u_42->value, u_0->value, u_4b->value}, - false, true); + mod.types.Get(u, 3u), + utils::Vector{u_42->Value(), u_0->Value(), u_4b->Value()}, false, true); generator_.Constant(b.Constant(v)); EXPECT_EQ(DumpTypes(), R"(%3 = OpTypeInt 32 0 %2 = OpTypeVector %3 3 @@ -118,7 +119,7 @@ TEST_F(SpvGeneratorImplTest, Constant_Vec4f) { auto* f_n1 = b.Constant(f32(-1)); auto* v = mod.constants_arena.Create( mod.types.Get(f, 4u), - utils::Vector{f_42->value, f_0->value, f_q->value, f_n1->value}, false, true); + utils::Vector{f_42->Value(), f_0->Value(), f_q->Value(), f_n1->Value()}, false, true); generator_.Constant(b.Constant(v)); EXPECT_EQ(DumpTypes(), R"(%3 = OpTypeFloat 32 %2 = OpTypeVector %3 4 @@ -135,7 +136,8 @@ TEST_F(SpvGeneratorImplTest, Constant_Vec2h) { auto* h_42 = b.Constant(f16(42)); auto* h_q = b.Constant(f16(0.25)); auto* v = mod.constants_arena.Create( - mod.types.Get(h, 2u), utils::Vector{h_42->value, h_q->value}, false, false); + mod.types.Get(h, 2u), utils::Vector{h_42->Value(), h_q->Value()}, false, + false); generator_.Constant(b.Constant(v)); EXPECT_EQ(DumpTypes(), R"(%3 = OpTypeFloat 16 %2 = OpTypeVector %3 2 diff --git a/src/tint/writer/spirv/ir/generator_impl_ir_function_test.cc b/src/tint/writer/spirv/ir/generator_impl_ir_function_test.cc index 77b4a62a57..d2af246a79 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir_function_test.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir_function_test.cc @@ -19,7 +19,7 @@ namespace { TEST_F(SpvGeneratorImplTest, Function_Empty) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -35,7 +35,7 @@ OpFunctionEnd // Test that we do not emit the same function type more than once. TEST_F(SpvGeneratorImplTest, Function_DeduplicateType) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); generator_.EmitFunction(func); generator_.EmitFunction(func); @@ -48,7 +48,7 @@ TEST_F(SpvGeneratorImplTest, Function_DeduplicateType) { TEST_F(SpvGeneratorImplTest, Function_EntryPoint_Compute) { auto* func = b.CreateFunction(mod.symbols.Register("main"), mod.types.Get(), ir::Function::PipelineStage::kCompute, {{32, 4, 1}}); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpEntryPoint GLCompute %1 "main" @@ -66,7 +66,7 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Function_EntryPoint_Fragment) { auto* func = b.CreateFunction(mod.symbols.Register("main"), mod.types.Get(), ir::Function::PipelineStage::kFragment); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpEntryPoint Fragment %1 "main" @@ -84,7 +84,7 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Function_EntryPoint_Vertex) { auto* func = b.CreateFunction(mod.symbols.Register("main"), mod.types.Get(), ir::Function::PipelineStage::kVertex); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpEntryPoint Vertex %1 "main" @@ -101,15 +101,15 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, Function_EntryPoint_Multiple) { auto* f1 = b.CreateFunction(mod.symbols.Register("main1"), mod.types.Get(), ir::Function::PipelineStage::kCompute, {{32, 4, 1}}); - b.Branch(f1->start_target, f1->end_target); + f1->StartTarget()->BranchTo(f1->EndTarget()); auto* f2 = b.CreateFunction(mod.symbols.Register("main2"), mod.types.Get(), ir::Function::PipelineStage::kCompute, {{8, 2, 16}}); - b.Branch(f2->start_target, f2->end_target); + f2->StartTarget()->BranchTo(f2->EndTarget()); auto* f3 = b.CreateFunction(mod.symbols.Register("main3"), mod.types.Get(), ir::Function::PipelineStage::kFragment); - b.Branch(f3->start_target, f3->end_target); + f3->StartTarget()->BranchTo(f3->EndTarget()); generator_.EmitFunction(f1); generator_.EmitFunction(f2); diff --git a/src/tint/writer/spirv/ir/generator_impl_ir_if_test.cc b/src/tint/writer/spirv/ir/generator_impl_ir_if_test.cc index 7b41184ac3..fa8cc6bc11 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir_if_test.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir_if_test.cc @@ -23,11 +23,11 @@ TEST_F(SpvGeneratorImplTest, If_TrueEmpty_FalseEmpty) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); auto* i = b.CreateIf(b.Constant(true)); - b.Branch(i->true_.target->As(), i->merge.target); - b.Branch(i->false_.target->As(), i->merge.target); - b.Branch(i->merge.target->As(), func->end_target); + i->True().target->As()->BranchTo(i->Merge().target); + i->False().target->As()->BranchTo(i->Merge().target); + i->Merge().target->As()->BranchTo(func->EndTarget()); - b.Branch(func->start_target, i); + func->StartTarget()->BranchTo(i); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -49,15 +49,15 @@ TEST_F(SpvGeneratorImplTest, If_FalseEmpty) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); auto* i = b.CreateIf(b.Constant(true)); - b.Branch(i->false_.target->As(), i->merge.target); - b.Branch(i->merge.target->As(), func->end_target); + i->False().target->As()->BranchTo(i->Merge().target); + i->Merge().target->As()->BranchTo(func->EndTarget()); - auto* true_block = i->true_.target->As(); - true_block->instructions.Push( - b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(1_i))); - b.Branch(true_block, i->merge.target); + auto* true_block = i->True().target->As(); + true_block->SetInstructions( + utils::Vector{b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(1_i))}); + true_block->BranchTo(i->Merge().target); - b.Branch(func->start_target, i); + func->StartTarget()->BranchTo(i); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -84,15 +84,15 @@ TEST_F(SpvGeneratorImplTest, If_TrueEmpty) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); auto* i = b.CreateIf(b.Constant(true)); - b.Branch(i->true_.target->As(), i->merge.target); - b.Branch(i->merge.target->As(), func->end_target); + i->True().target->As()->BranchTo(i->Merge().target); + i->Merge().target->As()->BranchTo(func->EndTarget()); - auto* false_block = i->false_.target->As(); - false_block->instructions.Push( - b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(1_i))); - b.Branch(false_block, i->merge.target); + auto* false_block = i->False().target->As(); + false_block->SetInstructions( + utils::Vector{b.Add(mod.types.Get(), b.Constant(1_i), b.Constant(1_i))}); + false_block->BranchTo(i->Merge().target); - b.Branch(func->start_target, i); + func->StartTarget()->BranchTo(i); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -119,11 +119,11 @@ TEST_F(SpvGeneratorImplTest, If_BothBranchesReturn) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); auto* i = b.CreateIf(b.Constant(true)); - b.Branch(i->true_.target->As(), func->end_target); - b.Branch(i->false_.target->As(), func->end_target); - i->merge.target->As()->branch.target = nullptr; + i->True().target->As()->BranchTo(func->EndTarget()); + i->False().target->As()->BranchTo(func->EndTarget()); + i->Merge().target->As()->BranchTo(nullptr); - b.Branch(func->start_target, i); + func->StartTarget()->BranchTo(i); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" diff --git a/src/tint/writer/spirv/ir/generator_impl_ir_var_test.cc b/src/tint/writer/spirv/ir/generator_impl_ir_var_test.cc index d9bc60c67f..20e5167980 100644 --- a/src/tint/writer/spirv/ir/generator_impl_ir_var_test.cc +++ b/src/tint/writer/spirv/ir/generator_impl_ir_var_test.cc @@ -22,12 +22,11 @@ namespace { TEST_F(SpvGeneratorImplTest, FunctionVar_NoInit) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* ty = mod.types.Get( mod.types.Get(), builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); - auto* v = b.Declare(ty); - func->start_target->instructions.Push(v); + func->StartTarget()->SetInstructions(utils::Vector{b.Declare(ty)}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -45,13 +44,14 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, FunctionVar_WithInit) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* ty = mod.types.Get( mod.types.Get(), builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); auto* v = b.Declare(ty); - func->start_target->instructions.Push(v); - v->initializer = b.Constant(42_i); + v->SetInitializer(b.Constant(42_i)); + + func->StartTarget()->SetInstructions(utils::Vector{v}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -71,12 +71,12 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, FunctionVar_Name) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* ty = mod.types.Get( mod.types.Get(), builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); auto* v = b.Declare(ty); - func->start_target->instructions.Push(v); + func->StartTarget()->SetInstructions(utils::Vector{v}); mod.SetName(v, "myvar"); generator_.EmitFunction(func); @@ -96,22 +96,22 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, FunctionVar_DeclInsideBlock) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* ty = mod.types.Get( mod.types.Get(), builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); auto* v = b.Declare(ty); - v->initializer = b.Constant(42_i); + v->SetInitializer(b.Constant(42_i)); auto* i = b.CreateIf(b.Constant(true)); - b.Branch(i->false_.target->As(), func->end_target); - b.Branch(i->merge.target->As(), func->end_target); + i->False().target->As()->BranchTo(func->EndTarget()); + i->Merge().target->As()->BranchTo(func->EndTarget()); - auto* true_block = i->true_.target->As(); - true_block->instructions.Push(v); - b.Branch(true_block, i->merge.target); + auto* true_block = i->True().target->As(); + true_block->SetInstructions(utils::Vector{v}); + true_block->BranchTo(i->Merge().target); - b.Branch(func->start_target, i); + func->StartTarget()->BranchTo(i); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -140,14 +140,13 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, FunctionVar_Load) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* store_ty = mod.types.Get(); auto* ty = mod.types.Get(store_ty, builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); auto* v = b.Declare(ty); - func->start_target->instructions.Push(v); - func->start_target->instructions.Push(b.Load(v)); + func->StartTarget()->SetInstructions(utils::Vector{v, b.Load(v)}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo" @@ -166,13 +165,12 @@ OpFunctionEnd TEST_F(SpvGeneratorImplTest, FunctionVar_Store) { auto* func = b.CreateFunction(mod.symbols.Register("foo"), mod.types.Get()); - b.Branch(func->start_target, func->end_target); + func->StartTarget()->BranchTo(func->EndTarget()); auto* ty = mod.types.Get( mod.types.Get(), builtin::AddressSpace::kFunction, builtin::Access::kReadWrite); auto* v = b.Declare(ty); - func->start_target->instructions.Push(v); - func->start_target->instructions.Push(b.Store(v, b.Constant(42_i))); + func->StartTarget()->SetInstructions(utils::Vector{v, b.Store(v, b.Constant(42_i))}); generator_.EmitFunction(func); EXPECT_EQ(DumpModule(generator_.Module()), R"(OpName %1 "foo"