1/*
2 * Copyright 2015-2021 Arm Limited
3 * SPDX-License-Identifier: Apache-2.0 OR MIT
4 *
5 * Licensed under the Apache License, Version 2.0 (the "License");
6 * you may not use this file except in compliance with the License.
7 * You may obtain a copy of the License at
8 *
9 * http://www.apache.org/licenses/LICENSE-2.0
10 *
11 * Unless required by applicable law or agreed to in writing, software
12 * distributed under the License is distributed on an "AS IS" BASIS,
13 * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
14 * See the License for the specific language governing permissions and
15 * limitations under the License.
16 */
17
18/*
19 * At your option, you may choose to accept this material under either:
20 * 1. The Apache License, Version 2.0, found at <http://www.apache.org/licenses/LICENSE-2.0>, or
21 * 2. The MIT License, found at <http://opensource.org/licenses/MIT>.
22 */
23
24#include "spirv_cpp.hpp"
25
26using namespace spv;
27using namespace SPIRV_CROSS_NAMESPACE;
28using namespace std;
29
30void CompilerCPP::emit_buffer_block(const SPIRVariable &var)
31{
32 add_resource_name(id: var.self);
33
34 auto &type = get<SPIRType>(id: var.basetype);
35 auto instance_name = to_name(id: var.self);
36
37 uint32_t descriptor_set = ir.meta[var.self].decoration.set;
38 uint32_t binding = ir.meta[var.self].decoration.binding;
39
40 emit_block_struct(type);
41 auto buffer_name = to_name(id: type.self);
42
43 statement(ts: "internal::Resource<", ts&: buffer_name, ts: type_to_array_glsl(type), ts: "> ", ts&: instance_name, ts: "__;");
44 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name, ts: "__.get()");
45 resource_registrations.push_back(
46 t: join(ts: "s.register_resource(", ts&: instance_name, ts: "__", ts: ", ", ts&: descriptor_set, ts: ", ", ts&: binding, ts: ");"));
47 statement(ts: "");
48}
49
50void CompilerCPP::emit_interface_block(const SPIRVariable &var)
51{
52 add_resource_name(id: var.self);
53
54 auto &type = get<SPIRType>(id: var.basetype);
55
56 const char *qual = var.storage == StorageClassInput ? "StageInput" : "StageOutput";
57 const char *lowerqual = var.storage == StorageClassInput ? "stage_input" : "stage_output";
58 auto instance_name = to_name(id: var.self);
59 uint32_t location = ir.meta[var.self].decoration.location;
60
61 string buffer_name;
62 auto flags = ir.meta[type.self].decoration.decoration_flags;
63 if (flags.get(bit: DecorationBlock))
64 {
65 emit_block_struct(type);
66 buffer_name = to_name(id: type.self);
67 }
68 else
69 buffer_name = type_to_glsl(type);
70
71 statement(ts: "internal::", ts&: qual, ts: "<", ts&: buffer_name, ts: type_to_array_glsl(type), ts: "> ", ts&: instance_name, ts: "__;");
72 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name, ts: "__.get()");
73 resource_registrations.push_back(t: join(ts: "s.register_", ts&: lowerqual, ts: "(", ts&: instance_name, ts: "__", ts: ", ", ts&: location, ts: ");"));
74 statement(ts: "");
75}
76
77void CompilerCPP::emit_shared(const SPIRVariable &var)
78{
79 add_resource_name(id: var.self);
80
81 auto instance_name = to_name(id: var.self);
82 statement(ts: CompilerGLSL::variable_decl(variable: var), ts: ";");
83 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name);
84}
85
86void CompilerCPP::emit_uniform(const SPIRVariable &var)
87{
88 add_resource_name(id: var.self);
89
90 auto &type = get<SPIRType>(id: var.basetype);
91 auto instance_name = to_name(id: var.self);
92
93 uint32_t descriptor_set = ir.meta[var.self].decoration.set;
94 uint32_t binding = ir.meta[var.self].decoration.binding;
95 uint32_t location = ir.meta[var.self].decoration.location;
96
97 string type_name = type_to_glsl(type);
98 remap_variable_type_name(type, var_name: instance_name, type_name);
99
100 if (type.basetype == SPIRType::Image || type.basetype == SPIRType::SampledImage ||
101 type.basetype == SPIRType::AtomicCounter)
102 {
103 statement(ts: "internal::Resource<", ts&: type_name, ts: type_to_array_glsl(type), ts: "> ", ts&: instance_name, ts: "__;");
104 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name, ts: "__.get()");
105 resource_registrations.push_back(
106 t: join(ts: "s.register_resource(", ts&: instance_name, ts: "__", ts: ", ", ts&: descriptor_set, ts: ", ", ts&: binding, ts: ");"));
107 }
108 else
109 {
110 statement(ts: "internal::UniformConstant<", ts&: type_name, ts: type_to_array_glsl(type), ts: "> ", ts&: instance_name, ts: "__;");
111 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name, ts: "__.get()");
112 resource_registrations.push_back(
113 t: join(ts: "s.register_uniform_constant(", ts&: instance_name, ts: "__", ts: ", ", ts&: location, ts: ");"));
114 }
115
116 statement(ts: "");
117}
118
119void CompilerCPP::emit_push_constant_block(const SPIRVariable &var)
120{
121 add_resource_name(id: var.self);
122
123 auto &type = get<SPIRType>(id: var.basetype);
124 auto &flags = ir.meta[var.self].decoration.decoration_flags;
125 if (flags.get(bit: DecorationBinding) || flags.get(bit: DecorationDescriptorSet))
126 SPIRV_CROSS_THROW("Push constant blocks cannot be compiled to GLSL with Binding or Set syntax. "
127 "Remap to location with reflection API first or disable these decorations.");
128
129 emit_block_struct(type);
130 auto buffer_name = to_name(id: type.self);
131 auto instance_name = to_name(id: var.self);
132
133 statement(ts: "internal::PushConstant<", ts&: buffer_name, ts: type_to_array_glsl(type), ts: "> ", ts&: instance_name, ts: ";");
134 statement_no_indent(ts: "#define ", ts&: instance_name, ts: " __res->", ts&: instance_name, ts: ".get()");
135 resource_registrations.push_back(t: join(ts: "s.register_push_constant(", ts&: instance_name, ts: "__", ts: ");"));
136 statement(ts: "");
137}
138
139void CompilerCPP::emit_block_struct(SPIRType &type)
140{
141 // C++ can't do interface blocks, so we fake it by emitting a separate struct.
142 // However, these structs are not allowed to alias anything, so remove it before
143 // emitting the struct.
144 //
145 // The type we have here needs to be resolved to the non-pointer type so we can remove aliases.
146 auto &self = get<SPIRType>(id: type.self);
147 self.type_alias = 0;
148 emit_struct(type&: self);
149}
150
151void CompilerCPP::emit_resources()
152{
153 for (auto &id : ir.ids)
154 {
155 if (id.get_type() == TypeConstant)
156 {
157 auto &c = id.get<SPIRConstant>();
158
159 bool needs_declaration = c.specialization || c.is_used_as_lut;
160
161 if (needs_declaration)
162 {
163 if (!options.vulkan_semantics && c.specialization)
164 {
165 c.specialization_constant_macro_name =
166 constant_value_macro_name(id: get_decoration(id: c.self, decoration: DecorationSpecId));
167 }
168 emit_constant(constant: c);
169 }
170 }
171 else if (id.get_type() == TypeConstantOp)
172 {
173 emit_specialization_constant_op(constant: id.get<SPIRConstantOp>());
174 }
175 }
176
177 // Output all basic struct types which are not Block or BufferBlock as these are declared inplace
178 // when such variables are instantiated.
179 for (auto &id : ir.ids)
180 {
181 if (id.get_type() == TypeType)
182 {
183 auto &type = id.get<SPIRType>();
184 if (type.basetype == SPIRType::Struct && type.array.empty() && !type.pointer &&
185 (!ir.meta[type.self].decoration.decoration_flags.get(bit: DecorationBlock) &&
186 !ir.meta[type.self].decoration.decoration_flags.get(bit: DecorationBufferBlock)))
187 {
188 emit_struct(type);
189 }
190 }
191 }
192
193 statement(ts: "struct Resources : ", ts&: resource_type);
194 begin_scope();
195
196 // Output UBOs and SSBOs
197 for (auto &id : ir.ids)
198 {
199 if (id.get_type() == TypeVariable)
200 {
201 auto &var = id.get<SPIRVariable>();
202 auto &type = get<SPIRType>(id: var.basetype);
203
204 if (var.storage != StorageClassFunction && type.pointer && type.storage == StorageClassUniform &&
205 !is_hidden_variable(var) &&
206 (ir.meta[type.self].decoration.decoration_flags.get(bit: DecorationBlock) ||
207 ir.meta[type.self].decoration.decoration_flags.get(bit: DecorationBufferBlock)))
208 {
209 emit_buffer_block(var);
210 }
211 }
212 }
213
214 // Output push constant blocks
215 for (auto &id : ir.ids)
216 {
217 if (id.get_type() == TypeVariable)
218 {
219 auto &var = id.get<SPIRVariable>();
220 auto &type = get<SPIRType>(id: var.basetype);
221 if (!is_hidden_variable(var) && var.storage != StorageClassFunction && type.pointer &&
222 type.storage == StorageClassPushConstant)
223 {
224 emit_push_constant_block(var);
225 }
226 }
227 }
228
229 // Output in/out interfaces.
230 for (auto &id : ir.ids)
231 {
232 if (id.get_type() == TypeVariable)
233 {
234 auto &var = id.get<SPIRVariable>();
235 auto &type = get<SPIRType>(id: var.basetype);
236
237 if (var.storage != StorageClassFunction && !is_hidden_variable(var) && type.pointer &&
238 (var.storage == StorageClassInput || var.storage == StorageClassOutput) &&
239 interface_variable_exists_in_entry_point(id: var.self))
240 {
241 emit_interface_block(var);
242 }
243 }
244 }
245
246 // Output Uniform Constants (values, samplers, images, etc).
247 for (auto &id : ir.ids)
248 {
249 if (id.get_type() == TypeVariable)
250 {
251 auto &var = id.get<SPIRVariable>();
252 auto &type = get<SPIRType>(id: var.basetype);
253
254 if (var.storage != StorageClassFunction && !is_hidden_variable(var) && type.pointer &&
255 (type.storage == StorageClassUniformConstant || type.storage == StorageClassAtomicCounter))
256 {
257 emit_uniform(var);
258 }
259 }
260 }
261
262 // Global variables.
263 bool emitted = false;
264 for (auto global : global_variables)
265 {
266 auto &var = get<SPIRVariable>(id: global);
267 if (var.storage == StorageClassWorkgroup)
268 {
269 emit_shared(var);
270 emitted = true;
271 }
272 }
273
274 if (emitted)
275 statement(ts: "");
276
277 declare_undefined_values();
278
279 statement(ts: "inline void init(spirv_cross_shader& s)");
280 begin_scope();
281 statement(ts&: resource_type, ts: "::init(s);");
282 for (auto &reg : resource_registrations)
283 statement(ts&: reg);
284 end_scope();
285 resource_registrations.clear();
286
287 end_scope_decl();
288
289 statement(ts: "");
290 statement(ts: "Resources* __res;");
291 if (get_entry_point().model == ExecutionModelGLCompute)
292 statement(ts: "ComputePrivateResources __priv_res;");
293 statement(ts: "");
294
295 // Emit regular globals which are allocated per invocation.
296 emitted = false;
297 for (auto global : global_variables)
298 {
299 auto &var = get<SPIRVariable>(id: global);
300 if (var.storage == StorageClassPrivate)
301 {
302 if (var.storage == StorageClassWorkgroup)
303 emit_shared(var);
304 else
305 statement(ts: CompilerGLSL::variable_decl(variable: var), ts: ";");
306 emitted = true;
307 }
308 }
309
310 if (emitted)
311 statement(ts: "");
312}
313
314string CompilerCPP::compile()
315{
316 ir.fixup_reserved_names();
317
318 // Do not deal with ES-isms like precision, older extensions and such.
319 options.es = false;
320 options.version = 450;
321 backend.float_literal_suffix = true;
322 backend.double_literal_suffix = false;
323 backend.long_long_literal_suffix = true;
324 backend.uint32_t_literal_suffix = true;
325 backend.basic_int_type = "int32_t";
326 backend.basic_uint_type = "uint32_t";
327 backend.swizzle_is_function = true;
328 backend.shared_is_implied = true;
329 backend.unsized_array_supported = false;
330 backend.explicit_struct_type = true;
331 backend.use_initializer_list = true;
332
333 fixup_type_alias();
334 reorder_type_alias();
335 build_function_control_flow_graphs_and_analyze();
336 update_active_builtins();
337
338 uint32_t pass_count = 0;
339 do
340 {
341 resource_registrations.clear();
342 reset(iteration_count: pass_count);
343
344 // Move constructor for this type is broken on GCC 4.9 ...
345 buffer.reset();
346
347 emit_header();
348 emit_resources();
349
350 emit_function(func&: get<SPIRFunction>(id: ir.default_entry_point), return_flags: Bitset());
351
352 pass_count++;
353 } while (is_forcing_recompilation());
354
355 // Match opening scope of emit_header().
356 end_scope_decl();
357 // namespace
358 end_scope();
359
360 // Emit C entry points
361 emit_c_linkage();
362
363 // Entry point in CPP is always main() for the time being.
364 get_entry_point().name = "main";
365
366 return buffer.str();
367}
368
369void CompilerCPP::emit_c_linkage()
370{
371 statement(ts: "");
372
373 statement(ts: "spirv_cross_shader_t *spirv_cross_construct(void)");
374 begin_scope();
375 statement(ts: "return new ", ts&: impl_type, ts: "();");
376 end_scope();
377
378 statement(ts: "");
379 statement(ts: "void spirv_cross_destruct(spirv_cross_shader_t *shader)");
380 begin_scope();
381 statement(ts: "delete static_cast<", ts&: impl_type, ts: "*>(shader);");
382 end_scope();
383
384 statement(ts: "");
385 statement(ts: "void spirv_cross_invoke(spirv_cross_shader_t *shader)");
386 begin_scope();
387 statement(ts: "static_cast<", ts&: impl_type, ts: "*>(shader)->invoke();");
388 end_scope();
389
390 statement(ts: "");
391 statement(ts: "static const struct spirv_cross_interface vtable =");
392 begin_scope();
393 statement(ts: "spirv_cross_construct,");
394 statement(ts: "spirv_cross_destruct,");
395 statement(ts: "spirv_cross_invoke,");
396 end_scope_decl();
397
398 statement(ts: "");
399 statement(ts: "const struct spirv_cross_interface *",
400 ts: interface_name.empty() ? string("spirv_cross_get_interface") : interface_name, ts: "(void)");
401 begin_scope();
402 statement(ts: "return &vtable;");
403 end_scope();
404}
405
406void CompilerCPP::emit_function_prototype(SPIRFunction &func, const Bitset &)
407{
408 if (func.self != ir.default_entry_point)
409 add_function_overload(func);
410
411 local_variable_names = resource_names;
412 string decl;
413
414 auto &type = get<SPIRType>(id: func.return_type);
415 decl += "inline ";
416 decl += type_to_glsl(type);
417 decl += " ";
418
419 if (func.self == ir.default_entry_point)
420 {
421 decl += "main";
422 processing_entry_point = true;
423 }
424 else
425 decl += to_name(id: func.self);
426
427 decl += "(";
428 for (auto &arg : func.arguments)
429 {
430 add_local_variable_name(id: arg.id);
431
432 decl += argument_decl(arg);
433 if (&arg != &func.arguments.back())
434 decl += ", ";
435
436 // Hold a pointer to the parameter so we can invalidate the readonly field if needed.
437 auto *var = maybe_get<SPIRVariable>(id: arg.id);
438 if (var)
439 var->parameter = &arg;
440 }
441
442 decl += ")";
443 statement(ts&: decl);
444}
445
446string CompilerCPP::argument_decl(const SPIRFunction::Parameter &arg)
447{
448 auto &type = expression_type(id: arg.id);
449 bool constref = !type.pointer || arg.write_count == 0;
450
451 auto &var = get<SPIRVariable>(id: arg.id);
452
453 string base = type_to_glsl(type);
454 string variable_name = to_name(id: var.self);
455 remap_variable_type_name(type, var_name: variable_name, type_name&: base);
456
457 for (uint32_t i = 0; i < type.array.size(); i++)
458 base = join(ts: "std::array<", ts&: base, ts: ", ", ts: to_array_size(type, index: i), ts: ">");
459
460 return join(ts: constref ? "const " : "", ts&: base, ts: " &", ts&: variable_name);
461}
462
463string CompilerCPP::variable_decl(const SPIRType &type, const string &name, uint32_t /* id */)
464{
465 string base = type_to_glsl(type);
466 remap_variable_type_name(type, var_name: name, type_name&: base);
467 bool runtime = false;
468
469 for (uint32_t i = 0; i < type.array.size(); i++)
470 {
471 auto &array = type.array[i];
472 if (!array && type.array_size_literal[i])
473 {
474 // Avoid using runtime arrays with std::array since this is undefined.
475 // Runtime arrays cannot be passed around as values, so this is fine.
476 runtime = true;
477 }
478 else
479 base = join(ts: "std::array<", ts&: base, ts: ", ", ts: to_array_size(type, index: i), ts: ">");
480 }
481 base += ' ';
482 return base + name + (runtime ? "[1]" : "");
483}
484
485void CompilerCPP::emit_header()
486{
487 auto &execution = get_entry_point();
488
489 statement(ts: "// This C++ shader is autogenerated by spirv-cross.");
490 statement(ts: "#include \"spirv_cross/internal_interface.hpp\"");
491 statement(ts: "#include \"spirv_cross/external_interface.h\"");
492 // Needed to properly implement GLSL-style arrays.
493 statement(ts: "#include <array>");
494 statement(ts: "#include <stdint.h>");
495 statement(ts: "");
496 statement(ts: "using namespace spirv_cross;");
497 statement(ts: "using namespace glm;");
498 statement(ts: "");
499
500 statement(ts: "namespace Impl");
501 begin_scope();
502
503 switch (execution.model)
504 {
505 case ExecutionModelGeometry:
506 case ExecutionModelTessellationControl:
507 case ExecutionModelTessellationEvaluation:
508 case ExecutionModelGLCompute:
509 case ExecutionModelFragment:
510 case ExecutionModelVertex:
511 statement(ts: "struct Shader");
512 begin_scope();
513 break;
514
515 default:
516 SPIRV_CROSS_THROW("Unsupported execution model.");
517 }
518
519 switch (execution.model)
520 {
521 case ExecutionModelGeometry:
522 impl_type = "GeometryShader<Impl::Shader, Impl::Shader::Resources>";
523 resource_type = "GeometryResources";
524 break;
525
526 case ExecutionModelVertex:
527 impl_type = "VertexShader<Impl::Shader, Impl::Shader::Resources>";
528 resource_type = "VertexResources";
529 break;
530
531 case ExecutionModelFragment:
532 impl_type = "FragmentShader<Impl::Shader, Impl::Shader::Resources>";
533 resource_type = "FragmentResources";
534 break;
535
536 case ExecutionModelGLCompute:
537 impl_type = join(ts: "ComputeShader<Impl::Shader, Impl::Shader::Resources, ", ts&: execution.workgroup_size.x, ts: ", ",
538 ts&: execution.workgroup_size.y, ts: ", ", ts&: execution.workgroup_size.z, ts: ">");
539 resource_type = "ComputeResources";
540 break;
541
542 case ExecutionModelTessellationControl:
543 impl_type = "TessControlShader<Impl::Shader, Impl::Shader::Resources>";
544 resource_type = "TessControlResources";
545 break;
546
547 case ExecutionModelTessellationEvaluation:
548 impl_type = "TessEvaluationShader<Impl::Shader, Impl::Shader::Resources>";
549 resource_type = "TessEvaluationResources";
550 break;
551
552 default:
553 SPIRV_CROSS_THROW("Unsupported execution model.");
554 }
555}
556

source code of qtshadertools/src/3rdparty/SPIRV-Cross/spirv_cpp.cpp