| Flutter Impeller |

spirv_sksl.cc
Go to the documentation of this file.
1 // Copyright 2013 The Flutter Authors. All rights reserved.
2 // Use of this source code is governed by a BSD-style license that can be
3 // found in the LICENSE file.
4
5 #include "impeller/compiler/spirv_sksl.h"
6 #include "impeller/compiler/uniform_sorter.h"
7
8 using namespace spv;
9 using namespace SPIRV_CROSS_NAMESPACE;
10
11 namespace impeller {
12 namespace compiler {
13
14 // This replaces the SPIRV_CROSS_THROW which aborts and drops the
15 // error message in non-debug modes.
16 void report_and_exit(const std::string& msg) {
17 fprintf(stderr, "There was a compiler error: %s\n", msg.c_str());
18 fflush(stderr);
19 exit(1);
20 }
21
22 #define FLUTTER_CROSS_THROW(x) report_and_exit(x)
23
24 std::string CompilerSkSL::compile() {
25 ir.fixup_reserved_names();
26
27 if (get_execution_model() != ExecutionModelFragment) {
28 FLUTTER_CROSS_THROW("Only fragment shaders are supported.'");
29 return "";
30 }
31
32 options.es = false;
33 options.version = 100;
34 options.vulkan_semantics = false;
35 options.enable_420pack_extension = false;
36 options.flatten_multidimensional_arrays = true;
37
38 backend.allow_precision_qualifiers = false;
39 backend.basic_int16_type = "short";
40 backend.basic_int_type = "int";
41 backend.basic_uint16_type = "ushort";
42 backend.basic_uint_type = "uint";
43 backend.double_literal_suffix = false;
44 backend.float_literal_suffix = false;
45 backend.long_long_literal_suffix = false;
46 backend.needs_row_major_load_workaround = true;
47 backend.nonuniform_qualifier = "";
48 backend.support_precise_qualifier = false;
49 backend.uint32_t_literal_suffix = false;
50 backend.use_array_constructor = true;
51 backend.workgroup_size_is_hidden = true;
52
53 fixup_user_functions();
54
55 fixup_anonymous_struct_names();
56 fixup_type_alias();
57 reorder_type_alias();
58 build_function_control_flow_graphs_and_analyze();
59 fixup_image_load_store_access();
60 update_active_builtins();
61 analyze_image_and_sampler_usage();
62 analyze_interlocked_resource_usage();
63
64 uint32_t pass_count = 0;
65 do {
66 reset(pass_count);
67
68 // Move constructor for this type is broken on GCC 4.9 ...
69 buffer.reset();
70
71 emit_header();
72 emit_resources();
73
74 emit_function(get<SPIRFunction>(ir.default_entry_point), Bitset());
75
76 pass_count++;
77 } while (is_forcing_recompilation());
78
79 statement("half4 main(float2 iFragCoord)");
80 begin_scope();
81 statement(" flutter_FragCoord = float4(iFragCoord, 0, 0);");
82 statement(" FLT_main();");
83 statement(" return " + output_name_ + ";");
84 end_scope();
85
86 return buffer.str();
87 }
88
89 void CompilerSkSL::fixup_user_functions() {
90 const std::string prefix = "FLT_flutter_local_";
91 ir.for_each_typed_id<SPIRFunction>([&](uint32_t, const SPIRFunction& func) {
92 const auto& original_name = get_name(func.self);
93 // Just in case. Don't add the prefix a second time.
94 if (original_name.rfind(prefix, 0) == 0) {
95 return;
96 }
97 std::string new_name = prefix + original_name;
98 set_name(func.self, new_name);
99 });
100
101 ir.for_each_typed_id<SPIRFunctionPrototype>(
102 [&](uint32_t, const SPIRFunctionPrototype& func) {
103 const auto& original_name = get_name(func.self);
104 // Just in case. Don't add the prefix a second time.
105 if (original_name.rfind(prefix, 0) == 0) {
106 return;
107 }
108 std::string new_name = prefix + original_name;
109 set_name(func.self, new_name);
110 });
111 }
112
113 void CompilerSkSL::emit_header() {
114 statement("// This SkSL shader is autogenerated by spirv-cross.");
115 statement("");
116 statement("float4 flutter_FragCoord;");
117 statement("");
118 }
119
120 void CompilerSkSL::emit_uniform(const SPIRVariable& var) {
121 auto& type = get<SPIRType>(var.basetype);
122 if (type.basetype == SPIRType::UInt && is_legacy()) {
123 FLUTTER_CROSS_THROW("SkSL does not support unsigned integers: '" +
124 get_name(var.self) + "'");
125 }
126
127 add_resource_name(var.self);
128 statement(variable_decl(var), ";");
129
130 // The Flutter FragmentProgram implementation passes additional uniforms along
131 // with shader uniforms that encode the shader width and height.
132 if (type.basetype == SPIRType::SampledImage) {
133 std::string name = to_name(var.self);
134 statement("uniform half2 " + name + "_size;");
135 }
136 }
137
138 bool CompilerSkSL::emit_constant_resources() {
139 bool emitted = false;
140
141 for (auto& id : ir.ids) {
142 if (id.get_type() == TypeConstant) {
143 auto& c = id.get<SPIRConstant>();
144 bool needs_declaration = c.specialization || c.is_used_as_lut;
145 if (needs_declaration) {
146 if (!options.vulkan_semantics && c.specialization) {
147 c.specialization_constant_macro_name = constant_value_macro_name(
148 get_decoration(c.self, DecorationSpecId));
149 }
150 emit_constant(c);
151 emitted = true;
152 }
153 } else if (id.get_type() == TypeConstantOp) {
154 emit_specialization_constant_op(id.get<SPIRConstantOp>());
155 emitted = true;
156 }
157 }
158
159 return emitted;
160 }
161
162 bool CompilerSkSL::emit_struct_resources() {
163 bool emitted = false;
164
165 // Output all basic struct types which are not Block or BufferBlock as these
166 // are declared inplace when such variables are instantiated.
167 for (auto& id : ir.ids) {
168 if (id.get_type() == TypeType) {
169 auto& type = id.get<SPIRType>();
170 if (type.basetype == SPIRType::Struct && type.array.empty() &&
171 !type.pointer &&
172 (!ir.meta[type.self].decoration.decoration_flags.get(
173 DecorationBlock) &&
174 !ir.meta[type.self].decoration.decoration_flags.get(
175 DecorationBufferBlock))) {
176 emit_struct(type);
177 emitted = true;
178 }
179 }
180 }
181
182 return emitted;
183 }
184
185 void CompilerSkSL::detect_unsupported_resources() {
186 for (auto& id : ir.ids) {
187 if (id.get_type() == TypeVariable) {
188 auto& var = id.get<SPIRVariable>();
189 auto& type = get<SPIRType>(var.basetype);
190
191 // UBOs and SSBOs are not supported.
192 if (var.storage != StorageClassFunction && type.pointer &&
193 type.storage == StorageClassUniform && !is_hidden_variable(var) &&
194 (ir.meta[type.self].decoration.decoration_flags.get(
195 DecorationBlock) ||
196 ir.meta[type.self].decoration.decoration_flags.get(
197 DecorationBufferBlock))) {
198 FLUTTER_CROSS_THROW("SkSL does not support UBOs or SSBOs: '" +
199 get_name(var.self) + "'");
200 }
201
202 // Push constant blocks are not supported.
203 if (!is_hidden_variable(var) && var.storage != StorageClassFunction &&
204 type.pointer && type.storage == StorageClassPushConstant) {
205 FLUTTER_CROSS_THROW("SkSL does not support push constant blocks: '" +
206 get_name(var.self) + "'");
207 }
208
209 // User specified inputs are not supported.
210 if (!is_hidden_variable(var) && var.storage != StorageClassFunction &&
211 type.pointer && type.storage == StorageClassInput) {
212 FLUTTER_CROSS_THROW("SkSL does not support inputs: '" +
213 get_name(var.self) + "'");
214 }
215 }
216 }
217 }
218
219 bool CompilerSkSL::emit_uniform_resources() {
220 bool emitted = false;
221
222 // Output Uniform Constants (values, samplers, images, etc).
223 std::vector<ID> regular_uniforms =
224 SortUniforms(&ir, this, SPIRType::SampledImage, /*include=*/false);
225 std::vector<ID> shader_uniforms =
226 SortUniforms(&ir, this, SPIRType::SampledImage);
227 if (regular_uniforms.size() > 0 || shader_uniforms.size() > 0) {
228 emitted = true;
229 }
230
231 for (const auto& id : regular_uniforms) {
232 auto& var = get<SPIRVariable>(id);
233 emit_uniform(var);
234 }
235
236 for (const auto& id : shader_uniforms) {
237 auto& var = get<SPIRVariable>(id);
238 emit_uniform(var);
239 }
240
241 return emitted;
242 }
243
244 bool CompilerSkSL::emit_output_resources() {
245 bool emitted = false;
246
247 // Output 'out' variables. These are restricted to the cases handled by
248 // SkSL in 'emit_interface_block'.
249 for (auto& id : ir.ids) {
250 if (id.get_type() == TypeVariable) {
251 auto& var = id.get<SPIRVariable>();
252 auto& type = get<SPIRType>(var.basetype);
253 if (var.storage != StorageClassFunction && !is_hidden_variable(var) &&
254 type.pointer &&
255 (var.storage == StorageClassInput ||
256 var.storage == StorageClassOutput) &&
257 interface_variable_exists_in_entry_point(var.self)) {
258 emit_interface_block(var);
259 emitted = true;
260 }
261 }
262 }
263
264 return emitted;
265 }
266
267 bool CompilerSkSL::emit_global_variable_resources() {
268 bool emitted = false;
269
270 for (auto global : global_variables) {
271 auto& var = get<SPIRVariable>(global);
272 if (is_hidden_variable(var, true)) {
273 continue;
274 }
275 if (var.storage != StorageClassOutput) {
276 if (!variable_is_lut(var)) {
277 add_resource_name(var.self);
278 std::string initializer;
279 if (options.force_zero_initialized_variables &&
280 var.storage == StorageClassPrivate && !var.initializer &&
281 !var.static_expression &&
282 type_can_zero_initialize(get_variable_data_type(var))) {
283 initializer = join(" = ", to_zero_initialized_expression(
284 get_variable_data_type_id(var)));
285 }
286 statement(variable_decl(var), initializer, ";");
287 emitted = true;
288 }
289 } else if (var.initializer &&
290 maybe_get<SPIRConstant>(var.initializer) != nullptr) {
291 emit_output_variable_initializer(var);
292 }
293 }
294
295 return emitted;
296 }
297
298 bool CompilerSkSL::emit_undefined_values() {
299 bool emitted = false;
300
301 ir.for_each_typed_id<SPIRUndef>([&](uint32_t, const SPIRUndef& undef) {
302 auto& type = this->get<SPIRType>(undef.basetype);
303 // OpUndef can be void for some reason ...
304 if (type.basetype == SPIRType::Void) {
305 return;
306 }
307
308 std::string initializer;
309 if (options.force_zero_initialized_variables &&
310 type_can_zero_initialize(type)) {
311 initializer = join(" = ", to_zero_initialized_expression(undef.basetype));
312 }
313
314 statement(variable_decl(type, to_name(undef.self), undef.self), initializer,
315 ";");
316 emitted = true;
317 });
318
319 return emitted;
320 }
321
322 void CompilerSkSL::emit_resources() {
323 detect_unsupported_resources();
324
325 if (emit_constant_resources()) {
326 statement("");
327 }
328
329 if (emit_struct_resources()) {
330 statement("");
331 }
332
333 if (emit_uniform_resources()) {
334 statement("");
335 }
336
337 if (emit_output_resources()) {
338 statement("");
339 }
340
341 if (emit_global_variable_resources()) {
342 statement("");
343 }
344
345 if (emit_undefined_values()) {
346 statement("");
347 }
348 }
349
350 void CompilerSkSL::emit_interface_block(const SPIRVariable& var) {
351 auto& type = get<SPIRType>(var.basetype);
352 bool block =
353 ir.meta[type.self].decoration.decoration_flags.get(DecorationBlock);
354 if (block) {
355 FLUTTER_CROSS_THROW("Interface blocks are not supported: '" +
356 to_name(var.self) + "'");
357 }
358
359 // The output is emitted as a global variable, which is returned from the
360 // wrapper around the 'main' function. Only one output variable is allowed.
361 add_resource_name(var.self);
362 statement(variable_decl(type, to_name(var.self), var.self), ";");
363 if (output_name_.empty()) {
364 output_name_ = to_name(var.self);
365 } else if (to_name(var.self) != output_name_) {
366 FLUTTER_CROSS_THROW("Only one output variable is supported: '" +
367 to_name(var.self) + "'");
368 }
369 }
370
371 void CompilerSkSL::emit_function_prototype(SPIRFunction& func,
372 const Bitset& return_flags) {
373 // If this is not the entrypoint, then no special processsing for SkSL is
374 // required.
375 if (func.self != ir.default_entry_point) {
376 CompilerGLSL::emit_function_prototype(func, return_flags);
377 return;
378 }
379
380 auto& type = get<SPIRType>(func.return_type);
381 if (type.basetype != SPIRType::Void) {
382 FLUTTER_CROSS_THROW(
383 "Return type of the entrypoint function must be 'void'");
384 }
385
386 if (func.arguments.size() != 0) {
387 FLUTTER_CROSS_THROW(
388 "The entry point function should not acept any parameters.");
389 }
390
391 processing_entry_point = true;
392
393 // If this is the entrypoint of a fragment shader, then GLSL requires the
394 // prototype to be "void main()", and so it is safe to rewrite as
395 // "void FLT_main()".
396 statement("void FLT_main()");
397 }
398
399 std::string CompilerSkSL::image_type_glsl(const SPIRType& type,
400 uint32_t id,
401 bool member) {
402 if (type.basetype != SPIRType::SampledImage || type.image.dim != Dim2D) {
403 FLUTTER_CROSS_THROW("Only sampler2D uniform image types are supported.");
404 return "???";
405 }
406 return "shader";
407 }
408
409 std::string CompilerSkSL::builtin_to_glsl(BuiltIn builtin,
410 StorageClass storage) {
411 std::string gl_builtin = CompilerGLSL::builtin_to_glsl(builtin, storage);
412 switch (builtin) {
413 case BuiltInFragCoord:
414 return "flutter_FragCoord";
415 default:
416 FLUTTER_CROSS_THROW("Builtin '" + gl_builtin + "' is not supported.");
417 break;
418 }
419
420 return "???";
421 }
422
423 std::string CompilerSkSL::to_texture_op(
424 const Instruction& i,
425 bool sparse,
426 bool* forward,
427 SmallVector<uint32_t>& inherited_expressions) {
428 auto op = static_cast<Op>(i.op);
429 if (op != OpImageSampleImplicitLod) {
430 FLUTTER_CROSS_THROW("Only simple shader sampling is supported.");
431 return "???";
432 }
433 return CompilerGLSL::to_texture_op(i, sparse, forward, inherited_expressions);
434 }
435
436 std::string CompilerSkSL::to_function_name(
437 const CompilerGLSL::TextureFunctionNameArguments& args) {
438 std::string name = to_expression(args.base.img);
439 return name + ".eval";
440 }
441
442 std::string CompilerSkSL::to_function_args(const TextureFunctionArguments& args,
443 bool* p_forward) {
444 std::string name = to_expression(args.base.img);
445
446 std::string glsl_args = CompilerGLSL::to_function_args(args, p_forward);
447 // SkSL only supports coordinates. All other arguments to texture are
448 // unsupported and will generate invalid SkSL.
449 if (args.grad_x || args.grad_y || args.lod || args.offset || args.sample ||
450 args.min_lod || args.sparse_texel || args.bias || args.component) {
451 FLUTTER_CROSS_THROW(
452 "Only sampler and position arguments are supported in texture() "
453 "calls.");
454 }
455
456 // GLSL puts the shader as the first argument, but in SkSL the shader is
457 // implicitly passed as the reciever of the 'eval' method. Therefore, the
458 // shader is removed from the GLSL argument list.
459 std::string no_shader;
460 auto npos = glsl_args.find(", "); // The first ','.
461 if (npos != std::string::npos) {
462 no_shader = glsl_args.substr(npos + 1); // The string after the first ','.
463 }
464
465 if (no_shader.empty()) {
466 FLUTTER_CROSS_THROW("Unexpected shader sampling arguments: '(" + glsl_args +
467 ")'");
468 return "()";
469 }
470
471 return name + "_size * (" + no_shader + ")";
472 }
473
474 } // namespace compiler
475 } // namespace impeller
uniform_sorter.h
impeller::compiler::report_and_exit
void report_and_exit(const std::string &msg)
Definition: spirv_sksl.cc:16
impeller::SortUniforms
std::vector< spirv_cross::ID > SortUniforms(const spirv_cross::ParsedIR *ir, const spirv_cross::Compiler *compiler, std::optional< spirv_cross::SPIRType::BaseType > type_filter, bool include)
Sorts uniform declarations in an IR according to decoration order.
Definition: uniform_sorter.cc:11
FLUTTER_CROSS_THROW
#define FLUTTER_CROSS_THROW(x)
Definition: spirv_sksl.cc:22
spirv_sksl.h
impeller
Definition: aiks_blur_unittests.cc:20