diff --git a/src/tint/BUILD.gn b/src/tint/BUILD.gn index 1b2e7d592b..c0c0b62236 100644 --- a/src/tint/BUILD.gn +++ b/src/tint/BUILD.gn @@ -420,6 +420,7 @@ libtint_source_set("libtint_core_all_src") { "sem/constant.h", "sem/depth_multisampled_texture.h", "sem/depth_texture.h", + "sem/evaluation_stage.h", "sem/expression.h", "sem/external_texture.h", "sem/f16.h", @@ -621,6 +622,7 @@ libtint_source_set("libtint_sem_src") { "sem/depth_multisampled_texture.h", "sem/depth_texture.cc", "sem/depth_texture.h", + "sem/evaluation_stage.h", "sem/expression.cc", "sem/expression.h", "sem/external_texture.cc", @@ -1093,6 +1095,7 @@ if (tint_build_unittests) { "resolver/control_block_validation_test.cc", "resolver/dependency_graph_test.cc", "resolver/entry_point_validation_test.cc", + "resolver/evaluation_stage_test.cc", "resolver/function_validation_test.cc", "resolver/host_shareable_validation_test.cc", "resolver/increment_decrement_validation_test.cc", diff --git a/src/tint/CMakeLists.txt b/src/tint/CMakeLists.txt index 1d22340509..6c611c306a 100644 --- a/src/tint/CMakeLists.txt +++ b/src/tint/CMakeLists.txt @@ -298,6 +298,7 @@ set(TINT_LIB_SRCS sem/depth_multisampled_texture.h sem/depth_texture.cc sem/depth_texture.h + sem/evaluation_stage.h sem/expression.cc sem/expression.h sem/external_texture.cc @@ -777,6 +778,7 @@ if(TINT_BUILD_TESTS) resolver/control_block_validation_test.cc resolver/dependency_graph_test.cc resolver/entry_point_validation_test.cc + resolver/evaluation_stage_test.cc resolver/function_validation_test.cc resolver/host_shareable_validation_test.cc resolver/increment_decrement_validation_test.cc diff --git a/src/tint/resolver/evaluation_stage_test.cc b/src/tint/resolver/evaluation_stage_test.cc new file mode 100644 index 0000000000..236e4b353f --- /dev/null +++ b/src/tint/resolver/evaluation_stage_test.cc @@ -0,0 +1,297 @@ +// Copyright 2022 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#include "src/tint/resolver/resolver.h" + +#include "gmock/gmock.h" +#include "src/tint/resolver/resolver_test_helper.h" + +using namespace tint::number_suffixes; // NOLINT + +namespace tint::resolver { +namespace { + +using ResolverEvaluationStageTest = ResolverTest; + +TEST_F(ResolverEvaluationStageTest, Literal_i32) { + auto* expr = Expr(123_i); + WrapInFunction(expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Literal_f32) { + auto* expr = Expr(123_f); + WrapInFunction(expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Vector_Ctor) { + auto* expr = vec3(); + WrapInFunction(expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Vector_Ctor_Const_Const) { + // const f = 1.f; + // vec2(f, f); + auto* f = Const("f", nullptr, Expr(1_f)); + auto* expr = vec2(f, f); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Vector_Ctor_Runtime_Runtime) { + // var f = 1.f; + // vec2(f, f); + auto* f = Var("f", nullptr, Expr(1_f)); + auto* expr = vec2(f, f); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, Vector_Conv_Const) { + // const f = 1.f; + // vec2(vec2(f)); + auto* f = Const("f", nullptr, Expr(1_f)); + auto* expr = vec2(vec2(f)); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Vector_Conv_Runtime) { + // var f = 1.f; + // vec2(vec2(f)); + auto* f = Var("f", nullptr, Expr(1_f)); + auto* expr = vec2(vec2(f)); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, Matrix_Ctor) { + auto* expr = mat2x2(); + WrapInFunction(expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor) { + auto* expr = array(); + WrapInFunction(expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor_Const_Const) { + // const f = 1.f; + // array(f, f); + auto* f = Const("f", nullptr, Expr(1_f)); + auto* expr = Construct(ty.array(), f, f); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor_Const_Override) { + // const f1 = 1.f; + // override f2 = 2.f; + // array(f1, f2); + auto* f1 = Const("f1", nullptr, Expr(1_f)); + auto* f2 = Override("f2", nullptr, Expr(2_f)); + auto* expr = Construct(ty.array(), f1, f2); + WrapInFunction(f1, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f1)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(f2)->Stage(), sem::EvaluationStage::kOverride); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kOverride); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor_Override_Runtime) { + // override f1 = 1.f; + // var f2 = 2.f; + // array(f1, f2); + auto* f1 = Override("f1", nullptr, Expr(1_f)); + auto* f2 = Var("f2", nullptr, Expr(2_f)); + auto* expr = Construct(ty.array(), f1, f2); + WrapInFunction(f2, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f1)->Stage(), sem::EvaluationStage::kOverride); + EXPECT_EQ(Sem().Get(f2)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor_Const_Runtime) { + // const f1 = 1.f; + // var f2 = 2.f; + // array(f1, f2); + auto* f1 = Const("f1", nullptr, Expr(1_f)); + auto* f2 = Var("f2", nullptr, Expr(2_f)); + auto* expr = Construct(ty.array(), f1, f2); + WrapInFunction(f1, f2, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f1)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(f2)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, Array_Ctor_Runtime_Runtime) { + // var f = 1.f; + // array(f, f); + auto* f = Var("f", nullptr, Expr(1_f)); + auto* expr = Construct(ty.array(), f, f); + WrapInFunction(f, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(f)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, IndexAccessor_Const_Const) { + // const vec = vec4(); + // const idx = 1_i; + // vec[idx] + auto* vec = Const("vec", nullptr, vec4()); + auto* idx = Const("idx", nullptr, Expr(1_i)); + auto* expr = IndexAccessor(vec, idx); + WrapInFunction(vec, idx, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(idx)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, IndexAccessor_Runtime_Const) { + // var vec = vec4(); + // const idx = 1_i; + // vec[idx] + auto* vec = Var("vec", nullptr, vec4()); + auto* idx = Const("idx", nullptr, Expr(1_i)); + auto* expr = IndexAccessor(vec, idx); + WrapInFunction(vec, idx, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(idx)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, IndexAccessor_Const_Override) { + // const vec = vec4(); + // override idx = 1_i; + // vec[idx] + auto* vec = Const("vec", nullptr, vec4()); + auto* idx = Override("idx", nullptr, Expr(1_i)); + auto* expr = IndexAccessor(vec, idx); + WrapInFunction(vec, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(idx)->Stage(), sem::EvaluationStage::kOverride); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kOverride); +} + +TEST_F(ResolverEvaluationStageTest, IndexAccessor_Const_Runtime) { + // const vec = vec4(); + // let idx = 1_i; + // vec[idx] + auto* vec = Const("vec", nullptr, vec4()); + auto* idx = Let("idx", nullptr, Expr(1_i)); + auto* expr = IndexAccessor(vec, idx); + WrapInFunction(vec, idx, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(idx)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, Swizzle_Const) { + // const vec = S(); + // vec.m + auto* vec = Const("vec", nullptr, vec4()); + auto* expr = MemberAccessor(vec, "xz"); + WrapInFunction(vec, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, Swizzle_Runtime) { + // var vec = S(); + // vec.m + auto* vec = Var("vec", nullptr, vec4()); + auto* expr = MemberAccessor(vec, "rg"); + WrapInFunction(vec, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(vec)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +TEST_F(ResolverEvaluationStageTest, MemberAccessor_Const) { + // struct S { m : i32 }; + // const str = S(); + // str.m + Structure("S", {Member("m", ty.i32())}); + auto* str = Const("str", nullptr, Construct(ty.type_name("S"))); + auto* expr = MemberAccessor(str, "m"); + WrapInFunction(str, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(str)->Stage(), sem::EvaluationStage::kConstant); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kConstant); +} + +TEST_F(ResolverEvaluationStageTest, MemberAccessor_Runtime) { + // struct S { m : i32 }; + // var str = S(); + // str.m + Structure("S", {Member("m", ty.i32())}); + auto* str = Var("str", nullptr, Construct(ty.type_name("S"))); + auto* expr = MemberAccessor(str, "m"); + WrapInFunction(str, expr); + + ASSERT_TRUE(r()->Resolve()) << r()->error(); + EXPECT_EQ(Sem().Get(str)->Stage(), sem::EvaluationStage::kRuntime); + EXPECT_EQ(Sem().Get(expr)->Stage(), sem::EvaluationStage::kRuntime); +} + +} // namespace +} // namespace tint::resolver diff --git a/src/tint/resolver/intrinsic_table.cc b/src/tint/resolver/intrinsic_table.cc index bd7dc30933..68a1068d09 100644 --- a/src/tint/resolver/intrinsic_table.cc +++ b/src/tint/resolver/intrinsic_table.cc @@ -1119,8 +1119,10 @@ Impl::Builtin Impl::Lookup(sem::BuiltinType builtin_type, if (match.overload->flags.Contains(OverloadFlag::kSupportsComputePipeline)) { supported_stages.Add(ast::PipelineStage::kCompute); } + auto eval_stage = match.overload->const_eval_fn ? sem::EvaluationStage::kConstant + : sem::EvaluationStage::kRuntime; return builder.create( - builtin_type, match.return_type, std::move(params), supported_stages, + builtin_type, match.return_type, std::move(params), eval_stage, supported_stages, match.overload->flags.Contains(OverloadFlag::kIsDeprecated)); }); return Builtin{sem, match.overload->const_eval_fn}; @@ -1292,8 +1294,11 @@ IntrinsicTable::CtorOrConv Impl::Lookup(CtorConvIntrinsic type, nullptr, static_cast(params.size()), p.type, ast::StorageClass::kNone, ast::Access::kUndefined, p.usage)); } + auto eval_stage = match.overload->const_eval_fn ? sem::EvaluationStage::kConstant + : sem::EvaluationStage::kRuntime; auto* target = utils::GetOrCreate(constructors, match, [&]() { - return builder.create(match.return_type, std::move(params)); + return builder.create(match.return_type, std::move(params), + eval_stage); }); return CtorOrConv{target, match.overload->const_eval_fn}; } @@ -1303,7 +1308,9 @@ IntrinsicTable::CtorOrConv Impl::Lookup(CtorConvIntrinsic type, auto param = builder.create( nullptr, 0u, match.parameters[0].type, ast::StorageClass::kNone, ast::Access::kUndefined, match.parameters[0].usage); - return builder.create(match.return_type, param); + auto eval_stage = match.overload->const_eval_fn ? sem::EvaluationStage::kConstant + : sem::EvaluationStage::kRuntime; + return builder.create(match.return_type, param, eval_stage); }); return CtorOrConv{target, match.overload->const_eval_fn}; } diff --git a/src/tint/resolver/resolver.cc b/src/tint/resolver/resolver.cc index 8d09d19388..7fc09cc91a 100644 --- a/src/tint/resolver/resolver.cc +++ b/src/tint/resolver/resolver.cc @@ -367,10 +367,11 @@ sem::Variable* Resolver::Let(const ast::Let* v, bool is_global) { sem::Variable* sem = nullptr; if (is_global) { sem = builder_->create( - v, ty, ast::StorageClass::kNone, ast::Access::kUndefined, /* constant_value */ nullptr, - sem::BindingPoint{}); + v, ty, sem::EvaluationStage::kRuntime, ast::StorageClass::kNone, + ast::Access::kUndefined, /* constant_value */ nullptr, sem::BindingPoint{}); } else { - sem = builder_->create(v, ty, ast::StorageClass::kNone, + sem = builder_->create(v, ty, sem::EvaluationStage::kRuntime, + ast::StorageClass::kNone, ast::Access::kUndefined, current_statement_, /* constant_value */ nullptr); } @@ -421,8 +422,8 @@ sem::Variable* Resolver::Override(const ast::Override* v) { } auto* sem = builder_->create( - v, ty, ast::StorageClass::kNone, ast::Access::kUndefined, /* constant_value */ nullptr, - sem::BindingPoint{}); + v, ty, sem::EvaluationStage::kOverride, ast::StorageClass::kNone, ast::Access::kUndefined, + /* constant_value */ nullptr, sem::BindingPoint{}); if (auto* id = ast::GetAttribute(v->attributes)) { sem->SetConstantId(static_cast(id->value)); @@ -482,11 +483,11 @@ sem::Variable* Resolver::Const(const ast::Const* c, bool is_global) { } auto* sem = is_global ? static_cast(builder_->create( - c, ty, ast::StorageClass::kNone, ast::Access::kUndefined, value, - sem::BindingPoint{})) + c, ty, sem::EvaluationStage::kConstant, ast::StorageClass::kNone, + ast::Access::kUndefined, value, sem::BindingPoint{})) : static_cast(builder_->create( - c, ty, ast::StorageClass::kNone, ast::Access::kUndefined, - current_statement_, value)); + c, ty, sem::EvaluationStage::kConstant, ast::StorageClass::kNone, + ast::Access::kUndefined, current_statement_, value)); sem->SetConstructor(rhs); builder_->Sem().Add(c, sem); @@ -567,12 +568,14 @@ sem::Variable* Resolver::Var(const ast::Var* var, bool is_global) { if (auto bp = var->BindingPoint()) { binding_point = {bp.group->value, bp.binding->value}; } - sem = builder_->create(var, var_ty, storage_class, access, + sem = builder_->create(var, var_ty, sem::EvaluationStage::kRuntime, + storage_class, access, /* constant_value */ nullptr, binding_point); } else { - sem = builder_->create( - var, var_ty, storage_class, access, current_statement_, /* constant_value */ nullptr); + sem = builder_->create(var, var_ty, sem::EvaluationStage::kRuntime, + storage_class, access, current_statement_, + /* constant_value */ nullptr); } sem->SetConstructor(rhs); @@ -1270,6 +1273,7 @@ sem::Expression* Resolver::Expression(const ast::Expression* root) { [&](const ast::UnaryOpExpression* unary) -> sem::Expression* { return UnaryOp(unary); }, [&](const ast::PhonyExpression*) -> sem::Expression* { return builder_->create(expr, builder_->create(), + sem::EvaluationStage::kRuntime, current_statement_, /* constant_value */ nullptr, /* has_side_effects */ false); @@ -1423,10 +1427,11 @@ sem::Expression* Resolver::IndexAccessor(const ast::IndexAccessorExpression* exp ty = builder_->create(ty, ref->StorageClass(), ref->Access()); } + auto stage = sem::EarliestStage(obj->Stage(), idx->Stage()); auto val = const_eval_.Index(obj, idx); bool has_side_effects = idx->HasSideEffects() || obj->HasSideEffects(); auto* sem = builder_->create( - expr, ty, obj, idx, current_statement_, std::move(val), has_side_effects, + expr, ty, stage, obj, idx, current_statement_, std::move(val), has_side_effects, obj->SourceVariable()); sem->Behaviors() = idx->Behaviors() + obj->Behaviors(); return sem; @@ -1443,8 +1448,9 @@ sem::Expression* Resolver::Bitcast(const ast::BitcastExpression* expr) { } auto val = const_eval_.Bitcast(ty, inner); - auto* sem = builder_->create(expr, ty, current_statement_, std::move(val), - inner->HasSideEffects()); + auto stage = sem::EvaluationStage::kRuntime; // TODO(crbug.com/tint/1581) + auto* sem = builder_->create(expr, ty, stage, current_statement_, + std::move(val), inner->HasSideEffects()); sem->Behaviors() = inner->Behaviors(); @@ -1464,6 +1470,7 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { // Resolve all of the arguments, their types and the set of behaviors. std::vector args(expr->args.size()); + auto args_stage = sem::EvaluationStage::kConstant; sem::Behaviors arg_behaviors; for (size_t i = 0; i < expr->args.size(); i++) { auto* arg = sem_.Get(expr->args[i]); @@ -1471,6 +1478,7 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { return nullptr; } args[i] = arg; + args_stage = sem::EarliestStage(args_stage, arg->Stage()); arg_behaviors.Add(arg->Behaviors()); } arg_behaviors.Remove(sem::Behavior::kNext); @@ -1491,14 +1499,38 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { return nullptr; } const sem::Constant* value = nullptr; - if (ctor_or_conv.const_eval_fn) { + auto stage = sem::EarliestStage(ctor_or_conv.target->Stage(), args_stage); + if (stage == sem::EvaluationStage::kConstant) { value = (const_eval_.*ctor_or_conv.const_eval_fn)(ctor_or_conv.target->ReturnType(), args.data(), args.size()); } - return builder_->create(expr, ctor_or_conv.target, std::move(args), + return builder_->create(expr, ctor_or_conv.target, stage, std::move(args), current_statement_, value, has_side_effects); }; + // ct_ctor_or_conv is a helper for building a sem::TypeConstructor for an array or structure + // constructor call target. + auto arr_or_str_ctor = [&](const sem::Type* ty, + const sem::CallTarget* call_target) -> sem::Call* { + if (!MaterializeArguments(args, call_target)) { + return nullptr; + } + + auto stage = args_stage; // The evaluation stage of the call + const sem::Constant* value = nullptr; // The constant value for the call + if (stage == sem::EvaluationStage::kConstant) { + value = const_eval_.ArrayOrStructCtor(ty, args); + if (!value) { + // Constant evaluation failed. + // Can happen for expressions that will fail validation (later). + stage = sem::EvaluationStage::kRuntime; + } + } + + return builder_->create(expr, call_target, stage, std::move(args), + current_statement_, std::move(value), has_side_effects); + }; + // ct_ctor_or_conv is a helper for building either a sem::TypeConstructor or sem::TypeConversion // call for the given semantic type. auto ty_ctor_or_conv = [&](const sem::Type* ty) { @@ -1517,7 +1549,7 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { [&](const sem::Bool*) { return ct_ctor_or_conv(CtorConvIntrinsic::kBool, nullptr); }, [&](const sem::Array* arr) -> sem::Call* { auto* call_target = utils::GetOrCreate( - array_ctors_, ArrayConstructorSig{{arr, args.size()}}, + array_ctors_, ArrayConstructorSig{{arr, args.size(), args_stage}}, [&]() -> sem::TypeConstructor* { sem::ParameterList params(args.size()); for (size_t i = 0; i < args.size(); i++) { @@ -1528,18 +1560,14 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { ast::StorageClass::kNone, // storage_class ast::Access::kUndefined); // access } - return builder_->create(arr, std::move(params)); + return builder_->create(arr, std::move(params), + args_stage); }); - if (!MaterializeArguments(args, call_target)) { - return nullptr; - } - auto val = const_eval_.ArrayOrStructCtor(arr, args); - return builder_->create(expr, call_target, std::move(args), - current_statement_, val, has_side_effects); + return arr_or_str_ctor(arr, call_target); }, [&](const sem::Struct* str) -> sem::Call* { auto* call_target = utils::GetOrCreate( - struct_ctors_, StructConstructorSig{{str, args.size()}}, + struct_ctors_, StructConstructorSig{{str, args.size(), args_stage}}, [&]() -> sem::TypeConstructor* { sem::ParameterList params(std::min(args.size(), str->Members().size())); for (size_t i = 0, n = params.size(); i < n; i++) { @@ -1550,15 +1578,10 @@ sem::Call* Resolver::Call(const ast::CallExpression* expr) { ast::StorageClass::kNone, // storage_class ast::Access::kUndefined); // access } - return builder_->create(str, std::move(params)); + return builder_->create(str, std::move(params), + args_stage); }); - if (!MaterializeArguments(args, call_target)) { - return nullptr; - } - auto val = const_eval_.ArrayOrStructCtor(str, args); - return builder_->create(expr, call_target, std::move(args), - current_statement_, std::move(val), - has_side_effects); + return arr_or_str_ctor(str, call_target); }, [&](Default) { AddError("type is not constructible", expr->source); @@ -1686,9 +1709,16 @@ sem::Call* Resolver::BuiltinCall(const ast::CallExpression* expr, AddWarning("use of deprecated builtin", expr->source); } + auto stage = builtin.sem->Stage(); + if (stage == sem::EvaluationStage::kConstant) { + for (auto* arg : args) { + stage = sem::EarliestStage(stage, arg->Stage()); + } + } + // If the builtin is @const, and all arguments have constant values, evaluate the builtin now. const sem::Constant* value = nullptr; - if (builtin.const_eval_fn) { + if (stage == sem::EvaluationStage::kConstant && builtin.const_eval_fn) { value = (const_eval_.*builtin.const_eval_fn)(builtin.sem->ReturnType(), args.data(), args.size()); } @@ -1696,8 +1726,8 @@ sem::Call* Resolver::BuiltinCall(const ast::CallExpression* expr, bool has_side_effects = builtin.sem->HasSideEffects() || std::any_of(args.begin(), args.end(), [](auto* e) { return e->HasSideEffects(); }); - auto* call = builder_->create(expr, builtin.sem, std::move(args), current_statement_, - value, has_side_effects); + auto* call = builder_->create(expr, builtin.sem, stage, std::move(args), + current_statement_, value, has_side_effects); if (current_function_) { current_function_->AddDirectlyCalledBuiltin(builtin.sem); @@ -1755,7 +1785,8 @@ sem::Call* Resolver::FunctionCall(const ast::CallExpression* expr, // TODO(crbug.com/tint/1420): For now, assume all function calls have side // effects. bool has_side_effects = true; - auto* call = builder_->create(expr, target, std::move(args), current_statement_, + auto* call = builder_->create(expr, target, sem::EvaluationStage::kRuntime, + std::move(args), current_statement_, /* constant_value */ nullptr, has_side_effects); target->AddCallSite(call); @@ -1851,7 +1882,8 @@ sem::Expression* Resolver::Literal(const ast::LiteralExpression* literal) { } auto val = const_eval_.Literal(ty, literal); - return builder_->create(literal, ty, current_statement_, std::move(val), + return builder_->create(literal, ty, sem::EvaluationStage::kConstant, + current_statement_, std::move(val), /* has_side_effects */ false); } @@ -2054,6 +2086,7 @@ sem::Expression* Resolver::Binary(const ast::BinaryExpression* expr) { const auto* rhs = sem_.Get(expr->rhs); auto* lhs_ty = lhs->Type()->UnwrapRef(); auto* rhs_ty = rhs->Type()->UnwrapRef(); + auto stage = sem::EvaluationStage::kRuntime; // TODO(crbug.com/tint/1581) auto op = intrinsic_table_->Lookup(expr->op, lhs_ty, rhs_ty, expr->source, false); if (!op.result) { @@ -2079,7 +2112,7 @@ sem::Expression* Resolver::Binary(const ast::BinaryExpression* expr) { } bool has_side_effects = lhs->HasSideEffects() || rhs->HasSideEffects(); - auto* sem = builder_->create(expr, op.result, current_statement_, value, + auto* sem = builder_->create(expr, op.result, stage, current_statement_, value, has_side_effects); sem->Behaviors() = lhs->Behaviors() + rhs->Behaviors(); @@ -2096,6 +2129,7 @@ sem::Expression* Resolver::UnaryOp(const ast::UnaryOpExpression* unary) { const sem::Type* ty = nullptr; const sem::Variable* source_var = nullptr; const sem::Constant* value = nullptr; + auto stage = sem::EvaluationStage::kRuntime; // TODO(crbug.com/tint/1581) switch (unary->op) { case ast::UnaryOp::kAddressOf: @@ -2155,7 +2189,7 @@ sem::Expression* Resolver::UnaryOp(const ast::UnaryOpExpression* unary) { } } - auto* sem = builder_->create(unary, ty, current_statement_, value, + auto* sem = builder_->create(unary, ty, stage, current_statement_, value, expr->HasSideEffects(), source_var); sem->Behaviors() = expr->Behaviors(); return sem; diff --git a/src/tint/resolver/resolver.h b/src/tint/resolver/resolver.h index 245512749f..f3aa38b145 100644 --- a/src/tint/resolver/resolver.h +++ b/src/tint/resolver/resolver.h @@ -400,12 +400,15 @@ class Resolver { bool IsBuiltin(Symbol) const; // ArrayConstructorSig represents a unique array constructor signature. - // It is a tuple of the array type and number of arguments provided. - using ArrayConstructorSig = utils::UnorderedKeyWrapper>; + // It is a tuple of the array type, number of arguments provided and earliest evaluation stage. + using ArrayConstructorSig = + utils::UnorderedKeyWrapper>; // StructConstructorSig represents a unique structure constructor signature. - // It is a tuple of the structure type and number of arguments provided. - using StructConstructorSig = utils::UnorderedKeyWrapper>; + // It is a tuple of the structure type, number of arguments provided and earliest evaluation + // stage. + using StructConstructorSig = + utils::UnorderedKeyWrapper>; ProgramBuilder* const builder_; diag::List& diagnostics_; diff --git a/src/tint/resolver/validator.cc b/src/tint/resolver/validator.cc index c7afb543bb..29e06a2521 100644 --- a/src/tint/resolver/validator.cc +++ b/src/tint/resolver/validator.cc @@ -547,10 +547,9 @@ bool Validator::LocalVariable(const sem::Variable* v) const { } } return Var(v); - }, // - [&](const ast::Let*) { return Let(v); }, // - [&](const ast::Override*) { return Override(v); }, // - [&](const ast::Const*) { return true; }, // + }, // + [&](const ast::Let*) { return Let(v); }, // + [&](const ast::Const*) { return true; }, // [&](Default) { TINT_ICE(Resolver, diagnostics_) << "Validator::Variable() called with a unknown variable type: " @@ -567,6 +566,13 @@ bool Validator::GlobalVariable( bool ok = Switch( decl, // [&](const ast::Var* var) { + if (auto* init = global->Constructor(); + init && init->Stage() > sem::EvaluationStage::kOverride) { + AddError("module-scope 'var' initializer must be a constant or override expression", + init->Declaration()->source); + return false; + } + if (global->StorageClass() == ast::StorageClass::kNone) { AddError("module-scope 'var' declaration must have a storage class", decl->source); return false; @@ -619,31 +625,7 @@ bool Validator::GlobalVariable( return Var(global); }, - [&](const ast::Override*) { - for (auto* attr : decl->attributes) { - if (auto* id_attr = attr->As()) { - uint32_t id = id_attr->value; - auto it = constant_ids.find(id); - if (it != constant_ids.end() && it->second != global) { - AddError("pipeline constant IDs must be unique", attr->source); - AddNote("a pipeline constant with an ID of " + std::to_string(id) + - " was previously declared here:", - ast::GetAttribute( - it->second->Declaration()->attributes) - ->source); - return false; - } - if (id > 65535) { - AddError("pipeline constant IDs must be between 0 and 65535", attr->source); - return false; - } - } else { - AddError("attribute is not valid for 'override' declaration", attr->source); - return false; - } - } - return Override(global); - }, + [&](const ast::Override*) { return Override(global, constant_ids); }, [&](const ast::Const*) { if (!decl->attributes.empty()) { AddError("attribute is not valid for module-scope 'const' declaration", @@ -777,10 +759,39 @@ bool Validator::Let(const sem::Variable* v) const { return true; } -bool Validator::Override(const sem::Variable* v) const { +bool Validator::Override(const sem::Variable* v, + std::unordered_map constant_ids) const { auto* decl = v->Declaration(); auto* storage_ty = v->Type()->UnwrapRef(); + if (auto* init = v->Constructor(); init && init->Stage() > sem::EvaluationStage::kOverride) { + AddError("'override' initializer must be an override expression", + init->Declaration()->source); + return false; + } + + for (auto* attr : decl->attributes) { + if (auto* id_attr = attr->As()) { + uint32_t id = id_attr->value; + auto it = constant_ids.find(id); + if (it != constant_ids.end() && it->second != v) { + AddError("pipeline constant IDs must be unique", attr->source); + AddNote("a pipeline constant with an ID of " + std::to_string(id) + + " was previously declared here:", + ast::GetAttribute(it->second->Declaration()->attributes) + ->source); + return false; + } + if (id > 65535) { + AddError("pipeline constant IDs must be between 0 and 65535", attr->source); + return false; + } + } else { + AddError("attribute is not valid for 'override' declaration", attr->source); + return false; + } + } + auto name = symbols_.NameFor(decl->symbol); if (sem::ParseBuiltinType(name) != sem::BuiltinType::kNone) { AddError("'" + name + "' is a builtin and cannot be redeclared as a 'override'", diff --git a/src/tint/resolver/validator.h b/src/tint/resolver/validator.h index 4d41959777..1f15a9e8c1 100644 --- a/src/tint/resolver/validator.h +++ b/src/tint/resolver/validator.h @@ -371,8 +371,10 @@ class Validator { /// Validates a 'override' variable declaration /// @param v the variable to validate + /// @param constant_ids the set of constant ids in the module /// @returns true on success, false otherwise. - bool Override(const sem::Variable* v) const; + bool Override(const sem::Variable* v, + std::unordered_map constant_ids) const; /// Validates a 'const' variable declaration /// @param v the variable to validate diff --git a/src/tint/sem/builtin.cc b/src/tint/sem/builtin.cc index bb2878be3a..c688c4c015 100644 --- a/src/tint/sem/builtin.cc +++ b/src/tint/sem/builtin.cc @@ -90,9 +90,10 @@ bool IsDP4aBuiltin(BuiltinType i) { Builtin::Builtin(BuiltinType type, const sem::Type* return_type, std::vector parameters, + EvaluationStage eval_stage, PipelineStageSet supported_stages, bool is_deprecated) - : Base(return_type, utils::ToConstPtrVec(parameters)), + : Base(return_type, utils::ToConstPtrVec(parameters), eval_stage), type_(type), supported_stages_(supported_stages), is_deprecated_(is_deprecated) { diff --git a/src/tint/sem/builtin.h b/src/tint/sem/builtin.h index 1dc61ad9d5..783e6ca915 100644 --- a/src/tint/sem/builtin.h +++ b/src/tint/sem/builtin.h @@ -83,6 +83,7 @@ class Builtin final : public Castable { /// @param type the builtin type /// @param return_type the return type for the builtin call /// @param parameters the parameters for the builtin overload + /// @param eval_stage the earliest evaluation stage for a call to the builtin /// @param supported_stages the pipeline stages that this builtin can be /// used in /// @param is_deprecated true if the particular overload is considered @@ -90,6 +91,7 @@ class Builtin final : public Castable { Builtin(BuiltinType type, const sem::Type* return_type, std::vector parameters, + EvaluationStage eval_stage, PipelineStageSet supported_stages, bool is_deprecated); diff --git a/src/tint/sem/call.cc b/src/tint/sem/call.cc index bfce3b1c53..a20649db45 100644 --- a/src/tint/sem/call.cc +++ b/src/tint/sem/call.cc @@ -23,13 +23,17 @@ namespace tint::sem { Call::Call(const ast::CallExpression* declaration, const CallTarget* target, + EvaluationStage stage, std::vector arguments, const Statement* statement, const Constant* constant, bool has_side_effects) - : Base(declaration, target->ReturnType(), statement, constant, has_side_effects), + : Base(declaration, target->ReturnType(), stage, statement, constant, has_side_effects), target_(target), - arguments_(std::move(arguments)) {} + arguments_(std::move(arguments)) { + // Check that the stage is no earlier than the target supports + TINT_ASSERT(Semantic, target->Stage() <= stage); +} Call::~Call() = default; diff --git a/src/tint/sem/call.h b/src/tint/sem/call.h index 1955bf9640..2d8230657b 100644 --- a/src/tint/sem/call.h +++ b/src/tint/sem/call.h @@ -30,12 +30,14 @@ class Call final : public Castable { /// Constructor /// @param declaration the AST node /// @param target the call target + /// @param stage the earliest evaluation stage for the expression /// @param arguments the call arguments /// @param statement the statement that owns this expression /// @param constant the constant value of this expression /// @param has_side_effects whether this expression may have side effects Call(const ast::CallExpression* declaration, const CallTarget* target, + EvaluationStage stage, std::vector arguments, const Statement* statement, const Constant* constant, diff --git a/src/tint/sem/call_target.cc b/src/tint/sem/call_target.cc index 67bde0e144..f8bcdd8253 100644 --- a/src/tint/sem/call_target.cc +++ b/src/tint/sem/call_target.cc @@ -21,8 +21,10 @@ TINT_INSTANTIATE_TYPEINFO(tint::sem::CallTarget); namespace tint::sem { -CallTarget::CallTarget(const sem::Type* return_type, const ParameterList& parameters) - : signature_{return_type, parameters} { +CallTarget::CallTarget(const sem::Type* return_type, + const ParameterList& parameters, + EvaluationStage stage) + : signature_{return_type, parameters}, stage_(stage) { TINT_ASSERT(Semantic, return_type); } diff --git a/src/tint/sem/call_target.h b/src/tint/sem/call_target.h index 64716b2913..be1b96ac43 100644 --- a/src/tint/sem/call_target.h +++ b/src/tint/sem/call_target.h @@ -63,9 +63,12 @@ struct CallTargetSignature { class CallTarget : public Castable { public: /// Constructor + /// @param stage the earliest evaluation stage for a call to this target /// @param return_type the return type of the call target /// @param parameters the parameters for the call target - CallTarget(const sem::Type* return_type, const ParameterList& parameters); + CallTarget(const sem::Type* return_type, + const ParameterList& parameters, + EvaluationStage stage); /// Copy constructor CallTarget(const CallTarget&); @@ -82,8 +85,12 @@ class CallTarget : public Castable { /// @return the signature of the call target const CallTargetSignature& Signature() const { return signature_; } + /// @return the earliest evaluation stage for a call to this target + EvaluationStage Stage() const { return stage_; } + private: CallTargetSignature signature_; + EvaluationStage stage_; }; } // namespace tint::sem diff --git a/src/tint/sem/evaluation_stage.h b/src/tint/sem/evaluation_stage.h new file mode 100644 index 0000000000..b5e554d873 --- /dev/null +++ b/src/tint/sem/evaluation_stage.h @@ -0,0 +1,60 @@ +// Copyright 2022 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#ifndef SRC_TINT_SEM_EVALUATION_STAGE_H_ +#define SRC_TINT_SEM_EVALUATION_STAGE_H_ + +#include +#include + +namespace tint::sem { + +/// The earliest point in time that an expression can be evaluated +enum class EvaluationStage { + /// Expression can be evaluated at shader creation time + kConstant, + /// Expression can be evaluated at pipeline creation time + kOverride, + /// Expression can be evaluated at runtime + kRuntime, +}; + +/// @returns true if stage `a` comes earlier than stage `b` +inline bool operator<(EvaluationStage a, EvaluationStage b) { + return static_cast(a) < static_cast(b); +} + +/// @returns true if stage `a` comes later than stage `b` +inline bool operator>(EvaluationStage a, EvaluationStage b) { + return static_cast(a) > static_cast(b); +} + +/// @param stages a list of EvaluationStage. +/// @returns the earliest stage supported by all the provided stages +inline EvaluationStage EarliestStage(std::initializer_list stages) { + auto earliest = EvaluationStage::kConstant; + for (auto stage : stages) { + earliest = std::max(stage, earliest); + } + return static_cast(earliest); +} + +template +inline EvaluationStage EarliestStage(ARGS... args) { + return EarliestStage({args...}); +} + +} // namespace tint::sem + +#endif // SRC_TINT_SEM_EVALUATION_STAGE_H_ diff --git a/src/tint/sem/expression.cc b/src/tint/sem/expression.cc index 4415db5c48..cdaa2bca3d 100644 --- a/src/tint/sem/expression.cc +++ b/src/tint/sem/expression.cc @@ -24,6 +24,7 @@ namespace tint::sem { Expression::Expression(const ast::Expression* declaration, const sem::Type* type, + EvaluationStage stage, const Statement* statement, const Constant* constant, bool has_side_effects, @@ -31,10 +32,12 @@ Expression::Expression(const ast::Expression* declaration, : declaration_(declaration), source_variable_(source_var), type_(type), + stage_(stage), statement_(statement), constant_(std::move(constant)), has_side_effects_(has_side_effects) { TINT_ASSERT(Semantic, type_); + TINT_ASSERT(Semantic, (constant != nullptr) == (stage == EvaluationStage::kConstant)); } Expression::~Expression() = default; diff --git a/src/tint/sem/expression.h b/src/tint/sem/expression.h index 45a67cafde..5942a71108 100644 --- a/src/tint/sem/expression.h +++ b/src/tint/sem/expression.h @@ -18,6 +18,7 @@ #include "src/tint/ast/expression.h" #include "src/tint/sem/behavior.h" #include "src/tint/sem/constant.h" +#include "src/tint/sem/evaluation_stage.h" #include "src/tint/sem/node.h" // Forward declarations @@ -28,18 +29,21 @@ class Variable; } // namespace tint::sem namespace tint::sem { + /// Expression holds the semantic information for expression nodes. class Expression : public Castable { public: /// Constructor /// @param declaration the AST node /// @param type the resolved type of the expression + /// @param stage the earliest evaluation stage for the expression /// @param statement the statement that owns this expression /// @param constant the constant value of the expression. May be null /// @param has_side_effects true if this expression may have side-effects /// @param source_var the (optional) source variable for this expression Expression(const ast::Expression* declaration, const sem::Type* type, + EvaluationStage stage, const Statement* statement, const Constant* constant, bool has_side_effects, @@ -54,6 +58,9 @@ class Expression : public Castable { /// @return the resolved type of the expression const sem::Type* Type() const { return type_; } + /// @return the earliest evaluation stage for the expression + EvaluationStage Stage() const { return stage_; } + /// @return the statement that owns this expression const Statement* Stmt() const { return statement_; } @@ -87,6 +94,7 @@ class Expression : public Castable { private: const sem::Type* const type_; + const EvaluationStage stage_; const Statement* const statement_; const Constant* const constant_; sem::Behaviors behaviors_{sem::Behavior::kNext}; diff --git a/src/tint/sem/expression_test.cc b/src/tint/sem/expression_test.cc index cc4bf0e3a8..63e01fbce0 100644 --- a/src/tint/sem/expression_test.cc +++ b/src/tint/sem/expression_test.cc @@ -43,7 +43,8 @@ using ExpressionTest = TestHelper; TEST_F(ExpressionTest, UnwrapMaterialize) { MockConstant c(create()); - auto* a = create(/* declaration */ nullptr, create(), /* statement */ nullptr, + auto* a = create(/* declaration */ nullptr, create(), + sem::EvaluationStage::kRuntime, /* statement */ nullptr, /* constant_value */ nullptr, /* has_side_effects */ false, /* source_var */ nullptr); auto* b = create(a, /* statement */ nullptr, &c); diff --git a/src/tint/sem/function.cc b/src/tint/sem/function.cc index cd34eb737d..dcc80b03b5 100644 --- a/src/tint/sem/function.cc +++ b/src/tint/sem/function.cc @@ -30,7 +30,7 @@ namespace tint::sem { Function::Function(const ast::Function* declaration, Type* return_type, std::vector parameters) - : Base(return_type, utils::ToConstPtrVec(parameters)), + : Base(return_type, utils::ToConstPtrVec(parameters), EvaluationStage::kRuntime), declaration_(declaration), workgroup_size_{WorkgroupDimension{1}, WorkgroupDimension{1}, WorkgroupDimension{1}} { for (auto* parameter : parameters) { diff --git a/src/tint/sem/index_accessor_expression.cc b/src/tint/sem/index_accessor_expression.cc index cd74201cef..70d1d5f89c 100644 --- a/src/tint/sem/index_accessor_expression.cc +++ b/src/tint/sem/index_accessor_expression.cc @@ -24,13 +24,14 @@ namespace tint::sem { IndexAccessorExpression::IndexAccessorExpression(const ast::IndexAccessorExpression* declaration, const sem::Type* type, + EvaluationStage stage, const Expression* object, const Expression* index, const Statement* statement, const Constant* constant, bool has_side_effects, const Variable* source_var /* = nullptr */) - : Base(declaration, type, statement, constant, has_side_effects, source_var), + : Base(declaration, type, stage, statement, constant, has_side_effects, source_var), object_(object), index_(index) {} diff --git a/src/tint/sem/index_accessor_expression.h b/src/tint/sem/index_accessor_expression.h index 3ba10eadae..ea93df72ed 100644 --- a/src/tint/sem/index_accessor_expression.h +++ b/src/tint/sem/index_accessor_expression.h @@ -32,6 +32,7 @@ class IndexAccessorExpression final : public CastableDeclaration(), /* type */ constant->Type(), + /* stage */ EvaluationStage::kConstant, // Abstract can only be const-expr /* statement */ statement, /* constant */ constant, /* has_side_effects */ false, diff --git a/src/tint/sem/member_accessor_expression.cc b/src/tint/sem/member_accessor_expression.cc index 7c31b0e1e3..74b4833c63 100644 --- a/src/tint/sem/member_accessor_expression.cc +++ b/src/tint/sem/member_accessor_expression.cc @@ -25,12 +25,14 @@ namespace tint::sem { MemberAccessorExpression::MemberAccessorExpression(const ast::MemberAccessorExpression* declaration, const sem::Type* type, + EvaluationStage stage, const Statement* statement, const Constant* constant, const Expression* object, bool has_side_effects, const Variable* source_var /* = nullptr */) - : Base(declaration, type, statement, constant, has_side_effects, source_var), object_(object) {} + : Base(declaration, type, stage, statement, constant, has_side_effects, source_var), + object_(object) {} MemberAccessorExpression::~MemberAccessorExpression() = default; @@ -42,7 +44,14 @@ StructMemberAccess::StructMemberAccess(const ast::MemberAccessorExpression* decl const StructMember* member, bool has_side_effects, const Variable* source_var /* = nullptr */) - : Base(declaration, type, statement, constant, object, has_side_effects, source_var), + : Base(declaration, + type, + object->Stage(), + statement, + constant, + object, + has_side_effects, + source_var), member_(member) {} StructMemberAccess::~StructMemberAccess() = default; @@ -55,7 +64,14 @@ Swizzle::Swizzle(const ast::MemberAccessorExpression* declaration, std::vector indices, bool has_side_effects, const Variable* source_var /* = nullptr */) - : Base(declaration, type, statement, constant, object, has_side_effects, source_var), + : Base(declaration, + type, + object->Stage(), + statement, + constant, + object, + has_side_effects, + source_var), indices_(std::move(indices)) {} Swizzle::~Swizzle() = default; diff --git a/src/tint/sem/member_accessor_expression.h b/src/tint/sem/member_accessor_expression.h index d8484a8c7e..43e1466218 100644 --- a/src/tint/sem/member_accessor_expression.h +++ b/src/tint/sem/member_accessor_expression.h @@ -37,6 +37,7 @@ class MemberAccessorExpression : public Castable { /// Constructor /// @param type the type that's being constructed /// @param parameters the type constructor parameters - TypeConstructor(const sem::Type* type, const ParameterList& parameters); + /// @param stage the earliest evaluation stage for the expression + TypeConstructor(const sem::Type* type, const ParameterList& parameters, EvaluationStage stage); /// Destructor ~TypeConstructor() override; diff --git a/src/tint/sem/type_conversion.cc b/src/tint/sem/type_conversion.cc index 5da2928582..262e3a0792 100644 --- a/src/tint/sem/type_conversion.cc +++ b/src/tint/sem/type_conversion.cc @@ -18,8 +18,10 @@ TINT_INSTANTIATE_TYPEINFO(tint::sem::TypeConversion); namespace tint::sem { -TypeConversion::TypeConversion(const sem::Type* type, const sem::Parameter* parameter) - : Base(type, ParameterList{parameter}) {} +TypeConversion::TypeConversion(const sem::Type* type, + const sem::Parameter* parameter, + EvaluationStage stage) + : Base(type, ParameterList{parameter}, stage) {} TypeConversion::~TypeConversion() = default; diff --git a/src/tint/sem/type_conversion.h b/src/tint/sem/type_conversion.h index e40056508e..5584b36fe3 100644 --- a/src/tint/sem/type_conversion.h +++ b/src/tint/sem/type_conversion.h @@ -25,7 +25,8 @@ class TypeConversion final : public Castable { /// Constructor /// @param type the target type of the cast /// @param parameter the type cast parameter - TypeConversion(const sem::Type* type, const sem::Parameter* parameter); + /// @param stage the earliest evaluation stage for the expression + TypeConversion(const sem::Type* type, const sem::Parameter* parameter, EvaluationStage stage); /// Destructor ~TypeConversion() override; diff --git a/src/tint/sem/variable.cc b/src/tint/sem/variable.cc index 38498073e7..28a373b613 100644 --- a/src/tint/sem/variable.cc +++ b/src/tint/sem/variable.cc @@ -28,14 +28,15 @@ TINT_INSTANTIATE_TYPEINFO(tint::sem::Parameter); TINT_INSTANTIATE_TYPEINFO(tint::sem::VariableUser); namespace tint::sem { - Variable::Variable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const Constant* constant_value) : declaration_(declaration), type_(type), + stage_(stage), storage_class_(storage_class), access_(access), constant_value_(constant_value) {} @@ -44,21 +45,24 @@ Variable::~Variable() = default; LocalVariable::LocalVariable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const sem::Statement* statement, const Constant* constant_value) - : Base(declaration, type, storage_class, access, constant_value), statement_(statement) {} + : Base(declaration, type, stage, storage_class, access, constant_value), + statement_(statement) {} LocalVariable::~LocalVariable() = default; GlobalVariable::GlobalVariable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const Constant* constant_value, sem::BindingPoint binding_point) - : Base(declaration, type, storage_class, access, constant_value), + : Base(declaration, type, stage, storage_class, access, constant_value), binding_point_(binding_point) {} GlobalVariable::~GlobalVariable() = default; @@ -69,7 +73,9 @@ Parameter::Parameter(const ast::Parameter* declaration, ast::StorageClass storage_class, ast::Access access, const ParameterUsage usage /* = ParameterUsage::kNone */) - : Base(declaration, type, storage_class, access, nullptr), index_(index), usage_(usage) {} + : Base(declaration, type, EvaluationStage::kRuntime, storage_class, access, nullptr), + index_(index), + usage_(usage) {} Parameter::~Parameter() = default; @@ -78,6 +84,7 @@ VariableUser::VariableUser(const ast::IdentifierExpression* declaration, sem::Variable* variable) : Base(declaration, variable->Type(), + variable->Stage(), statement, variable->ConstantValue(), /* has_side_effects */ false), diff --git a/src/tint/sem/variable.h b/src/tint/sem/variable.h index e5a5ceccb4..fe54f507e1 100644 --- a/src/tint/sem/variable.h +++ b/src/tint/sem/variable.h @@ -45,11 +45,13 @@ class Variable : public Castable { /// Constructor /// @param declaration the AST declaration node /// @param type the variable type + /// @param stage the evaluation stage for an expression of this variable type /// @param storage_class the variable storage class /// @param access the variable access control type /// @param constant_value the constant value for the variable. May be null Variable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const Constant* constant_value); @@ -63,6 +65,9 @@ class Variable : public Castable { /// @returns the canonical type for the variable const sem::Type* Type() const { return type_; } + /// @returns the evaluation stage for an expression of this variable type + EvaluationStage Stage() const { return stage_; } + /// @returns the storage class for the variable ast::StorageClass StorageClass() const { return storage_class_; } @@ -89,6 +94,7 @@ class Variable : public Castable { private: const ast::Variable* const declaration_; const sem::Type* const type_; + const EvaluationStage stage_; const ast::StorageClass storage_class_; const ast::Access access_; const Constant* constant_value_; @@ -102,12 +108,14 @@ class LocalVariable final : public Castable { /// Constructor /// @param declaration the AST declaration node /// @param type the variable type + /// @param stage the evaluation stage for an expression of this variable type /// @param storage_class the variable storage class /// @param access the variable access control type /// @param statement the statement that declared this local variable /// @param constant_value the constant value for the variable. May be null LocalVariable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const sem::Statement* statement, @@ -137,12 +145,14 @@ class GlobalVariable final : public Castable { /// Constructor /// @param declaration the AST declaration node /// @param type the variable type + /// @param stage the evaluation stage for an expression of this variable type /// @param storage_class the variable storage class /// @param access the variable access control type /// @param constant_value the constant value for the variable. May be null /// @param binding_point the optional resource binding point of the variable GlobalVariable(const ast::Variable* declaration, const sem::Type* type, + EvaluationStage stage, ast::StorageClass storage_class, ast::Access access, const Constant* constant_value, diff --git a/src/tint/writer/append_vector.cc b/src/tint/writer/append_vector.cc index c5f184d4b5..28845d9f61 100644 --- a/src/tint/writer/append_vector.cc +++ b/src/tint/writer/append_vector.cc @@ -59,7 +59,8 @@ const sem::Expression* Zero(ProgramBuilder& b, const sem::Type* ty, const sem::S << "unsupported vector element type: " << ty->TypeInfo().name; return nullptr; } - auto* sem = b.create(expr, ty, stmt, /* constant_value */ nullptr, + auto* sem = b.create(expr, ty, sem::EvaluationStage::kRuntime, stmt, + /* constant_value */ nullptr, /* has_side_effects */ false); b.Sem().Add(expr, sem); return sem; @@ -136,10 +137,12 @@ const sem::Call* AppendVector(ProgramBuilder* b, auto* scalar_cast_target = b->create( packed_el_sem_ty, b->create(nullptr, 0u, scalar_sem->Type()->UnwrapRef(), - ast::StorageClass::kNone, ast::Access::kUndefined)); + ast::StorageClass::kNone, ast::Access::kUndefined), + sem::EvaluationStage::kRuntime); auto* scalar_cast_sem = b->create( - scalar_cast_ast, scalar_cast_target, std::vector{scalar_sem}, - statement, /* constant_value */ nullptr, /* has_side_effects */ false); + scalar_cast_ast, scalar_cast_target, sem::EvaluationStage::kRuntime, + std::vector{scalar_sem}, statement, + /* constant_value */ nullptr, /* has_side_effects */ false); b->Sem().Add(scalar_cast_ast, scalar_cast_sem); packed.emplace_back(scalar_cast_sem); } else { @@ -151,15 +154,17 @@ const sem::Call* AppendVector(ProgramBuilder* b, utils::Transform(packed, [&](const sem::Expression* expr) { return expr->Declaration(); })); auto* constructor_target = b->create( packed_sem_ty, - utils::Transform( - packed, [&](const tint::sem::Expression* arg, size_t i) -> const sem::Parameter* { - return b->create(nullptr, static_cast(i), - arg->Type()->UnwrapRef(), ast::StorageClass::kNone, - ast::Access::kUndefined); - })); - auto* constructor_sem = b->create(constructor_ast, constructor_target, packed, - statement, /* constant_value */ nullptr, - /* has_side_effects */ false); + utils::Transform(packed, + [&](const tint::sem::Expression* arg, size_t i) -> const sem::Parameter* { + return b->create( + nullptr, static_cast(i), arg->Type()->UnwrapRef(), + ast::StorageClass::kNone, ast::Access::kUndefined); + }), + sem::EvaluationStage::kRuntime); + auto* constructor_sem = + b->create(constructor_ast, constructor_target, sem::EvaluationStage::kRuntime, + packed, statement, /* constant_value */ nullptr, + /* has_side_effects */ false); b->Sem().Add(constructor_ast, constructor_sem); return constructor_sem; } diff --git a/src/tint/writer/glsl/generator_impl.cc b/src/tint/writer/glsl/generator_impl.cc index 1a2b5dc29b..d28f3f491a 100644 --- a/src/tint/writer/glsl/generator_impl.cc +++ b/src/tint/writer/glsl/generator_impl.cc @@ -1338,7 +1338,8 @@ bool GeneratorImpl::EmitBarrierCall(std::ostream& out, const sem::Builtin* built const ast::Expression* GeneratorImpl::CreateF32Zero(const sem::Statement* stmt) { auto* zero = builder_.Expr(0_f); auto* f32 = builder_.create(); - auto* sem_zero = builder_.create(zero, f32, stmt, /* constant_value */ nullptr, + auto* sem_zero = builder_.create(zero, f32, sem::EvaluationStage::kRuntime, + stmt, /* constant_value */ nullptr, /* has_side_effects */ false); builder_.Sem().Add(zero, sem_zero); return zero; diff --git a/src/tint/writer/hlsl/generator_impl.cc b/src/tint/writer/hlsl/generator_impl.cc index 0a6a196801..e8ab2c7c55 100644 --- a/src/tint/writer/hlsl/generator_impl.cc +++ b/src/tint/writer/hlsl/generator_impl.cc @@ -2392,7 +2392,8 @@ bool GeneratorImpl::EmitTextureCall(std::ostream& out, auto* zero = builder_.Expr(0_i); auto* stmt = builder_.Sem().Get(vector)->Stmt(); builder_.Sem().Add( - zero, builder_.create(zero, i32, stmt, /* constant_value */ nullptr, + zero, builder_.create(zero, i32, sem::EvaluationStage::kRuntime, stmt, + /* constant_value */ nullptr, /* has_side_effects */ false)); auto* packed = AppendVector(&builder_, vector, zero); return EmitExpression(out, packed->Declaration());