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, variable_id: var.self), 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, variable_id: var.self), 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, variable_id: var.self), 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, variable_id: var.self), 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, variable_id: var.self), 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 statement(ts: "inline void init(spirv_cross_shader& s)");
278 begin_scope();
279 statement(ts&: resource_type, ts: "::init(s);");
280 for (auto &reg : resource_registrations)
281 statement(ts&: reg);
282 end_scope();
283 resource_registrations.clear();
284
285 end_scope_decl();
286
287 statement(ts: "");
288 statement(ts: "Resources* __res;");
289 if (get_entry_point().model == ExecutionModelGLCompute)
290 statement(ts: "ComputePrivateResources __priv_res;");
291 statement(ts: "");
292
293 // Emit regular globals which are allocated per invocation.
294 emitted = false;
295 for (auto global : global_variables)
296 {
297 auto &var = get<SPIRVariable>(id: global);
298 if (var.storage == StorageClassPrivate)
299 {
300 if (var.storage == StorageClassWorkgroup)
301 emit_shared(var);
302 else
303 statement(ts: CompilerGLSL::variable_decl(variable: var), ts: ";");
304 emitted = true;
305 }
306 }
307
308 if (emitted)
309 statement(ts: "");
310}
311
312string CompilerCPP::compile()
313{
314 ir.fixup_reserved_names();
315
316 // Do not deal with ES-isms like precision, older extensions and such.
317 options.es = false;
318 options.version = 450;
319 backend.float_literal_suffix = true;
320 backend.double_literal_suffix = false;
321 backend.long_long_literal_suffix = true;
322 backend.uint32_t_literal_suffix = true;
323 backend.basic_int_type = "int32_t";
324 backend.basic_uint_type = "uint32_t";
325 backend.swizzle_is_function = true;
326 backend.shared_is_implied = true;
327 backend.unsized_array_supported = false;
328 backend.explicit_struct_type = true;
329 backend.use_initializer_list = true;
330
331 fixup_type_alias();
332 reorder_type_alias();
333 build_function_control_flow_graphs_and_analyze();
334 update_active_builtins();
335
336 uint32_t pass_count = 0;
337 do
338 {
339 resource_registrations.clear();
340 reset(iteration_count: pass_count);
341
342 // Move constructor for this type is broken on GCC 4.9 ...
343 buffer.reset();
344
345 emit_header();
346 emit_resources();
347
348 emit_function(func&: get<SPIRFunction>(id: ir.default_entry_point), return_flags: Bitset());
349
350 pass_count++;
351 } while (is_forcing_recompilation());
352
353 // Match opening scope of emit_header().
354 end_scope_decl();
355 // namespace
356 end_scope();
357
358 // Emit C entry points
359 emit_c_linkage();
360
361 // Entry point in CPP is always main() for the time being.
362 get_entry_point().name = "main";
363
364 return buffer.str();
365}
366
367void CompilerCPP::emit_c_linkage()
368{
369 statement(ts: "");
370
371 statement(ts: "spirv_cross_shader_t *spirv_cross_construct(void)");
372 begin_scope();
373 statement(ts: "return new ", ts&: impl_type, ts: "();");
374 end_scope();
375
376 statement(ts: "");
377 statement(ts: "void spirv_cross_destruct(spirv_cross_shader_t *shader)");
378 begin_scope();
379 statement(ts: "delete static_cast<", ts&: impl_type, ts: "*>(shader);");
380 end_scope();
381
382 statement(ts: "");
383 statement(ts: "void spirv_cross_invoke(spirv_cross_shader_t *shader)");
384 begin_scope();
385 statement(ts: "static_cast<", ts&: impl_type, ts: "*>(shader)->invoke();");
386 end_scope();
387
388 statement(ts: "");
389 statement(ts: "static const struct spirv_cross_interface vtable =");
390 begin_scope();
391 statement(ts: "spirv_cross_construct,");
392 statement(ts: "spirv_cross_destruct,");
393 statement(ts: "spirv_cross_invoke,");
394 end_scope_decl();
395
396 statement(ts: "");
397 statement(ts: "const struct spirv_cross_interface *",
398 ts: interface_name.empty() ? string("spirv_cross_get_interface") : interface_name, ts: "(void)");
399 begin_scope();
400 statement(ts: "return &vtable;");
401 end_scope();
402}
403
404void CompilerCPP::emit_function_prototype(SPIRFunction &func, const Bitset &)
405{
406 if (func.self != ir.default_entry_point)
407 add_function_overload(func);
408
409 local_variable_names = resource_names;
410 string decl;
411
412 auto &type = get<SPIRType>(id: func.return_type);
413 decl += "inline ";
414 decl += type_to_glsl(type);
415 decl += " ";
416
417 if (func.self == ir.default_entry_point)
418 {
419 decl += "main";
420 processing_entry_point = true;
421 }
422 else
423 decl += to_name(id: func.self);
424
425 decl += "(";
426 for (auto &arg : func.arguments)
427 {
428 add_local_variable_name(id: arg.id);
429
430 decl += argument_decl(arg);
431 if (&arg != &func.arguments.back())
432 decl += ", ";
433
434 // Hold a pointer to the parameter so we can invalidate the readonly field if needed.
435 auto *var = maybe_get<SPIRVariable>(id: arg.id);
436 if (var)
437 var->parameter = &arg;
438 }
439
440 decl += ")";
441 statement(ts&: decl);
442}
443
444string CompilerCPP::argument_decl(const SPIRFunction::Parameter &arg)
445{
446 auto &type = expression_type(id: arg.id);
447 bool constref = !type.pointer || arg.write_count == 0;
448
449 auto &var = get<SPIRVariable>(id: arg.id);
450
451 string base = type_to_glsl(type);
452 string variable_name = to_name(id: var.self);
453 remap_variable_type_name(type, var_name: variable_name, type_name&: base);
454
455 for (uint32_t i = 0; i < type.array.size(); i++)
456 base = join(ts: "std::array<", ts&: base, ts: ", ", ts: to_array_size(type, index: i), ts: ">");
457
458 return join(ts: constref ? "const " : "", ts&: base, ts: " &", ts&: variable_name);
459}
460
461string CompilerCPP::variable_decl(const SPIRType &type, const string &name, uint32_t /* id */)
462{
463 string base = type_to_glsl(type);
464 remap_variable_type_name(type, var_name: name, type_name&: base);
465 bool runtime = false;
466
467 for (uint32_t i = 0; i < type.array.size(); i++)
468 {
469 auto &array = type.array[i];
470 if (!array && type.array_size_literal[i])
471 {
472 // Avoid using runtime arrays with std::array since this is undefined.
473 // Runtime arrays cannot be passed around as values, so this is fine.
474 runtime = true;
475 }
476 else
477 base = join(ts: "std::array<", ts&: base, ts: ", ", ts: to_array_size(type, index: i), ts: ">");
478 }
479 base += ' ';
480 return base + name + (runtime ? "[1]" : "");
481}
482
483void CompilerCPP::emit_header()
484{
485 auto &execution = get_entry_point();
486
487 statement(ts: "// This C++ shader is autogenerated by spirv-cross.");
488 statement(ts: "#include \"spirv_cross/internal_interface.hpp\"");
489 statement(ts: "#include \"spirv_cross/external_interface.h\"");
490 // Needed to properly implement GLSL-style arrays.
491 statement(ts: "#include <array>");
492 statement(ts: "#include <stdint.h>");
493 statement(ts: "");
494 statement(ts: "using namespace spirv_cross;");
495 statement(ts: "using namespace glm;");
496 statement(ts: "");
497
498 statement(ts: "namespace Impl");
499 begin_scope();
500
501 switch (execution.model)
502 {
503 case ExecutionModelGeometry:
504 case ExecutionModelTessellationControl:
505 case ExecutionModelTessellationEvaluation:
506 case ExecutionModelGLCompute:
507 case ExecutionModelFragment:
508 case ExecutionModelVertex:
509 statement(ts: "struct Shader");
510 begin_scope();
511 break;
512
513 default:
514 SPIRV_CROSS_THROW("Unsupported execution model.");
515 }
516
517 switch (execution.model)
518 {
519 case ExecutionModelGeometry:
520 impl_type = "GeometryShader<Impl::Shader, Impl::Shader::Resources>";
521 resource_type = "GeometryResources";
522 break;
523
524 case ExecutionModelVertex:
525 impl_type = "VertexShader<Impl::Shader, Impl::Shader::Resources>";
526 resource_type = "VertexResources";
527 break;
528
529 case ExecutionModelFragment:
530 impl_type = "FragmentShader<Impl::Shader, Impl::Shader::Resources>";
531 resource_type = "FragmentResources";
532 break;
533
534 case ExecutionModelGLCompute:
535 impl_type = join(ts: "ComputeShader<Impl::Shader, Impl::Shader::Resources, ", ts&: execution.workgroup_size.x, ts: ", ",
536 ts&: execution.workgroup_size.y, ts: ", ", ts&: execution.workgroup_size.z, ts: ">");
537 resource_type = "ComputeResources";
538 break;
539
540 case ExecutionModelTessellationControl:
541 impl_type = "TessControlShader<Impl::Shader, Impl::Shader::Resources>";
542 resource_type = "TessControlResources";
543 break;
544
545 case ExecutionModelTessellationEvaluation:
546 impl_type = "TessEvaluationShader<Impl::Shader, Impl::Shader::Resources>";
547 resource_type = "TessEvaluationResources";
548 break;
549
550 default:
551 SPIRV_CROSS_THROW("Unsupported execution model.");
552 }
553}
554

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