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/transform/binding_remapper.h"
16
17 #include <utility>
18
19 #include "src/transform/test_helper.h"
20
21 namespace tint {
22 namespace transform {
23 namespace {
24
25 using BindingRemapperTest = TransformTest;
26
TEST_F(BindingRemapperTest,NoRemappings)27 TEST_F(BindingRemapperTest, NoRemappings) {
28 auto* src = R"(
29 [[block]]
30 struct S {
31 a : f32;
32 };
33
34 [[group(2), binding(1)]] var<storage, read> a : S;
35
36 [[group(3), binding(2)]] var<storage, read> b : S;
37
38 [[stage(compute), workgroup_size(1)]]
39 fn f() {
40 }
41 )";
42
43 auto* expect = src;
44
45 DataMap data;
46 data.Add<BindingRemapper::Remappings>(BindingRemapper::BindingPoints{},
47 BindingRemapper::AccessControls{});
48 auto got = Run<BindingRemapper>(src, data);
49
50 EXPECT_EQ(expect, str(got));
51 }
52
TEST_F(BindingRemapperTest,RemapBindingPoints)53 TEST_F(BindingRemapperTest, RemapBindingPoints) {
54 auto* src = R"(
55 [[block]]
56 struct S {
57 a : f32;
58 };
59
60 [[group(2), binding(1)]] var<storage, read> a : S;
61
62 [[group(3), binding(2)]] var<storage, read> b : S;
63
64 [[stage(compute), workgroup_size(1)]]
65 fn f() {
66 }
67 )";
68
69 auto* expect = R"(
70 [[block]]
71 struct S {
72 a : f32;
73 };
74
75 [[group(1), binding(2)]] var<storage, read> a : S;
76
77 [[group(3), binding(2)]] var<storage, read> b : S;
78
79 [[stage(compute), workgroup_size(1)]]
80 fn f() {
81 }
82 )";
83
84 DataMap data;
85 data.Add<BindingRemapper::Remappings>(
86 BindingRemapper::BindingPoints{
87 {{2, 1}, {1, 2}}, // Remap
88 {{4, 5}, {6, 7}}, // Not found
89 // Keep [[group(3), binding(2)]] as is
90 },
91 BindingRemapper::AccessControls{});
92 auto got = Run<BindingRemapper>(src, data);
93
94 EXPECT_EQ(expect, str(got));
95 }
96
TEST_F(BindingRemapperTest,RemapAccessControls)97 TEST_F(BindingRemapperTest, RemapAccessControls) {
98 auto* src = R"(
99 [[block]]
100 struct S {
101 a : f32;
102 };
103
104 [[group(2), binding(1)]] var<storage, read> a : S;
105
106 [[group(3), binding(2)]] var<storage, write> b : S;
107
108 [[group(4), binding(3)]] var<storage, read> c : S;
109
110 [[stage(compute), workgroup_size(1)]]
111 fn f() {
112 }
113 )";
114
115 auto* expect = R"(
116 [[block]]
117 struct S {
118 a : f32;
119 };
120
121 [[group(2), binding(1)]] var<storage, write> a : S;
122
123 [[group(3), binding(2)]] var<storage, write> b : S;
124
125 [[group(4), binding(3)]] var<storage, read> c : S;
126
127 [[stage(compute), workgroup_size(1)]]
128 fn f() {
129 }
130 )";
131
132 DataMap data;
133 data.Add<BindingRemapper::Remappings>(
134 BindingRemapper::BindingPoints{},
135 BindingRemapper::AccessControls{
136 {{2, 1}, ast::Access::kWrite}, // Modify access control
137 // Keep [[group(3), binding(2)]] as is
138 {{4, 3}, ast::Access::kRead}, // Add access control
139 });
140 auto got = Run<BindingRemapper>(src, data);
141
142 EXPECT_EQ(expect, str(got));
143 }
144
145 // TODO(crbug.com/676): Possibly enable if the spec allows for access
146 // decorations in type aliases. If not, just remove.
TEST_F(BindingRemapperTest,DISABLED_RemapAccessControlsWithAliases)147 TEST_F(BindingRemapperTest, DISABLED_RemapAccessControlsWithAliases) {
148 auto* src = R"(
149 [[block]]
150 struct S {
151 a : f32;
152 };
153
154 type, read ReadOnlyS = S;
155
156 type, write WriteOnlyS = S;
157
158 type A = S;
159
160 [[group(2), binding(1)]] var<storage> a : ReadOnlyS;
161
162 [[group(3), binding(2)]] var<storage> b : WriteOnlyS;
163
164 [[group(4), binding(3)]] var<storage> c : A;
165
166 [[stage(compute), workgroup_size(1)]]
167 fn f() {
168 }
169 )";
170
171 auto* expect = R"(
172 [[block]]
173 struct S {
174 a : f32;
175 };
176
177 type, read ReadOnlyS = S;
178
179 type, write WriteOnlyS = S;
180
181 type A = S;
182
183 [[group(2), binding(1)]] var<storage, write> a : S;
184
185 [[group(3), binding(2)]] var<storage> b : WriteOnlyS;
186
187 [[group(4), binding(3)]] var<storage, write> c : S;
188
189 [[stage(compute), workgroup_size(1)]]
190 fn f() {
191 }
192 )";
193
194 DataMap data;
195 data.Add<BindingRemapper::Remappings>(
196 BindingRemapper::BindingPoints{},
197 BindingRemapper::AccessControls{
198 {{2, 1}, ast::Access::kWrite}, // Modify access control
199 // Keep [[group(3), binding(2)]] as is
200 {{4, 3}, ast::Access::kRead}, // Add access control
201 });
202 auto got = Run<BindingRemapper>(src, data);
203
204 EXPECT_EQ(expect, str(got));
205 }
206
TEST_F(BindingRemapperTest,RemapAll)207 TEST_F(BindingRemapperTest, RemapAll) {
208 auto* src = R"(
209 [[block]]
210 struct S {
211 a : f32;
212 };
213
214 [[group(2), binding(1)]] var<storage, read> a : S;
215
216 [[group(3), binding(2)]] var<storage, read> b : S;
217
218 [[stage(compute), workgroup_size(1)]]
219 fn f() {
220 }
221 )";
222
223 auto* expect = R"(
224 [[block]]
225 struct S {
226 a : f32;
227 };
228
229 [[group(4), binding(5)]] var<storage, write> a : S;
230
231 [[group(6), binding(7)]] var<storage, write> b : S;
232
233 [[stage(compute), workgroup_size(1)]]
234 fn f() {
235 }
236 )";
237
238 DataMap data;
239 data.Add<BindingRemapper::Remappings>(
240 BindingRemapper::BindingPoints{
241 {{2, 1}, {4, 5}},
242 {{3, 2}, {6, 7}},
243 },
244 BindingRemapper::AccessControls{
245 {{2, 1}, ast::Access::kWrite},
246 {{3, 2}, ast::Access::kWrite},
247 });
248 auto got = Run<BindingRemapper>(src, data);
249
250 EXPECT_EQ(expect, str(got));
251 }
252
TEST_F(BindingRemapperTest,BindingCollisionsSameEntryPoint)253 TEST_F(BindingRemapperTest, BindingCollisionsSameEntryPoint) {
254 auto* src = R"(
255 [[block]]
256 struct S {
257 i : i32;
258 };
259
260 [[group(2), binding(1)]] var<storage, read> a : S;
261
262 [[group(3), binding(2)]] var<storage, read> b : S;
263
264 [[group(4), binding(3)]] var<storage, read> c : S;
265
266 [[group(5), binding(4)]] var<storage, read> d : S;
267
268 [[stage(compute), workgroup_size(1)]]
269 fn f() {
270 let x : i32 = (((a.i + b.i) + c.i) + d.i);
271 }
272 )";
273
274 auto* expect = R"(
275 [[block]]
276 struct S {
277 i : i32;
278 };
279
280 [[internal(disable_validation__binding_point_collision), group(1), binding(1)]] var<storage, read> a : S;
281
282 [[internal(disable_validation__binding_point_collision), group(1), binding(1)]] var<storage, read> b : S;
283
284 [[internal(disable_validation__binding_point_collision), group(5), binding(4)]] var<storage, read> c : S;
285
286 [[internal(disable_validation__binding_point_collision), group(5), binding(4)]] var<storage, read> d : S;
287
288 [[stage(compute), workgroup_size(1)]]
289 fn f() {
290 let x : i32 = (((a.i + b.i) + c.i) + d.i);
291 }
292 )";
293
294 DataMap data;
295 data.Add<BindingRemapper::Remappings>(
296 BindingRemapper::BindingPoints{
297 {{2, 1}, {1, 1}},
298 {{3, 2}, {1, 1}},
299 {{4, 3}, {5, 4}},
300 },
301 BindingRemapper::AccessControls{}, true);
302 auto got = Run<BindingRemapper>(src, data);
303
304 EXPECT_EQ(expect, str(got));
305 }
306
TEST_F(BindingRemapperTest,BindingCollisionsDifferentEntryPoints)307 TEST_F(BindingRemapperTest, BindingCollisionsDifferentEntryPoints) {
308 auto* src = R"(
309 [[block]]
310 struct S {
311 i : i32;
312 };
313
314 [[group(2), binding(1)]] var<storage, read> a : S;
315
316 [[group(3), binding(2)]] var<storage, read> b : S;
317
318 [[group(4), binding(3)]] var<storage, read> c : S;
319
320 [[group(5), binding(4)]] var<storage, read> d : S;
321
322 [[stage(compute), workgroup_size(1)]]
323 fn f1() {
324 let x : i32 = (a.i + c.i);
325 }
326
327 [[stage(compute), workgroup_size(1)]]
328 fn f2() {
329 let x : i32 = (b.i + d.i);
330 }
331 )";
332
333 auto* expect = R"(
334 [[block]]
335 struct S {
336 i : i32;
337 };
338
339 [[group(1), binding(1)]] var<storage, read> a : S;
340
341 [[group(1), binding(1)]] var<storage, read> b : S;
342
343 [[group(5), binding(4)]] var<storage, read> c : S;
344
345 [[group(5), binding(4)]] var<storage, read> d : S;
346
347 [[stage(compute), workgroup_size(1)]]
348 fn f1() {
349 let x : i32 = (a.i + c.i);
350 }
351
352 [[stage(compute), workgroup_size(1)]]
353 fn f2() {
354 let x : i32 = (b.i + d.i);
355 }
356 )";
357
358 DataMap data;
359 data.Add<BindingRemapper::Remappings>(
360 BindingRemapper::BindingPoints{
361 {{2, 1}, {1, 1}},
362 {{3, 2}, {1, 1}},
363 {{4, 3}, {5, 4}},
364 },
365 BindingRemapper::AccessControls{}, true);
366 auto got = Run<BindingRemapper>(src, data);
367
368 EXPECT_EQ(expect, str(got));
369 }
370
TEST_F(BindingRemapperTest,NoData)371 TEST_F(BindingRemapperTest, NoData) {
372 auto* src = R"(
373 [[block]]
374 struct S {
375 a : f32;
376 };
377
378 [[group(2), binding(1)]] var<storage, read> a : S;
379 [[group(3), binding(2)]] var<storage, read> b : S;
380
381 [[stage(compute), workgroup_size(1)]]
382 fn f() {}
383 )";
384
385 auto* expect =
386 "error: missing transform data for tint::transform::BindingRemapper";
387
388 auto got = Run<BindingRemapper>(src);
389
390 EXPECT_EQ(expect, str(got));
391 }
392
393 } // namespace
394 } // namespace transform
395 } // namespace tint
396