// Copyright 2020 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/ast/stage_decoration.h" #include "src/ast/struct_block_decoration.h" #include "src/ast/variable_decl_statement.h" #include "src/ast/workgroup_decoration.h" #include "src/type/access_control_type.h" #include "src/writer/wgsl/test_helper.h" namespace tint { namespace writer { namespace wgsl { namespace { using WgslGeneratorImplTest = TestHelper; TEST_F(WgslGeneratorImplTest, Emit_Function) { auto* func = Func("my_func", ast::VariableList{}, ty.void_(), ast::StatementList{ create(), create(), }, ast::DecorationList{}); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( fn my_func() { discard; return; } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_WithParams) { auto* func = Func( "my_func", ast::VariableList{Param("a", ty.f32()), Param("b", ty.i32())}, ty.void_(), ast::StatementList{ create(), create(), }, ast::DecorationList{}); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( fn my_func(a : f32, b : i32) { discard; return; } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_WithDecoration_WorkgroupSize) { auto* func = Func("my_func", ast::VariableList{}, ty.void_(), ast::StatementList{ create(), create(), }, ast::DecorationList{ create(2u, 4u, 6u), }); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( [[workgroup_size(2, 4, 6)]] fn my_func() { discard; return; } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_WithDecoration_Stage) { auto* func = Func("my_func", ast::VariableList{}, ty.void_(), ast::StatementList{ create(), create(), }, ast::DecorationList{ create(ast::PipelineStage::kFragment), }); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( [[stage(fragment)]] fn my_func() { discard; return; } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_WithDecoration_Multiple) { auto* func = Func("my_func", ast::VariableList{}, ty.void_(), ast::StatementList{ create(), create(), }, ast::DecorationList{ create(ast::PipelineStage::kFragment), create(2u, 4u, 6u), }); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( [[stage(fragment)]] [[workgroup_size(2, 4, 6)]] fn my_func() { discard; return; } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_EntryPoint_Parameters) { auto* vec4 = ty.vec4(); auto* coord = Param( "coord", vec4, {create(ast::Builtin::kPosition)}); auto* loc1 = Param("loc1", ty.f32(), {create(1u)}); auto* func = Func("frag_main", ast::VariableList{coord, loc1}, ty.void_(), ast::StatementList{}, ast::DecorationList{ create(ast::PipelineStage::kFragment), }); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( [[stage(fragment)]] fn frag_main([[builtin(position)]] coord : vec4, [[location(1)]] loc1 : f32) { } )"); } TEST_F(WgslGeneratorImplTest, Emit_Function_EntryPoint_ReturnValue) { auto* func = Func("frag_main", ast::VariableList{}, ty.f32(), ast::StatementList{ create(Expr(1.f)), }, ast::DecorationList{ create(ast::PipelineStage::kFragment), }, ast::DecorationList{ create(1u), }); GeneratorImpl& gen = Build(); gen.increment_indent(); ASSERT_TRUE(gen.EmitFunction(func)); EXPECT_EQ(gen.result(), R"( [[stage(fragment)]] fn frag_main() -> [[location(1)]] f32 { return 1.0; } )"); } // https://crbug.com/tint/297 TEST_F(WgslGeneratorImplTest, Emit_Function_Multiple_EntryPoint_With_Same_ModuleVar) { // [[block]] struct Data { // d : f32; // }; // [[binding(0), group(0)]] var data : Data; // // [[stage(compute)]] // fn a() { // return; // } // // [[stage(compute)]] // fn b() { // return; // } auto* s = Structure("Data", {Member("d", ty.f32())}, {create()}); type::AccessControl ac(ast::AccessControl::kReadWrite, s); Global("data", &ac, ast::StorageClass::kStorage, nullptr, ast::DecorationList{ create(0), create(0), }); { auto* var = Var("v", ty.f32(), ast::StorageClass::kFunction, create(Expr("data"), Expr("d"))); Func("a", ast::VariableList{}, ty.void_(), ast::StatementList{ create(var), create(), }, ast::DecorationList{ create(ast::PipelineStage::kCompute), }); } { auto* var = Var("v", ty.f32(), ast::StorageClass::kFunction, create(Expr("data"), Expr("d"))); Func("b", ast::VariableList{}, ty.void_(), ast::StatementList{ create(var), create(), }, ast::DecorationList{ create(ast::PipelineStage::kCompute), }); } GeneratorImpl& gen = Build(); ASSERT_TRUE(gen.Generate(nullptr)) << gen.error(); EXPECT_EQ(gen.result(), R"([[block]] struct Data { d : f32; }; [[binding(0), group(0)]] var data : [[access(read_write)]] Data; [[stage(compute)]] fn a() { var v : f32 = data.d; return; } [[stage(compute)]] fn b() { var v : f32 = data.d; return; } )"); } } // namespace } // namespace wgsl } // namespace writer } // namespace tint