1 // Copyright 2021 The Tint Authors.
2 //
3 // Licensed under the Apache License, Version 2.0 (the "License");
4 // you may not use this file except in compliance with the License.
5 // You may obtain a copy of the License at
6 //
7 // http://www.apache.org/licenses/LICENSE-2.0
8 //
9 // Unless required by applicable law or agreed to in writing, software
10 // distributed under the License is distributed on an "AS IS" BASIS,
11 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
12 // See the License for the specific language governing permissions and
13 // limitations under the License.
14
15 #include "src/ast/intrinsic_texture_helper_test.h"
16 #include "src/resolver/resolver_test_helper.h"
17
18 namespace tint {
19 namespace resolver {
20 namespace {
21
22 using ResolverIntrinsicValidationTest = ResolverTest;
23
TEST_F(ResolverIntrinsicValidationTest,FunctionTypeMustMatchReturnStatementType_void_fail)24 TEST_F(ResolverIntrinsicValidationTest,
25 FunctionTypeMustMatchReturnStatementType_void_fail) {
26 // fn func { return workgroupBarrier(); }
27 Func("func", {}, ty.void_(),
28 {
29 Return(Call(Source{Source::Location{12, 34}}, "workgroupBarrier")),
30 });
31
32 EXPECT_FALSE(r()->Resolve());
33 EXPECT_EQ(
34 r()->error(),
35 "12:34 error: intrinsic 'workgroupBarrier' does not return a value");
36 }
37
TEST_F(ResolverIntrinsicValidationTest,InvalidPipelineStageDirect)38 TEST_F(ResolverIntrinsicValidationTest, InvalidPipelineStageDirect) {
39 // [[stage(compute), workgroup_size(1)]] fn func { return dpdx(1.0); }
40
41 auto* dpdx = create<ast::CallExpression>(Source{{3, 4}}, Expr("dpdx"),
42 ast::ExpressionList{Expr(1.0f)});
43 Func(Source{{1, 2}}, "func", ast::VariableList{}, ty.void_(),
44 {CallStmt(dpdx)},
45 {Stage(ast::PipelineStage::kCompute), WorkgroupSize(1)});
46
47 EXPECT_FALSE(r()->Resolve());
48 EXPECT_EQ(r()->error(),
49 "3:4 error: built-in cannot be used by compute pipeline stage");
50 }
51
TEST_F(ResolverIntrinsicValidationTest,InvalidPipelineStageIndirect)52 TEST_F(ResolverIntrinsicValidationTest, InvalidPipelineStageIndirect) {
53 // fn f0 { return dpdx(1.0); }
54 // fn f1 { f0(); }
55 // fn f2 { f1(); }
56 // [[stage(compute), workgroup_size(1)]] fn main { return f2(); }
57
58 auto* dpdx = create<ast::CallExpression>(Source{{3, 4}}, Expr("dpdx"),
59 ast::ExpressionList{Expr(1.0f)});
60 Func(Source{{1, 2}}, "f0", {}, ty.void_(), {CallStmt(dpdx)});
61
62 Func(Source{{3, 4}}, "f1", {}, ty.void_(), {CallStmt(Call("f0"))});
63
64 Func(Source{{5, 6}}, "f2", {}, ty.void_(), {CallStmt(Call("f1"))});
65
66 Func(Source{{7, 8}}, "main", {}, ty.void_(), {CallStmt(Call("f2"))},
67 {Stage(ast::PipelineStage::kCompute), WorkgroupSize(1)});
68
69 EXPECT_FALSE(r()->Resolve());
70 EXPECT_EQ(r()->error(),
71 R"(3:4 error: built-in cannot be used by compute pipeline stage
72 1:2 note: called by function 'f0'
73 3:4 note: called by function 'f1'
74 5:6 note: called by function 'f2'
75 7:8 note: called by entry point 'main')");
76 }
77
TEST_F(ResolverIntrinsicValidationTest,IntrinsicRedeclaredAsFunction)78 TEST_F(ResolverIntrinsicValidationTest, IntrinsicRedeclaredAsFunction) {
79 Func(Source{{12, 34}}, "mix", {}, ty.i32(), {});
80
81 EXPECT_FALSE(r()->Resolve());
82 EXPECT_EQ(
83 r()->error(),
84 R"(12:34 error: 'mix' is a builtin and cannot be redeclared as a function)");
85 }
86
TEST_F(ResolverIntrinsicValidationTest,IntrinsicRedeclaredAsGlobalLet)87 TEST_F(ResolverIntrinsicValidationTest, IntrinsicRedeclaredAsGlobalLet) {
88 GlobalConst(Source{{12, 34}}, "mix", ty.i32(), Expr(1));
89
90 EXPECT_FALSE(r()->Resolve());
91 EXPECT_EQ(
92 r()->error(),
93 R"(12:34 error: 'mix' is a builtin and cannot be redeclared as a module-scope let)");
94 }
95
TEST_F(ResolverIntrinsicValidationTest,IntrinsicRedeclaredAsGlobalVar)96 TEST_F(ResolverIntrinsicValidationTest, IntrinsicRedeclaredAsGlobalVar) {
97 Global(Source{{12, 34}}, "mix", ty.i32(), Expr(1),
98 ast::StorageClass::kPrivate);
99
100 EXPECT_FALSE(r()->Resolve());
101 EXPECT_EQ(
102 r()->error(),
103 R"(12:34 error: 'mix' is a builtin and cannot be redeclared as a module-scope var)");
104 }
105
TEST_F(ResolverIntrinsicValidationTest,IntrinsicRedeclaredAsAlias)106 TEST_F(ResolverIntrinsicValidationTest, IntrinsicRedeclaredAsAlias) {
107 Alias(Source{{12, 34}}, "mix", ty.i32());
108
109 EXPECT_FALSE(r()->Resolve());
110 EXPECT_EQ(
111 r()->error(),
112 R"(12:34 error: 'mix' is a builtin and cannot be redeclared as an alias)");
113 }
114
TEST_F(ResolverIntrinsicValidationTest,IntrinsicRedeclaredAsStruct)115 TEST_F(ResolverIntrinsicValidationTest, IntrinsicRedeclaredAsStruct) {
116 Structure(Source{{12, 34}}, "mix", {Member("m", ty.i32())});
117
118 EXPECT_FALSE(r()->Resolve());
119 EXPECT_EQ(
120 r()->error(),
121 R"(12:34 error: 'mix' is a builtin and cannot be redeclared as a struct)");
122 }
123
124 namespace texture_constexpr_args {
125
126 using TextureOverloadCase = ast::intrinsic::test::TextureOverloadCase;
127 using ValidTextureOverload = ast::intrinsic::test::ValidTextureOverload;
128 using TextureKind = ast::intrinsic::test::TextureKind;
129 using TextureDataType = ast::intrinsic::test::TextureDataType;
130 using u32 = ProgramBuilder::u32;
131 using i32 = ProgramBuilder::i32;
132 using f32 = ProgramBuilder::f32;
133
TextureCases(std::unordered_set<ValidTextureOverload> overloads)134 static std::vector<TextureOverloadCase> TextureCases(
135 std::unordered_set<ValidTextureOverload> overloads) {
136 std::vector<TextureOverloadCase> cases;
137 for (auto c : TextureOverloadCase::ValidCases()) {
138 if (overloads.count(c.overload)) {
139 cases.push_back(c);
140 }
141 }
142 return cases;
143 }
144
145 enum class Position {
146 kFirst,
147 kLast,
148 };
149
150 struct Parameter {
151 const char* const name;
152 const Position position;
153 int min;
154 int max;
155 };
156
157 class Constexpr {
158 public:
159 enum class Kind {
160 kScalar,
161 kVec2,
162 kVec3,
163 kVec3_Scalar_Vec2,
164 kVec3_Vec2_Scalar,
165 kEmptyVec2,
166 kEmptyVec3,
167 };
168
Constexpr(int32_t invalid_idx,Kind k,int32_t x=0,int32_t y=0,int32_t z=0)169 Constexpr(int32_t invalid_idx,
170 Kind k,
171 int32_t x = 0,
172 int32_t y = 0,
173 int32_t z = 0)
174 : invalid_index(invalid_idx), kind(k), values{x, y, z} {}
175
operator ()(Source src,ProgramBuilder & b)176 const ast::Expression* operator()(Source src, ProgramBuilder& b) {
177 switch (kind) {
178 case Kind::kScalar:
179 return b.Expr(src, values[0]);
180 case Kind::kVec2:
181 return b.Construct(src, b.ty.vec2<i32>(), values[0], values[1]);
182 case Kind::kVec3:
183 return b.Construct(src, b.ty.vec3<i32>(), values[0], values[1],
184 values[2]);
185 case Kind::kVec3_Scalar_Vec2:
186 return b.Construct(src, b.ty.vec3<i32>(), values[0],
187 b.vec2<i32>(values[1], values[2]));
188 case Kind::kVec3_Vec2_Scalar:
189 return b.Construct(src, b.ty.vec3<i32>(),
190 b.vec2<i32>(values[0], values[1]), values[2]);
191 case Kind::kEmptyVec2:
192 return b.Construct(src, b.ty.vec2<i32>());
193 case Kind::kEmptyVec3:
194 return b.Construct(src, b.ty.vec3<i32>());
195 }
196 return nullptr;
197 }
198
199 static const constexpr int32_t kValid = -1;
200 const int32_t invalid_index; // Expected error value, or kValid
201 const Kind kind;
202 const std::array<int32_t, 3> values;
203 };
204
operator <<(std::ostream & out,Parameter param)205 static std::ostream& operator<<(std::ostream& out, Parameter param) {
206 return out << param.name;
207 }
208
operator <<(std::ostream & out,Constexpr expr)209 static std::ostream& operator<<(std::ostream& out, Constexpr expr) {
210 switch (expr.kind) {
211 case Constexpr::Kind::kScalar:
212 return out << expr.values[0];
213 case Constexpr::Kind::kVec2:
214 return out << "vec2(" << expr.values[0] << ", " << expr.values[1] << ")";
215 case Constexpr::Kind::kVec3:
216 return out << "vec3(" << expr.values[0] << ", " << expr.values[1] << ", "
217 << expr.values[2] << ")";
218 case Constexpr::Kind::kVec3_Scalar_Vec2:
219 return out << "vec3(" << expr.values[0] << ", vec2(" << expr.values[1]
220 << ", " << expr.values[2] << "))";
221 case Constexpr::Kind::kVec3_Vec2_Scalar:
222 return out << "vec3(vec2(" << expr.values[0] << ", " << expr.values[1]
223 << "), " << expr.values[2] << ")";
224 case Constexpr::Kind::kEmptyVec2:
225 return out << "vec2()";
226 case Constexpr::Kind::kEmptyVec3:
227 return out << "vec3()";
228 }
229 return out;
230 }
231
232 using IntrinsicTextureConstExprArgValidationTest = ResolverTestWithParam<
233 std::tuple<TextureOverloadCase, Parameter, Constexpr>>;
234
TEST_P(IntrinsicTextureConstExprArgValidationTest,Immediate)235 TEST_P(IntrinsicTextureConstExprArgValidationTest, Immediate) {
236 auto& p = GetParam();
237 auto overload = std::get<0>(p);
238 auto param = std::get<1>(p);
239 auto expr = std::get<2>(p);
240
241 overload.BuildTextureVariable(this);
242 overload.BuildSamplerVariable(this);
243
244 auto args = overload.args(this);
245 auto*& arg_to_replace =
246 (param.position == Position::kFirst) ? args.front() : args.back();
247
248 // BuildTextureVariable() uses a Literal for scalars, and a CallExpression for
249 // a vector constructor.
250 bool is_vector = arg_to_replace->Is<ast::CallExpression>();
251
252 // Make the expression to be replaced, reachable. This keeps the resolver
253 // happy.
254 WrapInFunction(arg_to_replace);
255
256 arg_to_replace = expr(Source{{12, 34}}, *this);
257
258 // Call the intrinsic with the constexpr argument replaced
259 Func("func", {}, ty.void_(), {CallStmt(Call(overload.function, args))},
260 {Stage(ast::PipelineStage::kFragment)});
261
262 if (expr.invalid_index == Constexpr::kValid) {
263 EXPECT_TRUE(r()->Resolve()) << r()->error();
264 } else {
265 EXPECT_FALSE(r()->Resolve());
266 std::stringstream err;
267 if (is_vector) {
268 err << "12:34 error: each component of the " << param.name
269 << " argument must be at least " << param.min << " and at most "
270 << param.max << ". " << param.name << " component "
271 << expr.invalid_index << " is "
272 << std::to_string(expr.values[expr.invalid_index]);
273 } else {
274 err << "12:34 error: the " << param.name << " argument must be at least "
275 << param.min << " and at most " << param.max << ". " << param.name
276 << " is " << std::to_string(expr.values[expr.invalid_index]);
277 }
278 EXPECT_EQ(r()->error(), err.str());
279 }
280 }
281
TEST_P(IntrinsicTextureConstExprArgValidationTest,GlobalConst)282 TEST_P(IntrinsicTextureConstExprArgValidationTest, GlobalConst) {
283 auto& p = GetParam();
284 auto overload = std::get<0>(p);
285 auto param = std::get<1>(p);
286 auto expr = std::get<2>(p);
287
288 // Build the global texture and sampler variables
289 overload.BuildTextureVariable(this);
290 overload.BuildSamplerVariable(this);
291
292 // Build the module-scope let 'G' with the offset value
293 GlobalConst("G", nullptr, expr({}, *this));
294
295 auto args = overload.args(this);
296 auto*& arg_to_replace =
297 (param.position == Position::kFirst) ? args.front() : args.back();
298
299 // Make the expression to be replaced, reachable. This keeps the resolver
300 // happy.
301 WrapInFunction(arg_to_replace);
302
303 arg_to_replace = Expr(Source{{12, 34}}, "G");
304
305 // Call the intrinsic with the constexpr argument replaced
306 Func("func", {}, ty.void_(), {CallStmt(Call(overload.function, args))},
307 {Stage(ast::PipelineStage::kFragment)});
308
309 EXPECT_FALSE(r()->Resolve());
310 std::stringstream err;
311 err << "12:34 error: the " << param.name
312 << " argument must be a const_expression";
313 EXPECT_EQ(r()->error(), err.str());
314 }
315
316 INSTANTIATE_TEST_SUITE_P(
317 Offset2D,
318 IntrinsicTextureConstExprArgValidationTest,
319 testing::Combine(
320 testing::ValuesIn(TextureCases({
321 ValidTextureOverload::kSample2dOffsetF32,
322 ValidTextureOverload::kSample2dArrayOffsetF32,
323 ValidTextureOverload::kSampleDepth2dOffsetF32,
324 ValidTextureOverload::kSampleDepth2dArrayOffsetF32,
325 ValidTextureOverload::kSampleBias2dOffsetF32,
326 ValidTextureOverload::kSampleBias2dArrayOffsetF32,
327 ValidTextureOverload::kSampleLevel2dOffsetF32,
328 ValidTextureOverload::kSampleLevel2dArrayOffsetF32,
329 ValidTextureOverload::kSampleLevelDepth2dOffsetF32,
330 ValidTextureOverload::kSampleLevelDepth2dArrayOffsetF32,
331 ValidTextureOverload::kSampleGrad2dOffsetF32,
332 ValidTextureOverload::kSampleGrad2dArrayOffsetF32,
333 ValidTextureOverload::kSampleCompareDepth2dOffsetF32,
334 ValidTextureOverload::kSampleCompareDepth2dArrayOffsetF32,
335 ValidTextureOverload::kSampleCompareLevelDepth2dOffsetF32,
336 ValidTextureOverload::kSampleCompareLevelDepth2dArrayOffsetF32,
337 })),
338 testing::Values(Parameter{"offset", Position::kLast, -8, 7}),
339 testing::Values(
340 Constexpr{Constexpr::kValid, Constexpr::Kind::kEmptyVec2},
341 Constexpr{Constexpr::kValid, Constexpr::Kind::kVec2, -1, 1},
342 Constexpr{Constexpr::kValid, Constexpr::Kind::kVec2, 7, -8},
343 Constexpr{0, Constexpr::Kind::kVec2, 8, 0},
344 Constexpr{1, Constexpr::Kind::kVec2, 0, 8},
345 Constexpr{0, Constexpr::Kind::kVec2, -9, 0},
346 Constexpr{1, Constexpr::Kind::kVec2, 0, -9},
347 Constexpr{0, Constexpr::Kind::kVec2, 8, 8},
348 Constexpr{0, Constexpr::Kind::kVec2, -9, -9})));
349
350 INSTANTIATE_TEST_SUITE_P(
351 Offset3D,
352 IntrinsicTextureConstExprArgValidationTest,
353 testing::Combine(
354 testing::ValuesIn(TextureCases({
355 ValidTextureOverload::kSample3dOffsetF32,
356 ValidTextureOverload::kSampleBias3dOffsetF32,
357 ValidTextureOverload::kSampleLevel3dOffsetF32,
358 ValidTextureOverload::kSampleGrad3dOffsetF32,
359 })),
360 testing::Values(Parameter{"offset", Position::kLast, -8, 7}),
361 testing::Values(
362 Constexpr{Constexpr::kValid, Constexpr::Kind::kEmptyVec3},
363 Constexpr{Constexpr::kValid, Constexpr::Kind::kVec3, 0, 0, 0},
364 Constexpr{Constexpr::kValid, Constexpr::Kind::kVec3, 7, -8, 7},
365 Constexpr{0, Constexpr::Kind::kVec3, 10, 0, 0},
366 Constexpr{1, Constexpr::Kind::kVec3, 0, 10, 0},
367 Constexpr{2, Constexpr::Kind::kVec3, 0, 0, 10},
368 Constexpr{0, Constexpr::Kind::kVec3, 10, 11, 12},
369 Constexpr{0, Constexpr::Kind::kVec3_Scalar_Vec2, 10, 0, 0},
370 Constexpr{1, Constexpr::Kind::kVec3_Scalar_Vec2, 0, 10, 0},
371 Constexpr{2, Constexpr::Kind::kVec3_Scalar_Vec2, 0, 0, 10},
372 Constexpr{0, Constexpr::Kind::kVec3_Scalar_Vec2, 10, 11, 12},
373 Constexpr{0, Constexpr::Kind::kVec3_Vec2_Scalar, 10, 0, 0},
374 Constexpr{1, Constexpr::Kind::kVec3_Vec2_Scalar, 0, 10, 0},
375 Constexpr{2, Constexpr::Kind::kVec3_Vec2_Scalar, 0, 0, 10},
376 Constexpr{0, Constexpr::Kind::kVec3_Vec2_Scalar, 10, 11, 12})));
377
378 INSTANTIATE_TEST_SUITE_P(
379 Component,
380 IntrinsicTextureConstExprArgValidationTest,
381 testing::Combine(
382 testing::ValuesIn(
383 TextureCases({ValidTextureOverload::kGather2dF32,
384 ValidTextureOverload::kGather2dOffsetF32,
385 ValidTextureOverload::kGather2dArrayF32,
386 ValidTextureOverload::kGather2dArrayOffsetF32,
387 ValidTextureOverload::kGatherCubeF32,
388 ValidTextureOverload::kGatherCubeArrayF32})),
389 testing::Values(Parameter{"component", Position::kFirst, 0, 3}),
390 testing::Values(
391 Constexpr{Constexpr::kValid, Constexpr::Kind::kScalar, 0},
392 Constexpr{Constexpr::kValid, Constexpr::Kind::kScalar, 1},
393 Constexpr{Constexpr::kValid, Constexpr::Kind::kScalar, 2},
394 Constexpr{Constexpr::kValid, Constexpr::Kind::kScalar, 3},
395 Constexpr{0, Constexpr::Kind::kScalar, 4},
396 Constexpr{0, Constexpr::Kind::kScalar, 123},
397 Constexpr{0, Constexpr::Kind::kScalar, -1})));
398
399 } // namespace texture_constexpr_args
400
401 } // namespace
402 } // namespace resolver
403 } // namespace tint
404