diff --git a/src/ast/function.h b/src/ast/function.h index 7e21c28263..25bb059a6d 100644 --- a/src/ast/function.h +++ b/src/ast/function.h @@ -113,6 +113,8 @@ class Function : public Node { /// is not already included. /// @param var the module variable to add void add_referenced_module_variable(Variable* var); + /// Note: If this function calls other functions, the return will also include + /// all of the referenced variables from the callees. /// @returns the referenced module variables const std::vector& referenced_module_variables() const { return referenced_module_vars_; diff --git a/src/inspector.cc b/src/inspector.cc index bdcb4df70f..cfe9d50a39 100644 --- a/src/inspector.cc +++ b/src/inspector.cc @@ -14,22 +14,41 @@ #include "src/inspector.h" +#include + #include "src/ast/function.h" namespace tint { namespace inspector { +EntryPoint::EntryPoint() = default; +EntryPoint::EntryPoint(EntryPoint&) = default; +EntryPoint::EntryPoint(EntryPoint&&) = default; +EntryPoint::~EntryPoint() = default; + Inspector::Inspector(const ast::Module& module) : module_(module) {} Inspector::~Inspector() = default; std::vector Inspector::GetEntryPoints() { std::vector result; + for (const auto& func : module_.functions()) { if (func->IsEntryPoint()) { - uint32_t x, y, z; - std::tie(x, y, z) = func->workgroup_size(); - result.push_back({func->name(), func->pipeline_stage(), x, y, z}); + EntryPoint entry_point; + entry_point.name = func->name(); + entry_point.stage = func->pipeline_stage(); + std::tie(entry_point.workgroup_size_x, entry_point.workgroup_size_y, + entry_point.workgroup_size_z) = func->workgroup_size(); + + for (auto* var : func->referenced_module_variables()) { + if (var->storage_class() == ast::StorageClass::kInput) { + entry_point.input_variables.push_back(var->name()); + } else { + entry_point.output_variables.push_back(var->name()); + } + } + result.push_back(std::move(entry_point)); } } diff --git a/src/inspector.h b/src/inspector.h index 6631765057..75f9ead453 100644 --- a/src/inspector.h +++ b/src/inspector.h @@ -28,6 +28,12 @@ namespace inspector { /// Container of reflection data for an entry point in the shader. struct EntryPoint { + /// Constructors + EntryPoint(); + EntryPoint(EntryPoint&); + EntryPoint(EntryPoint&&); + ~EntryPoint(); + /// The entry point name std::string name; /// The entry point stage @@ -38,6 +44,10 @@ struct EntryPoint { uint32_t workgroup_size_y; /// The workgroup z size uint32_t workgroup_size_z; + /// List of the input variable accessed via this entry point. + std::vector input_variables; + /// List of the output variable accessed via this entry point. + std::vector output_variables; /// @returns the size of the workgroup in {x,y,z} format std::tuple workgroup_size() { diff --git a/src/inspector_test.cc b/src/inspector_test.cc index 7bd3b4f438..d0a8f96f2f 100644 --- a/src/inspector_test.cc +++ b/src/inspector_test.cc @@ -15,12 +15,19 @@ #include "src/inspector.h" #include "gtest/gtest.h" +#include "src/ast/assignment_statement.h" +#include "src/ast/call_expression.h" +#include "src/ast/call_statement.h" #include "src/ast/function.h" +#include "src/ast/identifier_expression.h" #include "src/ast/pipeline_stage.h" +#include "src/ast/return_statement.h" #include "src/ast/stage_decoration.h" +#include "src/ast/type/u32_type.h" #include "src/ast/type/void_type.h" #include "src/ast/workgroup_decoration.h" #include "src/context.h" +#include "src/type_determiner.h" namespace tint { namespace inspector { @@ -29,33 +36,135 @@ namespace { class InspectorHelper { public: InspectorHelper() - : mod_(std::make_unique()), - inspector_(std::make_unique(*mod_)) {} + : td_(std::make_unique(&ctx_, &mod_)), + inspector_(std::make_unique(mod_)) {} - void AddFunction(const std::string& name, ast::PipelineStage stage) { - auto func = std::make_unique( - name, ast::VariableList{}, - ctx_.type_mgr().Get(std::make_unique())); - if (stage != ast::PipelineStage::kNone) { - func->add_decoration(std::make_unique(stage)); + /// Generates an empty function + /// @param name name of the function created + /// @returns a function object + std::unique_ptr GenerateEmptyBodyFunction(std::string name) { + auto body = std::make_unique(); + body->append(std::make_unique()); + std::unique_ptr func = + std::make_unique(name, ast::VariableList(), void_type()); + func->set_body(std::move(body)); + return func; + } + + /// Generates a function that calls another + /// @param caller name of the function created + /// @param callee name of the function to be called + /// @returns a function object + std::unique_ptr GenerateCallerBodyFunction( + std::string caller, + std::string callee) { + auto body = std::make_unique(); + auto ident_expr = std::make_unique(callee); + auto call_expr = std::make_unique( + std::move(ident_expr), ast::ExpressionList()); + body->append(std::make_unique(std::move(call_expr))); + body->append(std::make_unique()); + std::unique_ptr func = std::make_unique( + caller, ast::VariableList(), void_type()); + func->set_body(std::move(body)); + return func; + } + + /// Add In/Out variables to the global variables + /// @param inout_vars tuples of {in, out} that will be added as entries to the + /// global variables + void CreateInOutVariables( + std::vector> inout_vars) { + for (auto inout : inout_vars) { + std::string in, out; + std::tie(in, out) = inout; + auto in_var = std::make_unique( + in, ast::StorageClass::kInput, u32_type()); + auto out_var = std::make_unique( + out, ast::StorageClass::kOutput, u32_type()); + mod()->AddGlobalVariable(std::move(in_var)); + mod()->AddGlobalVariable(std::move(out_var)); } - last_function_ = func.get(); - mod()->AddFunction(std::move(func)); } - void AddWorkGroupSizeToLastFunction(uint32_t x, uint32_t y, uint32_t z) { - last_function_->add_decoration( - std::make_unique(x, y, z)); + /// Generates a function that references in/out variables + /// @param name name of the function created + /// @param inout_vars tuples of {in, out} that will be converted into out = in + /// calls in the function body + /// @returns a function object + std::unique_ptr GenerateInOutVariableBodyFunction( + std::string name, + std::vector> inout_vars) { + auto body = std::make_unique(); + for (auto inout : inout_vars) { + std::string in, out; + std::tie(in, out) = inout; + body->append(std::make_unique( + std::make_unique(out), + std::make_unique(in))); + } + body->append(std::make_unique()); + auto func = + std::make_unique(name, ast::VariableList(), void_type()); + func->set_body(std::move(body)); + return func; } - ast::Module* mod() { return mod_.get(); } + /// Generates a function that references in/out variables and calls another + /// function. + /// @param caller name of the function created + /// @param callee name of the function to be called + /// @param inout_vars tuples of {in, out} that will be converted into out = in + /// calls in the function body + /// @returns a function object + std::unique_ptr GenerateInOutVariableCallerBodyFunction( + std::string caller, + std::string callee, + std::vector> inout_vars) { + auto body = std::make_unique(); + for (auto inout : inout_vars) { + std::string in, out; + std::tie(in, out) = inout; + body->append(std::make_unique( + std::make_unique(out), + std::make_unique(in))); + } + auto ident_expr = std::make_unique(callee); + auto call_expr = std::make_unique( + std::move(ident_expr), ast::ExpressionList()); + body->append(std::make_unique(std::move(call_expr))); + body->append(std::make_unique()); + auto func = std::make_unique(caller, ast::VariableList(), + void_type()); + func->set_body(std::move(body)); + return func; + } + + bool ContainsString(const std::vector& vec, + const std::string& str) { + for (auto& s : vec) { + if (s == str) { + return true; + } + } + return false; + } + + ast::Module* mod() { return &mod_; } + TypeDeterminer* td() { return td_.get(); } Inspector* inspector() { return inspector_.get(); } + ast::type::VoidType* void_type() { return &void_type_; } + ast::type::U32Type* u32_type() { return &u32_type_; } + private: Context ctx_; - std::unique_ptr mod_; + ast::Module mod_; + std::unique_ptr td_; std::unique_ptr inspector_; - ast::Function* last_function_; + + ast::type::VoidType void_type_; + ast::type::U32Type u32_type_; }; class InspectorTest : public InspectorHelper, public testing::Test {}; @@ -70,7 +179,7 @@ TEST_F(InspectorGetEntryPointTest, NoFunctions) { } TEST_F(InspectorGetEntryPointTest, NoEntryPoints) { - AddFunction("foo", ast::PipelineStage::kNone); + mod()->AddFunction(GenerateEmptyBodyFunction("foo")); auto result = inspector()->GetEntryPoints(); ASSERT_FALSE(inspector()->has_error()); @@ -79,24 +188,34 @@ TEST_F(InspectorGetEntryPointTest, NoEntryPoints) { } TEST_F(InspectorGetEntryPointTest, OneEntryPoint) { - AddFunction("foo", ast::PipelineStage::kVertex); + auto foo = GenerateEmptyBodyFunction("foo"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); auto result = inspector()->GetEntryPoints(); ASSERT_FALSE(inspector()->has_error()); - EXPECT_EQ(1u, result.size()); + ASSERT_EQ(1u, result.size()); EXPECT_EQ("foo", result[0].name); EXPECT_EQ(ast::PipelineStage::kVertex, result[0].stage); } TEST_F(InspectorGetEntryPointTest, MultipleEntryPoints) { - AddFunction("foo", ast::PipelineStage::kVertex); - AddFunction("bar", ast::PipelineStage::kCompute); + auto foo = GenerateEmptyBodyFunction("foo"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + auto bar = GenerateEmptyBodyFunction("bar"); + bar->add_decoration( + std::make_unique(ast::PipelineStage::kCompute)); + mod()->AddFunction(std::move(bar)); auto result = inspector()->GetEntryPoints(); ASSERT_FALSE(inspector()->has_error()); - EXPECT_EQ(2u, result.size()); + ASSERT_EQ(2u, result.size()); EXPECT_EQ("foo", result[0].name); EXPECT_EQ(ast::PipelineStage::kVertex, result[0].stage); EXPECT_EQ("bar", result[1].name); @@ -104,22 +223,34 @@ TEST_F(InspectorGetEntryPointTest, MultipleEntryPoints) { } TEST_F(InspectorGetEntryPointTest, MixFunctionsAndEntryPoints) { - AddFunction("foo", ast::PipelineStage::kVertex); - AddFunction("func", ast::PipelineStage::kNone); - AddFunction("bar", ast::PipelineStage::kCompute); + auto func = GenerateEmptyBodyFunction("func"); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateCallerBodyFunction("foo", "func"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + auto bar = GenerateCallerBodyFunction("bar", "func"); + bar->add_decoration( + std::make_unique(ast::PipelineStage::kFragment)); + mod()->AddFunction(std::move(bar)); auto result = inspector()->GetEntryPoints(); EXPECT_FALSE(inspector()->has_error()); - EXPECT_EQ(2u, result.size()); + ASSERT_EQ(2u, result.size()); EXPECT_EQ("foo", result[0].name); EXPECT_EQ(ast::PipelineStage::kVertex, result[0].stage); EXPECT_EQ("bar", result[1].name); - EXPECT_EQ(ast::PipelineStage::kCompute, result[1].stage); + EXPECT_EQ(ast::PipelineStage::kFragment, result[1].stage); } TEST_F(InspectorGetEntryPointTest, DefaultWorkgroupSize) { - AddFunction("foo", ast::PipelineStage::kVertex); + auto foo = GenerateCallerBodyFunction("foo", "func"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); auto result = inspector()->GetEntryPoints(); ASSERT_FALSE(inspector()->has_error()); @@ -133,8 +264,12 @@ TEST_F(InspectorGetEntryPointTest, DefaultWorkgroupSize) { } TEST_F(InspectorGetEntryPointTest, NonDefaultWorkgroupSize) { - AddFunction("foo", ast::PipelineStage::kCompute); - AddWorkGroupSizeToLastFunction(8u, 2u, 1u); + auto foo = GenerateEmptyBodyFunction("foo"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kCompute)); + foo->add_decoration(std::make_unique(8u, 2u, 1u)); + mod()->AddFunction(std::move(foo)); + auto result = inspector()->GetEntryPoints(); ASSERT_FALSE(inspector()->has_error()); @@ -146,6 +281,219 @@ TEST_F(InspectorGetEntryPointTest, NonDefaultWorkgroupSize) { EXPECT_EQ(1u, z); } +TEST_F(InspectorGetEntryPointTest, NoInOutVariables) { + auto func = GenerateEmptyBodyFunction("func"); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateCallerBodyFunction("foo", "func"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + EXPECT_EQ(0u, result[0].input_variables.size()); + EXPECT_EQ(0u, result[0].output_variables.size()); +} + +TEST_F(InspectorGetEntryPointTest, EntryPointInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}}); + + auto foo = GenerateInOutVariableBodyFunction("foo", {{"in_var", "out_var"}}); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + + ASSERT_EQ(1u, result[0].input_variables.size()); + EXPECT_EQ("in_var", result[0].input_variables[0]); + ASSERT_EQ(1u, result[0].output_variables.size()); + EXPECT_EQ("out_var", result[0].output_variables[0]); +} + +TEST_F(InspectorGetEntryPointTest, FunctionInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}}); + + auto func = + GenerateInOutVariableBodyFunction("func", {{"in_var", "out_var"}}); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateCallerBodyFunction("foo", "func"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + + ASSERT_EQ(1u, result[0].input_variables.size()); + EXPECT_EQ("in_var", result[0].input_variables[0]); + ASSERT_EQ(1u, result[0].output_variables.size()); + EXPECT_EQ("out_var", result[0].output_variables[0]); +} + +TEST_F(InspectorGetEntryPointTest, RepeatedInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}}); + + auto func = + GenerateInOutVariableBodyFunction("func", {{"in_var", "out_var"}}); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateInOutVariableCallerBodyFunction("foo", "func", + {{"in_var", "out_var"}}); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + + ASSERT_EQ(1u, result[0].input_variables.size()); + EXPECT_EQ("in_var", result[0].input_variables[0]); + ASSERT_EQ(1u, result[0].output_variables.size()); + EXPECT_EQ("out_var", result[0].output_variables[0]); +} + +TEST_F(InspectorGetEntryPointTest, EntryPointMultipleInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + + auto foo = GenerateInOutVariableBodyFunction( + "foo", {{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + + ASSERT_EQ(2u, result[0].input_variables.size()); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in_var")); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in2_var")); + ASSERT_EQ(2u, result[0].output_variables.size()); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out_var")); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out2_var")); +} + +TEST_F(InspectorGetEntryPointTest, FunctionMultipleInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + + auto func = GenerateInOutVariableBodyFunction( + "func", {{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateCallerBodyFunction("foo", "func"); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(1u, result.size()); + + ASSERT_EQ(2u, result[0].input_variables.size()); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in_var")); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in2_var")); + ASSERT_EQ(2u, result[0].output_variables.size()); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out_var")); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out2_var")); +} + +TEST_F(InspectorGetEntryPointTest, MultipleEntryPointsInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + + auto foo = GenerateInOutVariableBodyFunction("foo", {{"in_var", "out2_var"}}); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + auto bar = GenerateInOutVariableBodyFunction("bar", {{"in2_var", "out_var"}}); + bar->add_decoration( + std::make_unique(ast::PipelineStage::kCompute)); + mod()->AddFunction(std::move(bar)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(2u, result.size()); + + ASSERT_EQ("foo", result[0].name); + ASSERT_EQ(1u, result[0].input_variables.size()); + EXPECT_EQ("in_var", result[0].input_variables[0]); + ASSERT_EQ(1u, result[0].output_variables.size()); + EXPECT_EQ("out2_var", result[0].output_variables[0]); + + ASSERT_EQ("bar", result[1].name); + ASSERT_EQ(1u, result[1].input_variables.size()); + EXPECT_EQ("in2_var", result[1].input_variables[0]); + ASSERT_EQ(1u, result[1].output_variables.size()); + EXPECT_EQ("out_var", result[1].output_variables[0]); +} + +TEST_F(InspectorGetEntryPointTest, MultipleEntryPointsSharedInOutVariables) { + CreateInOutVariables({{"in_var", "out_var"}, {"in2_var", "out2_var"}}); + + auto func = + GenerateInOutVariableBodyFunction("func", {{"in2_var", "out2_var"}}); + mod()->AddFunction(std::move(func)); + + auto foo = GenerateInOutVariableCallerBodyFunction("foo", "func", + {{"in_var", "out_var"}}); + foo->add_decoration( + std::make_unique(ast::PipelineStage::kVertex)); + mod()->AddFunction(std::move(foo)); + + auto bar = GenerateCallerBodyFunction("bar", "func"); + bar->add_decoration( + std::make_unique(ast::PipelineStage::kCompute)); + mod()->AddFunction(std::move(bar)); + + ASSERT_TRUE(td()->Determine()) << td()->error(); + + auto result = inspector()->GetEntryPoints(); + ASSERT_FALSE(inspector()->has_error()); + + ASSERT_EQ(2u, result.size()); + + ASSERT_EQ("foo", result[0].name); + EXPECT_EQ(2u, result[0].input_variables.size()); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in_var")); + EXPECT_TRUE(ContainsString(result[0].input_variables, "in2_var")); + EXPECT_EQ(2u, result[0].output_variables.size()); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out_var")); + EXPECT_TRUE(ContainsString(result[0].output_variables, "out2_var")); + + ASSERT_EQ("bar", result[1].name); + EXPECT_EQ(1u, result[1].input_variables.size()); + EXPECT_EQ("in2_var", result[1].input_variables[0]); + EXPECT_EQ(1u, result[1].output_variables.size()); + EXPECT_EQ("out2_var", result[1].output_variables[0]); +} + } // namespace } // namespace inspector } // namespace tint