• Home
  • Line#
  • Scopes#
  • Navigate#
  • Raw
  • Download
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 #ifndef SPIRV_CROSS_COMMON_HPP
25 #define SPIRV_CROSS_COMMON_HPP
26 
27 #include "spirv.hpp"
28 #include "spirv_cross_containers.hpp"
29 #include "spirv_cross_error_handling.hpp"
30 #include <functional>
31 
32 // A bit crude, but allows projects which embed SPIRV-Cross statically to
33 // effectively hide all the symbols from other projects.
34 // There is a case where we have:
35 // - Project A links against SPIRV-Cross statically.
36 // - Project A links against Project B statically.
37 // - Project B links against SPIRV-Cross statically (might be a different version).
38 // This leads to a conflict with extremely bizarre results.
39 // By overriding the namespace in one of the project builds, we can work around this.
40 // If SPIRV-Cross is embedded in dynamic libraries,
41 // prefer using -fvisibility=hidden on GCC/Clang instead.
42 #ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
43 #define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
44 #else
45 #define SPIRV_CROSS_NAMESPACE spirv_cross
46 #endif
47 
48 namespace SPIRV_CROSS_NAMESPACE
49 {
50 namespace inner
51 {
52 template <typename T>
join_helper(StringStream<> & stream,T && t)53 void join_helper(StringStream<> &stream, T &&t)
54 {
55 	stream << std::forward<T>(t);
56 }
57 
58 template <typename T, typename... Ts>
join_helper(StringStream<> & stream,T && t,Ts &&...ts)59 void join_helper(StringStream<> &stream, T &&t, Ts &&... ts)
60 {
61 	stream << std::forward<T>(t);
62 	join_helper(stream, std::forward<Ts>(ts)...);
63 }
64 } // namespace inner
65 
66 class Bitset
67 {
68 public:
69 	Bitset() = default;
Bitset(uint64_t lower_)70 	explicit inline Bitset(uint64_t lower_)
71 	    : lower(lower_)
72 	{
73 	}
74 
get(uint32_t bit) const75 	inline bool get(uint32_t bit) const
76 	{
77 		if (bit < 64)
78 			return (lower & (1ull << bit)) != 0;
79 		else
80 			return higher.count(bit) != 0;
81 	}
82 
set(uint32_t bit)83 	inline void set(uint32_t bit)
84 	{
85 		if (bit < 64)
86 			lower |= 1ull << bit;
87 		else
88 			higher.insert(bit);
89 	}
90 
clear(uint32_t bit)91 	inline void clear(uint32_t bit)
92 	{
93 		if (bit < 64)
94 			lower &= ~(1ull << bit);
95 		else
96 			higher.erase(bit);
97 	}
98 
get_lower() const99 	inline uint64_t get_lower() const
100 	{
101 		return lower;
102 	}
103 
reset()104 	inline void reset()
105 	{
106 		lower = 0;
107 		higher.clear();
108 	}
109 
merge_and(const Bitset & other)110 	inline void merge_and(const Bitset &other)
111 	{
112 		lower &= other.lower;
113 		std::unordered_set<uint32_t> tmp_set;
114 		for (auto &v : higher)
115 			if (other.higher.count(v) != 0)
116 				tmp_set.insert(v);
117 		higher = std::move(tmp_set);
118 	}
119 
merge_or(const Bitset & other)120 	inline void merge_or(const Bitset &other)
121 	{
122 		lower |= other.lower;
123 		for (auto &v : other.higher)
124 			higher.insert(v);
125 	}
126 
operator ==(const Bitset & other) const127 	inline bool operator==(const Bitset &other) const
128 	{
129 		if (lower != other.lower)
130 			return false;
131 
132 		if (higher.size() != other.higher.size())
133 			return false;
134 
135 		for (auto &v : higher)
136 			if (other.higher.count(v) == 0)
137 				return false;
138 
139 		return true;
140 	}
141 
operator !=(const Bitset & other) const142 	inline bool operator!=(const Bitset &other) const
143 	{
144 		return !(*this == other);
145 	}
146 
147 	template <typename Op>
for_each_bit(const Op & op) const148 	void for_each_bit(const Op &op) const
149 	{
150 		// TODO: Add ctz-based iteration.
151 		for (uint32_t i = 0; i < 64; i++)
152 		{
153 			if (lower & (1ull << i))
154 				op(i);
155 		}
156 
157 		if (higher.empty())
158 			return;
159 
160 		// Need to enforce an order here for reproducible results,
161 		// but hitting this path should happen extremely rarely, so having this slow path is fine.
162 		SmallVector<uint32_t> bits;
163 		bits.reserve(higher.size());
164 		for (auto &v : higher)
165 			bits.push_back(v);
166 		std::sort(std::begin(bits), std::end(bits));
167 
168 		for (auto &v : bits)
169 			op(v);
170 	}
171 
empty() const172 	inline bool empty() const
173 	{
174 		return lower == 0 && higher.empty();
175 	}
176 
177 private:
178 	// The most common bits to set are all lower than 64,
179 	// so optimize for this case. Bits spilling outside 64 go into a slower data structure.
180 	// In almost all cases, higher data structure will not be used.
181 	uint64_t lower = 0;
182 	std::unordered_set<uint32_t> higher;
183 };
184 
185 // Helper template to avoid lots of nasty string temporary munging.
186 template <typename... Ts>
join(Ts &&...ts)187 std::string join(Ts &&... ts)
188 {
189 	StringStream<> stream;
190 	inner::join_helper(stream, std::forward<Ts>(ts)...);
191 	return stream.str();
192 }
193 
merge(const SmallVector<std::string> & list,const char * between=", ")194 inline std::string merge(const SmallVector<std::string> &list, const char *between = ", ")
195 {
196 	StringStream<> stream;
197 	for (auto &elem : list)
198 	{
199 		stream << elem;
200 		if (&elem != &list.back())
201 			stream << between;
202 	}
203 	return stream.str();
204 }
205 
206 // Make sure we don't accidentally call this with float or doubles with SFINAE.
207 // Have to use the radix-aware overload.
208 template <typename T, typename std::enable_if<!std::is_floating_point<T>::value, int>::type = 0>
convert_to_string(const T & t)209 inline std::string convert_to_string(const T &t)
210 {
211 	return std::to_string(t);
212 }
213 
214 // Allow implementations to set a convenient standard precision
215 #ifndef SPIRV_CROSS_FLT_FMT
216 #define SPIRV_CROSS_FLT_FMT "%.32g"
217 #endif
218 
219 // Disable sprintf and strcat warnings.
220 // We cannot rely on snprintf and family existing because, ..., MSVC.
221 #if defined(__clang__) || defined(__GNUC__)
222 #pragma GCC diagnostic push
223 #pragma GCC diagnostic ignored "-Wdeprecated-declarations"
224 #elif defined(_MSC_VER)
225 #pragma warning(push)
226 #pragma warning(disable : 4996)
227 #endif
228 
fixup_radix_point(char * str,char radix_point)229 static inline void fixup_radix_point(char *str, char radix_point)
230 {
231 	// Setting locales is a very risky business in multi-threaded program,
232 	// so just fixup locales instead. We only need to care about the radix point.
233 	if (radix_point != '.')
234 	{
235 		while (*str != '\0')
236 		{
237 			if (*str == radix_point)
238 				*str = '.';
239 			str++;
240 		}
241 	}
242 }
243 
convert_to_string(float t,char locale_radix_point)244 inline std::string convert_to_string(float t, char locale_radix_point)
245 {
246 	// std::to_string for floating point values is broken.
247 	// Fallback to something more sane.
248 	char buf[64];
249 	sprintf(buf, SPIRV_CROSS_FLT_FMT, t);
250 	fixup_radix_point(buf, locale_radix_point);
251 
252 	// Ensure that the literal is float.
253 	if (!strchr(buf, '.') && !strchr(buf, 'e'))
254 		strcat(buf, ".0");
255 	return buf;
256 }
257 
convert_to_string(double t,char locale_radix_point)258 inline std::string convert_to_string(double t, char locale_radix_point)
259 {
260 	// std::to_string for floating point values is broken.
261 	// Fallback to something more sane.
262 	char buf[64];
263 	sprintf(buf, SPIRV_CROSS_FLT_FMT, t);
264 	fixup_radix_point(buf, locale_radix_point);
265 
266 	// Ensure that the literal is float.
267 	if (!strchr(buf, '.') && !strchr(buf, 'e'))
268 		strcat(buf, ".0");
269 	return buf;
270 }
271 
272 template <typename T>
273 struct ValueSaver
274 {
ValueSaverSPIRV_CROSS_NAMESPACE::ValueSaver275 	explicit ValueSaver(T &current_)
276 	    : current(current_)
277 	    , saved(current_)
278 	{
279 	}
280 
releaseSPIRV_CROSS_NAMESPACE::ValueSaver281 	void release()
282 	{
283 		current = saved;
284 	}
285 
~ValueSaverSPIRV_CROSS_NAMESPACE::ValueSaver286 	~ValueSaver()
287 	{
288 		release();
289 	}
290 
291 	T &current;
292 	T saved;
293 };
294 
295 #if defined(__clang__) || defined(__GNUC__)
296 #pragma GCC diagnostic pop
297 #elif defined(_MSC_VER)
298 #pragma warning(pop)
299 #endif
300 
301 struct Instruction
302 {
303 	uint16_t op = 0;
304 	uint16_t count = 0;
305 	// If offset is 0 (not a valid offset into the instruction stream),
306 	// we have an instruction stream which is embedded in the object.
307 	uint32_t offset = 0;
308 	uint32_t length = 0;
309 
is_embeddedSPIRV_CROSS_NAMESPACE::Instruction310 	inline bool is_embedded() const
311 	{
312 		return offset == 0;
313 	}
314 };
315 
316 struct EmbeddedInstruction : Instruction
317 {
318 	SmallVector<uint32_t> ops;
319 };
320 
321 enum Types
322 {
323 	TypeNone,
324 	TypeType,
325 	TypeVariable,
326 	TypeConstant,
327 	TypeFunction,
328 	TypeFunctionPrototype,
329 	TypeBlock,
330 	TypeExtension,
331 	TypeExpression,
332 	TypeConstantOp,
333 	TypeCombinedImageSampler,
334 	TypeAccessChain,
335 	TypeUndef,
336 	TypeString,
337 	TypeCount
338 };
339 
340 template <Types type>
341 class TypedID;
342 
343 template <>
344 class TypedID<TypeNone>
345 {
346 public:
347 	TypedID() = default;
TypedID(uint32_t id_)348 	TypedID(uint32_t id_)
349 	    : id(id_)
350 	{
351 	}
352 
353 	template <Types U>
TypedID(const TypedID<U> & other)354 	TypedID(const TypedID<U> &other)
355 	{
356 		*this = other;
357 	}
358 
359 	template <Types U>
operator =(const TypedID<U> & other)360 	TypedID &operator=(const TypedID<U> &other)
361 	{
362 		id = uint32_t(other);
363 		return *this;
364 	}
365 
366 	// Implicit conversion to u32 is desired here.
367 	// As long as we block implicit conversion between TypedID<A> and TypedID<B> we're good.
operator uint32_t() const368 	operator uint32_t() const
369 	{
370 		return id;
371 	}
372 
373 	template <Types U>
operator TypedID<U>() const374 	operator TypedID<U>() const
375 	{
376 		return TypedID<U>(*this);
377 	}
378 
379 private:
380 	uint32_t id = 0;
381 };
382 
383 template <Types type>
384 class TypedID
385 {
386 public:
387 	TypedID() = default;
TypedID(uint32_t id_)388 	TypedID(uint32_t id_)
389 	    : id(id_)
390 	{
391 	}
392 
TypedID(const TypedID<TypeNone> & other)393 	explicit TypedID(const TypedID<TypeNone> &other)
394 	    : id(uint32_t(other))
395 	{
396 	}
397 
operator uint32_t() const398 	operator uint32_t() const
399 	{
400 		return id;
401 	}
402 
403 private:
404 	uint32_t id = 0;
405 };
406 
407 using VariableID = TypedID<TypeVariable>;
408 using TypeID = TypedID<TypeType>;
409 using ConstantID = TypedID<TypeConstant>;
410 using FunctionID = TypedID<TypeFunction>;
411 using BlockID = TypedID<TypeBlock>;
412 using ID = TypedID<TypeNone>;
413 
414 // Helper for Variant interface.
415 struct IVariant
416 {
417 	virtual ~IVariant() = default;
418 	virtual IVariant *clone(ObjectPoolBase *pool) = 0;
419 	ID self = 0;
420 
421 protected:
422 	IVariant() = default;
423 	IVariant(const IVariant&) = default;
424 	IVariant &operator=(const IVariant&) = default;
425 };
426 
427 #define SPIRV_CROSS_DECLARE_CLONE(T)                                \
428 	IVariant *clone(ObjectPoolBase *pool) override                  \
429 	{                                                               \
430 		return static_cast<ObjectPool<T> *>(pool)->allocate(*this); \
431 	}
432 
433 struct SPIRUndef : IVariant
434 {
435 	enum
436 	{
437 		type = TypeUndef
438 	};
439 
SPIRUndefSPIRV_CROSS_NAMESPACE::SPIRUndef440 	explicit SPIRUndef(TypeID basetype_)
441 	    : basetype(basetype_)
442 	{
443 	}
444 	TypeID basetype;
445 
446 	SPIRV_CROSS_DECLARE_CLONE(SPIRUndef)
447 };
448 
449 struct SPIRString : IVariant
450 {
451 	enum
452 	{
453 		type = TypeString
454 	};
455 
SPIRStringSPIRV_CROSS_NAMESPACE::SPIRString456 	explicit SPIRString(std::string str_)
457 	    : str(std::move(str_))
458 	{
459 	}
460 
461 	std::string str;
462 
463 	SPIRV_CROSS_DECLARE_CLONE(SPIRString)
464 };
465 
466 // This type is only used by backends which need to access the combined image and sampler IDs separately after
467 // the OpSampledImage opcode.
468 struct SPIRCombinedImageSampler : IVariant
469 {
470 	enum
471 	{
472 		type = TypeCombinedImageSampler
473 	};
SPIRCombinedImageSamplerSPIRV_CROSS_NAMESPACE::SPIRCombinedImageSampler474 	SPIRCombinedImageSampler(TypeID type_, VariableID image_, VariableID sampler_)
475 	    : combined_type(type_)
476 	    , image(image_)
477 	    , sampler(sampler_)
478 	{
479 	}
480 	TypeID combined_type;
481 	VariableID image;
482 	VariableID sampler;
483 
484 	SPIRV_CROSS_DECLARE_CLONE(SPIRCombinedImageSampler)
485 };
486 
487 struct SPIRConstantOp : IVariant
488 {
489 	enum
490 	{
491 		type = TypeConstantOp
492 	};
493 
SPIRConstantOpSPIRV_CROSS_NAMESPACE::SPIRConstantOp494 	SPIRConstantOp(TypeID result_type, spv::Op op, const uint32_t *args, uint32_t length)
495 	    : opcode(op)
496 	    , basetype(result_type)
497 	{
498 		arguments.reserve(length);
499 		for (uint32_t i = 0; i < length; i++)
500 			arguments.push_back(args[i]);
501 	}
502 
503 	spv::Op opcode;
504 	SmallVector<uint32_t> arguments;
505 	TypeID basetype;
506 
507 	SPIRV_CROSS_DECLARE_CLONE(SPIRConstantOp)
508 };
509 
510 struct SPIRType : IVariant
511 {
512 	enum
513 	{
514 		type = TypeType
515 	};
516 
517 	enum BaseType
518 	{
519 		Unknown,
520 		Void,
521 		Boolean,
522 		SByte,
523 		UByte,
524 		Short,
525 		UShort,
526 		Int,
527 		UInt,
528 		Int64,
529 		UInt64,
530 		AtomicCounter,
531 		Half,
532 		Float,
533 		Double,
534 		Struct,
535 		Image,
536 		SampledImage,
537 		Sampler,
538 		AccelerationStructure,
539 		RayQuery,
540 
541 		// Keep internal types at the end.
542 		ControlPointArray,
543 		Interpolant,
544 		Char
545 	};
546 
547 	// Scalar/vector/matrix support.
548 	BaseType basetype = Unknown;
549 	uint32_t width = 0;
550 	uint32_t vecsize = 1;
551 	uint32_t columns = 1;
552 
553 	// Arrays, support array of arrays by having a vector of array sizes.
554 	SmallVector<uint32_t> array;
555 
556 	// Array elements can be either specialization constants or specialization ops.
557 	// This array determines how to interpret the array size.
558 	// If an element is true, the element is a literal,
559 	// otherwise, it's an expression, which must be resolved on demand.
560 	// The actual size is not really known until runtime.
561 	SmallVector<bool> array_size_literal;
562 
563 	// Pointers
564 	// Keep track of how many pointer layers we have.
565 	uint32_t pointer_depth = 0;
566 	bool pointer = false;
567 	bool forward_pointer = false;
568 
569 	spv::StorageClass storage = spv::StorageClassGeneric;
570 
571 	SmallVector<TypeID> member_types;
572 
573 	// If member order has been rewritten to handle certain scenarios with Offset,
574 	// allow codegen to rewrite the index.
575 	SmallVector<uint32_t> member_type_index_redirection;
576 
577 	struct ImageType
578 	{
579 		TypeID type;
580 		spv::Dim dim;
581 		bool depth;
582 		bool arrayed;
583 		bool ms;
584 		uint32_t sampled;
585 		spv::ImageFormat format;
586 		spv::AccessQualifier access;
587 	} image;
588 
589 	// Structs can be declared multiple times if they are used as part of interface blocks.
590 	// We want to detect this so that we only emit the struct definition once.
591 	// Since we cannot rely on OpName to be equal, we need to figure out aliases.
592 	TypeID type_alias = 0;
593 
594 	// Denotes the type which this type is based on.
595 	// Allows the backend to traverse how a complex type is built up during access chains.
596 	TypeID parent_type = 0;
597 
598 	// Used in backends to avoid emitting members with conflicting names.
599 	std::unordered_set<std::string> member_name_cache;
600 
601 	SPIRV_CROSS_DECLARE_CLONE(SPIRType)
602 };
603 
604 struct SPIRExtension : IVariant
605 {
606 	enum
607 	{
608 		type = TypeExtension
609 	};
610 
611 	enum Extension
612 	{
613 		Unsupported,
614 		GLSL,
615 		SPV_debug_info,
616 		SPV_AMD_shader_ballot,
617 		SPV_AMD_shader_explicit_vertex_parameter,
618 		SPV_AMD_shader_trinary_minmax,
619 		SPV_AMD_gcn_shader
620 	};
621 
SPIRExtensionSPIRV_CROSS_NAMESPACE::SPIRExtension622 	explicit SPIRExtension(Extension ext_)
623 	    : ext(ext_)
624 	{
625 	}
626 
627 	Extension ext;
628 	SPIRV_CROSS_DECLARE_CLONE(SPIRExtension)
629 };
630 
631 // SPIREntryPoint is not a variant since its IDs are used to decorate OpFunction,
632 // so in order to avoid conflicts, we can't stick them in the ids array.
633 struct SPIREntryPoint
634 {
SPIREntryPointSPIRV_CROSS_NAMESPACE::SPIREntryPoint635 	SPIREntryPoint(FunctionID self_, spv::ExecutionModel execution_model, const std::string &entry_name)
636 	    : self(self_)
637 	    , name(entry_name)
638 	    , orig_name(entry_name)
639 	    , model(execution_model)
640 	{
641 	}
642 	SPIREntryPoint() = default;
643 
644 	FunctionID self = 0;
645 	std::string name;
646 	std::string orig_name;
647 	SmallVector<VariableID> interface_variables;
648 
649 	Bitset flags;
650 	struct WorkgroupSize
651 	{
652 		uint32_t x = 0, y = 0, z = 0;
653 		uint32_t constant = 0; // Workgroup size can be expressed as a constant/spec-constant instead.
654 	} workgroup_size;
655 	uint32_t invocations = 0;
656 	uint32_t output_vertices = 0;
657 	spv::ExecutionModel model = spv::ExecutionModelMax;
658 	bool geometry_passthrough = false;
659 };
660 
661 struct SPIRExpression : IVariant
662 {
663 	enum
664 	{
665 		type = TypeExpression
666 	};
667 
668 	// Only created by the backend target to avoid creating tons of temporaries.
SPIRExpressionSPIRV_CROSS_NAMESPACE::SPIRExpression669 	SPIRExpression(std::string expr, TypeID expression_type_, bool immutable_)
670 	    : expression(move(expr))
671 	    , expression_type(expression_type_)
672 	    , immutable(immutable_)
673 	{
674 	}
675 
676 	// If non-zero, prepend expression with to_expression(base_expression).
677 	// Used in amortizing multiple calls to to_expression()
678 	// where in certain cases that would quickly force a temporary when not needed.
679 	ID base_expression = 0;
680 
681 	std::string expression;
682 	TypeID expression_type = 0;
683 
684 	// If this expression is a forwarded load,
685 	// allow us to reference the original variable.
686 	ID loaded_from = 0;
687 
688 	// If this expression will never change, we can avoid lots of temporaries
689 	// in high level source.
690 	// An expression being immutable can be speculative,
691 	// it is assumed that this is true almost always.
692 	bool immutable = false;
693 
694 	// Before use, this expression must be transposed.
695 	// This is needed for targets which don't support row_major layouts.
696 	bool need_transpose = false;
697 
698 	// Whether or not this is an access chain expression.
699 	bool access_chain = false;
700 
701 	// A list of expressions which this expression depends on.
702 	SmallVector<ID> expression_dependencies;
703 
704 	// By reading this expression, we implicitly read these expressions as well.
705 	// Used by access chain Store and Load since we read multiple expressions in this case.
706 	SmallVector<ID> implied_read_expressions;
707 
708 	// The expression was emitted at a certain scope. Lets us track when an expression read means multiple reads.
709 	uint32_t emitted_loop_level = 0;
710 
711 	SPIRV_CROSS_DECLARE_CLONE(SPIRExpression)
712 };
713 
714 struct SPIRFunctionPrototype : IVariant
715 {
716 	enum
717 	{
718 		type = TypeFunctionPrototype
719 	};
720 
SPIRFunctionPrototypeSPIRV_CROSS_NAMESPACE::SPIRFunctionPrototype721 	explicit SPIRFunctionPrototype(TypeID return_type_)
722 	    : return_type(return_type_)
723 	{
724 	}
725 
726 	TypeID return_type;
727 	SmallVector<uint32_t> parameter_types;
728 
729 	SPIRV_CROSS_DECLARE_CLONE(SPIRFunctionPrototype)
730 };
731 
732 struct SPIRBlock : IVariant
733 {
734 	enum
735 	{
736 		type = TypeBlock
737 	};
738 
739 	enum Terminator
740 	{
741 		Unknown,
742 		Direct, // Emit next block directly without a particular condition.
743 
744 		Select, // Block ends with an if/else block.
745 		MultiSelect, // Block ends with switch statement.
746 
747 		Return, // Block ends with return.
748 		Unreachable, // Noop
749 		Kill, // Discard
750 		IgnoreIntersection, // Ray Tracing
751 		TerminateRay // Ray Tracing
752 	};
753 
754 	enum Merge
755 	{
756 		MergeNone,
757 		MergeLoop,
758 		MergeSelection
759 	};
760 
761 	enum Hints
762 	{
763 		HintNone,
764 		HintUnroll,
765 		HintDontUnroll,
766 		HintFlatten,
767 		HintDontFlatten
768 	};
769 
770 	enum Method
771 	{
772 		MergeToSelectForLoop,
773 		MergeToDirectForLoop,
774 		MergeToSelectContinueForLoop
775 	};
776 
777 	enum ContinueBlockType
778 	{
779 		ContinueNone,
780 
781 		// Continue block is branchless and has at least one instruction.
782 		ForLoop,
783 
784 		// Noop continue block.
785 		WhileLoop,
786 
787 		// Continue block is conditional.
788 		DoWhileLoop,
789 
790 		// Highly unlikely that anything will use this,
791 		// since it is really awkward/impossible to express in GLSL.
792 		ComplexLoop
793 	};
794 
795 	enum : uint32_t
796 	{
797 		NoDominator = 0xffffffffu
798 	};
799 
800 	Terminator terminator = Unknown;
801 	Merge merge = MergeNone;
802 	Hints hint = HintNone;
803 	BlockID next_block = 0;
804 	BlockID merge_block = 0;
805 	BlockID continue_block = 0;
806 
807 	ID return_value = 0; // If 0, return nothing (void).
808 	ID condition = 0;
809 	BlockID true_block = 0;
810 	BlockID false_block = 0;
811 	BlockID default_block = 0;
812 
813 	SmallVector<Instruction> ops;
814 
815 	struct Phi
816 	{
817 		ID local_variable; // flush local variable ...
818 		BlockID parent; // If we're in from_block and want to branch into this block ...
819 		VariableID function_variable; // to this function-global "phi" variable first.
820 	};
821 
822 	// Before entering this block flush out local variables to magical "phi" variables.
823 	SmallVector<Phi> phi_variables;
824 
825 	// Declare these temporaries before beginning the block.
826 	// Used for handling complex continue blocks which have side effects.
827 	SmallVector<std::pair<TypeID, ID>> declare_temporary;
828 
829 	// Declare these temporaries, but only conditionally if this block turns out to be
830 	// a complex loop header.
831 	SmallVector<std::pair<TypeID, ID>> potential_declare_temporary;
832 
833 	struct Case
834 	{
835 		uint32_t value;
836 		BlockID block;
837 	};
838 	SmallVector<Case> cases;
839 
840 	// If we have tried to optimize code for this block but failed,
841 	// keep track of this.
842 	bool disable_block_optimization = false;
843 
844 	// If the continue block is complex, fallback to "dumb" for loops.
845 	bool complex_continue = false;
846 
847 	// Do we need a ladder variable to defer breaking out of a loop construct after a switch block?
848 	bool need_ladder_break = false;
849 
850 	// If marked, we have explicitly handled Phi from this block, so skip any flushes related to that on a branch.
851 	// Used to handle an edge case with switch and case-label fallthrough where fall-through writes to Phi.
852 	BlockID ignore_phi_from_block = 0;
853 
854 	// The dominating block which this block might be within.
855 	// Used in continue; blocks to determine if we really need to write continue.
856 	BlockID loop_dominator = 0;
857 
858 	// All access to these variables are dominated by this block,
859 	// so before branching anywhere we need to make sure that we declare these variables.
860 	SmallVector<VariableID> dominated_variables;
861 
862 	// These are variables which should be declared in a for loop header, if we
863 	// fail to use a classic for-loop,
864 	// we remove these variables, and fall back to regular variables outside the loop.
865 	SmallVector<VariableID> loop_variables;
866 
867 	// Some expressions are control-flow dependent, i.e. any instruction which relies on derivatives or
868 	// sub-group-like operations.
869 	// Make sure that we only use these expressions in the original block.
870 	SmallVector<ID> invalidate_expressions;
871 
872 	SPIRV_CROSS_DECLARE_CLONE(SPIRBlock)
873 };
874 
875 struct SPIRFunction : IVariant
876 {
877 	enum
878 	{
879 		type = TypeFunction
880 	};
881 
SPIRFunctionSPIRV_CROSS_NAMESPACE::SPIRFunction882 	SPIRFunction(TypeID return_type_, TypeID function_type_)
883 	    : return_type(return_type_)
884 	    , function_type(function_type_)
885 	{
886 	}
887 
888 	struct Parameter
889 	{
890 		TypeID type;
891 		ID id;
892 		uint32_t read_count;
893 		uint32_t write_count;
894 
895 		// Set to true if this parameter aliases a global variable,
896 		// used mostly in Metal where global variables
897 		// have to be passed down to functions as regular arguments.
898 		// However, for this kind of variable, we should not care about
899 		// read and write counts as access to the function arguments
900 		// is not local to the function in question.
901 		bool alias_global_variable;
902 	};
903 
904 	// When calling a function, and we're remapping separate image samplers,
905 	// resolve these arguments into combined image samplers and pass them
906 	// as additional arguments in this order.
907 	// It gets more complicated as functions can pull in their own globals
908 	// and combine them with parameters,
909 	// so we need to distinguish if something is local parameter index
910 	// or a global ID.
911 	struct CombinedImageSamplerParameter
912 	{
913 		VariableID id;
914 		VariableID image_id;
915 		VariableID sampler_id;
916 		bool global_image;
917 		bool global_sampler;
918 		bool depth;
919 	};
920 
921 	TypeID return_type;
922 	TypeID function_type;
923 	SmallVector<Parameter> arguments;
924 
925 	// Can be used by backends to add magic arguments.
926 	// Currently used by combined image/sampler implementation.
927 
928 	SmallVector<Parameter> shadow_arguments;
929 	SmallVector<VariableID> local_variables;
930 	BlockID entry_block = 0;
931 	SmallVector<BlockID> blocks;
932 	SmallVector<CombinedImageSamplerParameter> combined_parameters;
933 
934 	struct EntryLine
935 	{
936 		uint32_t file_id = 0;
937 		uint32_t line_literal = 0;
938 	};
939 	EntryLine entry_line;
940 
add_local_variableSPIRV_CROSS_NAMESPACE::SPIRFunction941 	void add_local_variable(VariableID id)
942 	{
943 		local_variables.push_back(id);
944 	}
945 
add_parameterSPIRV_CROSS_NAMESPACE::SPIRFunction946 	void add_parameter(TypeID parameter_type, ID id, bool alias_global_variable = false)
947 	{
948 		// Arguments are read-only until proven otherwise.
949 		arguments.push_back({ parameter_type, id, 0u, 0u, alias_global_variable });
950 	}
951 
952 	// Hooks to be run when the function returns.
953 	// Mostly used for lowering internal data structures onto flattened structures.
954 	// Need to defer this, because they might rely on things which change during compilation.
955 	// Intentionally not a small vector, this one is rare, and std::function can be large.
956 	Vector<std::function<void()>> fixup_hooks_out;
957 
958 	// Hooks to be run when the function begins.
959 	// Mostly used for populating internal data structures from flattened structures.
960 	// Need to defer this, because they might rely on things which change during compilation.
961 	// Intentionally not a small vector, this one is rare, and std::function can be large.
962 	Vector<std::function<void()>> fixup_hooks_in;
963 
964 	// On function entry, make sure to copy a constant array into thread addr space to work around
965 	// the case where we are passing a constant array by value to a function on backends which do not
966 	// consider arrays value types.
967 	SmallVector<ID> constant_arrays_needed_on_stack;
968 
969 	bool active = false;
970 	bool flush_undeclared = true;
971 	bool do_combined_parameters = true;
972 
973 	SPIRV_CROSS_DECLARE_CLONE(SPIRFunction)
974 };
975 
976 struct SPIRAccessChain : IVariant
977 {
978 	enum
979 	{
980 		type = TypeAccessChain
981 	};
982 
SPIRAccessChainSPIRV_CROSS_NAMESPACE::SPIRAccessChain983 	SPIRAccessChain(TypeID basetype_, spv::StorageClass storage_, std::string base_, std::string dynamic_index_,
984 	                int32_t static_index_)
985 	    : basetype(basetype_)
986 	    , storage(storage_)
987 	    , base(std::move(base_))
988 	    , dynamic_index(std::move(dynamic_index_))
989 	    , static_index(static_index_)
990 	{
991 	}
992 
993 	// The access chain represents an offset into a buffer.
994 	// Some backends need more complicated handling of access chains to be able to use buffers, like HLSL
995 	// which has no usable buffer type ala GLSL SSBOs.
996 	// StructuredBuffer is too limited, so our only option is to deal with ByteAddressBuffer which works with raw addresses.
997 
998 	TypeID basetype;
999 	spv::StorageClass storage;
1000 	std::string base;
1001 	std::string dynamic_index;
1002 	int32_t static_index;
1003 
1004 	VariableID loaded_from = 0;
1005 	uint32_t matrix_stride = 0;
1006 	uint32_t array_stride = 0;
1007 	bool row_major_matrix = false;
1008 	bool immutable = false;
1009 
1010 	// By reading this expression, we implicitly read these expressions as well.
1011 	// Used by access chain Store and Load since we read multiple expressions in this case.
1012 	SmallVector<ID> implied_read_expressions;
1013 
1014 	SPIRV_CROSS_DECLARE_CLONE(SPIRAccessChain)
1015 };
1016 
1017 struct SPIRVariable : IVariant
1018 {
1019 	enum
1020 	{
1021 		type = TypeVariable
1022 	};
1023 
1024 	SPIRVariable() = default;
SPIRVariableSPIRV_CROSS_NAMESPACE::SPIRVariable1025 	SPIRVariable(TypeID basetype_, spv::StorageClass storage_, ID initializer_ = 0, VariableID basevariable_ = 0)
1026 	    : basetype(basetype_)
1027 	    , storage(storage_)
1028 	    , initializer(initializer_)
1029 	    , basevariable(basevariable_)
1030 	{
1031 	}
1032 
1033 	TypeID basetype = 0;
1034 	spv::StorageClass storage = spv::StorageClassGeneric;
1035 	uint32_t decoration = 0;
1036 	ID initializer = 0;
1037 	VariableID basevariable = 0;
1038 
1039 	SmallVector<uint32_t> dereference_chain;
1040 	bool compat_builtin = false;
1041 
1042 	// If a variable is shadowed, we only statically assign to it
1043 	// and never actually emit a statement for it.
1044 	// When we read the variable as an expression, just forward
1045 	// shadowed_id as the expression.
1046 	bool statically_assigned = false;
1047 	ID static_expression = 0;
1048 
1049 	// Temporaries which can remain forwarded as long as this variable is not modified.
1050 	SmallVector<ID> dependees;
1051 	bool forwardable = true;
1052 
1053 	bool deferred_declaration = false;
1054 	bool phi_variable = false;
1055 
1056 	// Used to deal with Phi variable flushes. See flush_phi().
1057 	bool allocate_temporary_copy = false;
1058 
1059 	bool remapped_variable = false;
1060 	uint32_t remapped_components = 0;
1061 
1062 	// The block which dominates all access to this variable.
1063 	BlockID dominator = 0;
1064 	// If true, this variable is a loop variable, when accessing the variable
1065 	// outside a loop,
1066 	// we should statically forward it.
1067 	bool loop_variable = false;
1068 	// Set to true while we're inside the for loop.
1069 	bool loop_variable_enable = false;
1070 
1071 	SPIRFunction::Parameter *parameter = nullptr;
1072 
1073 	SPIRV_CROSS_DECLARE_CLONE(SPIRVariable)
1074 };
1075 
1076 struct SPIRConstant : IVariant
1077 {
1078 	enum
1079 	{
1080 		type = TypeConstant
1081 	};
1082 
1083 	union Constant
1084 	{
1085 		uint32_t u32;
1086 		int32_t i32;
1087 		float f32;
1088 
1089 		uint64_t u64;
1090 		int64_t i64;
1091 		double f64;
1092 	};
1093 
1094 	struct ConstantVector
1095 	{
1096 		Constant r[4];
1097 		// If != 0, this element is a specialization constant, and we should keep track of it as such.
1098 		ID id[4];
1099 		uint32_t vecsize = 1;
1100 
ConstantVectorSPIRV_CROSS_NAMESPACE::SPIRConstant::ConstantVector1101 		ConstantVector()
1102 		{
1103 			memset(r, 0, sizeof(r));
1104 		}
1105 	};
1106 
1107 	struct ConstantMatrix
1108 	{
1109 		ConstantVector c[4];
1110 		// If != 0, this column is a specialization constant, and we should keep track of it as such.
1111 		ID id[4];
1112 		uint32_t columns = 1;
1113 	};
1114 
f16_to_f32SPIRV_CROSS_NAMESPACE::SPIRConstant1115 	static inline float f16_to_f32(uint16_t u16_value)
1116 	{
1117 		// Based on the GLM implementation.
1118 		int s = (u16_value >> 15) & 0x1;
1119 		int e = (u16_value >> 10) & 0x1f;
1120 		int m = (u16_value >> 0) & 0x3ff;
1121 
1122 		union
1123 		{
1124 			float f32;
1125 			uint32_t u32;
1126 		} u;
1127 
1128 		if (e == 0)
1129 		{
1130 			if (m == 0)
1131 			{
1132 				u.u32 = uint32_t(s) << 31;
1133 				return u.f32;
1134 			}
1135 			else
1136 			{
1137 				while ((m & 0x400) == 0)
1138 				{
1139 					m <<= 1;
1140 					e--;
1141 				}
1142 
1143 				e++;
1144 				m &= ~0x400;
1145 			}
1146 		}
1147 		else if (e == 31)
1148 		{
1149 			if (m == 0)
1150 			{
1151 				u.u32 = (uint32_t(s) << 31) | 0x7f800000u;
1152 				return u.f32;
1153 			}
1154 			else
1155 			{
1156 				u.u32 = (uint32_t(s) << 31) | 0x7f800000u | (m << 13);
1157 				return u.f32;
1158 			}
1159 		}
1160 
1161 		e += 127 - 15;
1162 		m <<= 13;
1163 		u.u32 = (uint32_t(s) << 31) | (e << 23) | m;
1164 		return u.f32;
1165 	}
1166 
specialization_constant_idSPIRV_CROSS_NAMESPACE::SPIRConstant1167 	inline uint32_t specialization_constant_id(uint32_t col, uint32_t row) const
1168 	{
1169 		return m.c[col].id[row];
1170 	}
1171 
specialization_constant_idSPIRV_CROSS_NAMESPACE::SPIRConstant1172 	inline uint32_t specialization_constant_id(uint32_t col) const
1173 	{
1174 		return m.id[col];
1175 	}
1176 
scalarSPIRV_CROSS_NAMESPACE::SPIRConstant1177 	inline uint32_t scalar(uint32_t col = 0, uint32_t row = 0) const
1178 	{
1179 		return m.c[col].r[row].u32;
1180 	}
1181 
scalar_i16SPIRV_CROSS_NAMESPACE::SPIRConstant1182 	inline int16_t scalar_i16(uint32_t col = 0, uint32_t row = 0) const
1183 	{
1184 		return int16_t(m.c[col].r[row].u32 & 0xffffu);
1185 	}
1186 
scalar_u16SPIRV_CROSS_NAMESPACE::SPIRConstant1187 	inline uint16_t scalar_u16(uint32_t col = 0, uint32_t row = 0) const
1188 	{
1189 		return uint16_t(m.c[col].r[row].u32 & 0xffffu);
1190 	}
1191 
scalar_i8SPIRV_CROSS_NAMESPACE::SPIRConstant1192 	inline int8_t scalar_i8(uint32_t col = 0, uint32_t row = 0) const
1193 	{
1194 		return int8_t(m.c[col].r[row].u32 & 0xffu);
1195 	}
1196 
scalar_u8SPIRV_CROSS_NAMESPACE::SPIRConstant1197 	inline uint8_t scalar_u8(uint32_t col = 0, uint32_t row = 0) const
1198 	{
1199 		return uint8_t(m.c[col].r[row].u32 & 0xffu);
1200 	}
1201 
scalar_f16SPIRV_CROSS_NAMESPACE::SPIRConstant1202 	inline float scalar_f16(uint32_t col = 0, uint32_t row = 0) const
1203 	{
1204 		return f16_to_f32(scalar_u16(col, row));
1205 	}
1206 
scalar_f32SPIRV_CROSS_NAMESPACE::SPIRConstant1207 	inline float scalar_f32(uint32_t col = 0, uint32_t row = 0) const
1208 	{
1209 		return m.c[col].r[row].f32;
1210 	}
1211 
scalar_i32SPIRV_CROSS_NAMESPACE::SPIRConstant1212 	inline int32_t scalar_i32(uint32_t col = 0, uint32_t row = 0) const
1213 	{
1214 		return m.c[col].r[row].i32;
1215 	}
1216 
scalar_f64SPIRV_CROSS_NAMESPACE::SPIRConstant1217 	inline double scalar_f64(uint32_t col = 0, uint32_t row = 0) const
1218 	{
1219 		return m.c[col].r[row].f64;
1220 	}
1221 
scalar_i64SPIRV_CROSS_NAMESPACE::SPIRConstant1222 	inline int64_t scalar_i64(uint32_t col = 0, uint32_t row = 0) const
1223 	{
1224 		return m.c[col].r[row].i64;
1225 	}
1226 
scalar_u64SPIRV_CROSS_NAMESPACE::SPIRConstant1227 	inline uint64_t scalar_u64(uint32_t col = 0, uint32_t row = 0) const
1228 	{
1229 		return m.c[col].r[row].u64;
1230 	}
1231 
vectorSPIRV_CROSS_NAMESPACE::SPIRConstant1232 	inline const ConstantVector &vector() const
1233 	{
1234 		return m.c[0];
1235 	}
1236 
vector_sizeSPIRV_CROSS_NAMESPACE::SPIRConstant1237 	inline uint32_t vector_size() const
1238 	{
1239 		return m.c[0].vecsize;
1240 	}
1241 
columnsSPIRV_CROSS_NAMESPACE::SPIRConstant1242 	inline uint32_t columns() const
1243 	{
1244 		return m.columns;
1245 	}
1246 
make_nullSPIRV_CROSS_NAMESPACE::SPIRConstant1247 	inline void make_null(const SPIRType &constant_type_)
1248 	{
1249 		m = {};
1250 		m.columns = constant_type_.columns;
1251 		for (auto &c : m.c)
1252 			c.vecsize = constant_type_.vecsize;
1253 	}
1254 
constant_is_nullSPIRV_CROSS_NAMESPACE::SPIRConstant1255 	inline bool constant_is_null() const
1256 	{
1257 		if (specialization)
1258 			return false;
1259 		if (!subconstants.empty())
1260 			return false;
1261 
1262 		for (uint32_t col = 0; col < columns(); col++)
1263 			for (uint32_t row = 0; row < vector_size(); row++)
1264 				if (scalar_u64(col, row) != 0)
1265 					return false;
1266 
1267 		return true;
1268 	}
1269 
SPIRConstantSPIRV_CROSS_NAMESPACE::SPIRConstant1270 	explicit SPIRConstant(uint32_t constant_type_)
1271 	    : constant_type(constant_type_)
1272 	{
1273 	}
1274 
1275 	SPIRConstant() = default;
1276 
SPIRConstantSPIRV_CROSS_NAMESPACE::SPIRConstant1277 	SPIRConstant(TypeID constant_type_, const uint32_t *elements, uint32_t num_elements, bool specialized)
1278 	    : constant_type(constant_type_)
1279 	    , specialization(specialized)
1280 	{
1281 		subconstants.reserve(num_elements);
1282 		for (uint32_t i = 0; i < num_elements; i++)
1283 			subconstants.push_back(elements[i]);
1284 		specialization = specialized;
1285 	}
1286 
1287 	// Construct scalar (32-bit).
SPIRConstantSPIRV_CROSS_NAMESPACE::SPIRConstant1288 	SPIRConstant(TypeID constant_type_, uint32_t v0, bool specialized)
1289 	    : constant_type(constant_type_)
1290 	    , specialization(specialized)
1291 	{
1292 		m.c[0].r[0].u32 = v0;
1293 		m.c[0].vecsize = 1;
1294 		m.columns = 1;
1295 	}
1296 
1297 	// Construct scalar (64-bit).
SPIRConstantSPIRV_CROSS_NAMESPACE::SPIRConstant1298 	SPIRConstant(TypeID constant_type_, uint64_t v0, bool specialized)
1299 	    : constant_type(constant_type_)
1300 	    , specialization(specialized)
1301 	{
1302 		m.c[0].r[0].u64 = v0;
1303 		m.c[0].vecsize = 1;
1304 		m.columns = 1;
1305 	}
1306 
1307 	// Construct vectors and matrices.
SPIRConstantSPIRV_CROSS_NAMESPACE::SPIRConstant1308 	SPIRConstant(TypeID constant_type_, const SPIRConstant *const *vector_elements, uint32_t num_elements,
1309 	             bool specialized)
1310 	    : constant_type(constant_type_)
1311 	    , specialization(specialized)
1312 	{
1313 		bool matrix = vector_elements[0]->m.c[0].vecsize > 1;
1314 
1315 		if (matrix)
1316 		{
1317 			m.columns = num_elements;
1318 
1319 			for (uint32_t i = 0; i < num_elements; i++)
1320 			{
1321 				m.c[i] = vector_elements[i]->m.c[0];
1322 				if (vector_elements[i]->specialization)
1323 					m.id[i] = vector_elements[i]->self;
1324 			}
1325 		}
1326 		else
1327 		{
1328 			m.c[0].vecsize = num_elements;
1329 			m.columns = 1;
1330 
1331 			for (uint32_t i = 0; i < num_elements; i++)
1332 			{
1333 				m.c[0].r[i] = vector_elements[i]->m.c[0].r[0];
1334 				if (vector_elements[i]->specialization)
1335 					m.c[0].id[i] = vector_elements[i]->self;
1336 			}
1337 		}
1338 	}
1339 
1340 	TypeID constant_type = 0;
1341 	ConstantMatrix m;
1342 
1343 	// If this constant is a specialization constant (i.e. created with OpSpecConstant*).
1344 	bool specialization = false;
1345 	// If this constant is used as an array length which creates specialization restrictions on some backends.
1346 	bool is_used_as_array_length = false;
1347 
1348 	// If true, this is a LUT, and should always be declared in the outer scope.
1349 	bool is_used_as_lut = false;
1350 
1351 	// For composites which are constant arrays, etc.
1352 	SmallVector<ConstantID> subconstants;
1353 
1354 	// Non-Vulkan GLSL, HLSL and sometimes MSL emits defines for each specialization constant,
1355 	// and uses them to initialize the constant. This allows the user
1356 	// to still be able to specialize the value by supplying corresponding
1357 	// preprocessor directives before compiling the shader.
1358 	std::string specialization_constant_macro_name;
1359 
1360 	SPIRV_CROSS_DECLARE_CLONE(SPIRConstant)
1361 };
1362 
1363 // Variants have a very specific allocation scheme.
1364 struct ObjectPoolGroup
1365 {
1366 	std::unique_ptr<ObjectPoolBase> pools[TypeCount];
1367 };
1368 
1369 class Variant
1370 {
1371 public:
Variant(ObjectPoolGroup * group_)1372 	explicit Variant(ObjectPoolGroup *group_)
1373 	    : group(group_)
1374 	{
1375 	}
1376 
~Variant()1377 	~Variant()
1378 	{
1379 		if (holder)
1380 			group->pools[type]->free_opaque(holder);
1381 	}
1382 
1383 	// Marking custom move constructor as noexcept is important.
Variant(Variant && other)1384 	Variant(Variant &&other) SPIRV_CROSS_NOEXCEPT
1385 	{
1386 		*this = std::move(other);
1387 	}
1388 
1389 	// We cannot copy from other variant without our own pool group.
1390 	// Have to explicitly copy.
1391 	Variant(const Variant &variant) = delete;
1392 
1393 	// Marking custom move constructor as noexcept is important.
operator =(Variant && other)1394 	Variant &operator=(Variant &&other) SPIRV_CROSS_NOEXCEPT
1395 	{
1396 		if (this != &other)
1397 		{
1398 			if (holder)
1399 				group->pools[type]->free_opaque(holder);
1400 			holder = other.holder;
1401 			group = other.group;
1402 			type = other.type;
1403 			allow_type_rewrite = other.allow_type_rewrite;
1404 
1405 			other.holder = nullptr;
1406 			other.type = TypeNone;
1407 		}
1408 		return *this;
1409 	}
1410 
1411 	// This copy/clone should only be called in the Compiler constructor.
1412 	// If this is called inside ::compile(), we invalidate any references we took higher in the stack.
1413 	// This should never happen.
operator =(const Variant & other)1414 	Variant &operator=(const Variant &other)
1415 	{
1416 //#define SPIRV_CROSS_COPY_CONSTRUCTOR_SANITIZE
1417 #ifdef SPIRV_CROSS_COPY_CONSTRUCTOR_SANITIZE
1418 		abort();
1419 #endif
1420 		if (this != &other)
1421 		{
1422 			if (holder)
1423 				group->pools[type]->free_opaque(holder);
1424 
1425 			if (other.holder)
1426 				holder = other.holder->clone(group->pools[other.type].get());
1427 			else
1428 				holder = nullptr;
1429 
1430 			type = other.type;
1431 			allow_type_rewrite = other.allow_type_rewrite;
1432 		}
1433 		return *this;
1434 	}
1435 
set(IVariant * val,Types new_type)1436 	void set(IVariant *val, Types new_type)
1437 	{
1438 		if (holder)
1439 			group->pools[type]->free_opaque(holder);
1440 		holder = nullptr;
1441 
1442 		if (!allow_type_rewrite && type != TypeNone && type != new_type)
1443 		{
1444 			if (val)
1445 				group->pools[new_type]->free_opaque(val);
1446 			SPIRV_CROSS_THROW("Overwriting a variant with new type.");
1447 		}
1448 
1449 		holder = val;
1450 		type = new_type;
1451 		allow_type_rewrite = false;
1452 	}
1453 
1454 	template <typename T, typename... Ts>
allocate_and_set(Types new_type,Ts &&...ts)1455 	T *allocate_and_set(Types new_type, Ts &&... ts)
1456 	{
1457 		T *val = static_cast<ObjectPool<T> &>(*group->pools[new_type]).allocate(std::forward<Ts>(ts)...);
1458 		set(val, new_type);
1459 		return val;
1460 	}
1461 
1462 	template <typename T>
get()1463 	T &get()
1464 	{
1465 		if (!holder)
1466 			SPIRV_CROSS_THROW("nullptr");
1467 		if (static_cast<Types>(T::type) != type)
1468 			SPIRV_CROSS_THROW("Bad cast");
1469 		return *static_cast<T *>(holder);
1470 	}
1471 
1472 	template <typename T>
get() const1473 	const T &get() const
1474 	{
1475 		if (!holder)
1476 			SPIRV_CROSS_THROW("nullptr");
1477 		if (static_cast<Types>(T::type) != type)
1478 			SPIRV_CROSS_THROW("Bad cast");
1479 		return *static_cast<const T *>(holder);
1480 	}
1481 
get_type() const1482 	Types get_type() const
1483 	{
1484 		return type;
1485 	}
1486 
get_id() const1487 	ID get_id() const
1488 	{
1489 		return holder ? holder->self : ID(0);
1490 	}
1491 
empty() const1492 	bool empty() const
1493 	{
1494 		return !holder;
1495 	}
1496 
reset()1497 	void reset()
1498 	{
1499 		if (holder)
1500 			group->pools[type]->free_opaque(holder);
1501 		holder = nullptr;
1502 		type = TypeNone;
1503 	}
1504 
set_allow_type_rewrite()1505 	void set_allow_type_rewrite()
1506 	{
1507 		allow_type_rewrite = true;
1508 	}
1509 
1510 private:
1511 	ObjectPoolGroup *group = nullptr;
1512 	IVariant *holder = nullptr;
1513 	Types type = TypeNone;
1514 	bool allow_type_rewrite = false;
1515 };
1516 
1517 template <typename T>
variant_get(Variant & var)1518 T &variant_get(Variant &var)
1519 {
1520 	return var.get<T>();
1521 }
1522 
1523 template <typename T>
variant_get(const Variant & var)1524 const T &variant_get(const Variant &var)
1525 {
1526 	return var.get<T>();
1527 }
1528 
1529 template <typename T, typename... P>
variant_set(Variant & var,P &&...args)1530 T &variant_set(Variant &var, P &&... args)
1531 {
1532 	auto *ptr = var.allocate_and_set<T>(static_cast<Types>(T::type), std::forward<P>(args)...);
1533 	return *ptr;
1534 }
1535 
1536 struct AccessChainMeta
1537 {
1538 	uint32_t storage_physical_type = 0;
1539 	bool need_transpose = false;
1540 	bool storage_is_packed = false;
1541 	bool storage_is_invariant = false;
1542 	bool flattened_struct = false;
1543 };
1544 
1545 enum ExtendedDecorations
1546 {
1547 	// Marks if a buffer block is re-packed, i.e. member declaration might be subject to PhysicalTypeID remapping and padding.
1548 	SPIRVCrossDecorationBufferBlockRepacked = 0,
1549 
1550 	// A type in a buffer block might be declared with a different physical type than the logical type.
1551 	// If this is not set, PhysicalTypeID == the SPIR-V type as declared.
1552 	SPIRVCrossDecorationPhysicalTypeID,
1553 
1554 	// Marks if the physical type is to be declared with tight packing rules, i.e. packed_floatN on MSL and friends.
1555 	// If this is set, PhysicalTypeID might also be set. It can be set to same as logical type if all we're doing
1556 	// is converting float3 to packed_float3 for example.
1557 	// If this is marked on a struct, it means the struct itself must use only Packed types for all its members.
1558 	SPIRVCrossDecorationPhysicalTypePacked,
1559 
1560 	// The padding in bytes before declaring this struct member.
1561 	// If used on a struct type, marks the target size of a struct.
1562 	SPIRVCrossDecorationPaddingTarget,
1563 
1564 	SPIRVCrossDecorationInterfaceMemberIndex,
1565 	SPIRVCrossDecorationInterfaceOrigID,
1566 	SPIRVCrossDecorationResourceIndexPrimary,
1567 	// Used for decorations like resource indices for samplers when part of combined image samplers.
1568 	// A variable might need to hold two resource indices in this case.
1569 	SPIRVCrossDecorationResourceIndexSecondary,
1570 	// Used for resource indices for multiplanar images when part of combined image samplers.
1571 	SPIRVCrossDecorationResourceIndexTertiary,
1572 	SPIRVCrossDecorationResourceIndexQuaternary,
1573 
1574 	// Marks a buffer block for using explicit offsets (GLSL/HLSL).
1575 	SPIRVCrossDecorationExplicitOffset,
1576 
1577 	// Apply to a variable in the Input storage class; marks it as holding the base group passed to vkCmdDispatchBase(),
1578 	// or the base vertex and instance indices passed to vkCmdDrawIndexed().
1579 	// In MSL, this is used to adjust the WorkgroupId and GlobalInvocationId variables in compute shaders,
1580 	// and to hold the BaseVertex and BaseInstance variables in vertex shaders.
1581 	SPIRVCrossDecorationBuiltInDispatchBase,
1582 
1583 	// Apply to a variable that is a function parameter; marks it as being a "dynamic"
1584 	// combined image-sampler. In MSL, this is used when a function parameter might hold
1585 	// either a regular combined image-sampler or one that has an attached sampler
1586 	// Y'CbCr conversion.
1587 	SPIRVCrossDecorationDynamicImageSampler,
1588 
1589 	// Apply to a variable in the Input storage class; marks it as holding the size of the stage
1590 	// input grid.
1591 	// In MSL, this is used to hold the vertex and instance counts in a tessellation pipeline
1592 	// vertex shader.
1593 	SPIRVCrossDecorationBuiltInStageInputSize,
1594 
1595 	// Apply to any access chain of a tessellation I/O variable; stores the type of the sub-object
1596 	// that was chained to, as recorded in the input variable itself. This is used in case the pointer
1597 	// is itself used as the base of an access chain, to calculate the original type of the sub-object
1598 	// chained to, in case a swizzle needs to be applied. This should not happen normally with valid
1599 	// SPIR-V, but the MSL backend can change the type of input variables, necessitating the
1600 	// addition of swizzles to keep the generated code compiling.
1601 	SPIRVCrossDecorationTessIOOriginalInputTypeID,
1602 
1603 	// Apply to any access chain of an interface variable used with pull-model interpolation, where the variable is a
1604 	// vector but the resulting pointer is a scalar; stores the component index that is to be accessed by the chain.
1605 	// This is used when emitting calls to interpolation functions on the chain in MSL: in this case, the component
1606 	// must be applied to the result, since pull-model interpolants in MSL cannot be swizzled directly, but the
1607 	// results of interpolation can.
1608 	SPIRVCrossDecorationInterpolantComponentExpr,
1609 
1610 	SPIRVCrossDecorationCount
1611 };
1612 
1613 struct Meta
1614 {
1615 	struct Decoration
1616 	{
1617 		std::string alias;
1618 		std::string qualified_alias;
1619 		std::string hlsl_semantic;
1620 		Bitset decoration_flags;
1621 		spv::BuiltIn builtin_type = spv::BuiltInMax;
1622 		uint32_t location = 0;
1623 		uint32_t component = 0;
1624 		uint32_t set = 0;
1625 		uint32_t binding = 0;
1626 		uint32_t offset = 0;
1627 		uint32_t xfb_buffer = 0;
1628 		uint32_t xfb_stride = 0;
1629 		uint32_t stream = 0;
1630 		uint32_t array_stride = 0;
1631 		uint32_t matrix_stride = 0;
1632 		uint32_t input_attachment = 0;
1633 		uint32_t spec_id = 0;
1634 		uint32_t index = 0;
1635 		spv::FPRoundingMode fp_rounding_mode = spv::FPRoundingModeMax;
1636 		bool builtin = false;
1637 
1638 		struct Extended
1639 		{
ExtendedSPIRV_CROSS_NAMESPACE::Meta::Decoration::Extended1640 			Extended()
1641 			{
1642 				// MSVC 2013 workaround to init like this.
1643 				for (auto &v : values)
1644 					v = 0;
1645 			}
1646 
1647 			Bitset flags;
1648 			uint32_t values[SPIRVCrossDecorationCount];
1649 		} extended;
1650 	};
1651 
1652 	Decoration decoration;
1653 
1654 	// Intentionally not a SmallVector. Decoration is large and somewhat rare.
1655 	Vector<Decoration> members;
1656 
1657 	std::unordered_map<uint32_t, uint32_t> decoration_word_offset;
1658 
1659 	// For SPV_GOOGLE_hlsl_functionality1.
1660 	bool hlsl_is_magic_counter_buffer = false;
1661 	// ID for the sibling counter buffer.
1662 	uint32_t hlsl_magic_counter_buffer = 0;
1663 };
1664 
1665 // A user callback that remaps the type of any variable.
1666 // var_name is the declared name of the variable.
1667 // name_of_type is the textual name of the type which will be used in the code unless written to by the callback.
1668 using VariableTypeRemapCallback =
1669     std::function<void(const SPIRType &type, const std::string &var_name, std::string &name_of_type)>;
1670 
1671 class Hasher
1672 {
1673 public:
u32(uint32_t value)1674 	inline void u32(uint32_t value)
1675 	{
1676 		h = (h * 0x100000001b3ull) ^ value;
1677 	}
1678 
get() const1679 	inline uint64_t get() const
1680 	{
1681 		return h;
1682 	}
1683 
1684 private:
1685 	uint64_t h = 0xcbf29ce484222325ull;
1686 };
1687 
type_is_floating_point(const SPIRType & type)1688 static inline bool type_is_floating_point(const SPIRType &type)
1689 {
1690 	return type.basetype == SPIRType::Half || type.basetype == SPIRType::Float || type.basetype == SPIRType::Double;
1691 }
1692 
type_is_integral(const SPIRType & type)1693 static inline bool type_is_integral(const SPIRType &type)
1694 {
1695 	return type.basetype == SPIRType::SByte || type.basetype == SPIRType::UByte || type.basetype == SPIRType::Short ||
1696 	       type.basetype == SPIRType::UShort || type.basetype == SPIRType::Int || type.basetype == SPIRType::UInt ||
1697 	       type.basetype == SPIRType::Int64 || type.basetype == SPIRType::UInt64;
1698 }
1699 
to_signed_basetype(uint32_t width)1700 static inline SPIRType::BaseType to_signed_basetype(uint32_t width)
1701 {
1702 	switch (width)
1703 	{
1704 	case 8:
1705 		return SPIRType::SByte;
1706 	case 16:
1707 		return SPIRType::Short;
1708 	case 32:
1709 		return SPIRType::Int;
1710 	case 64:
1711 		return SPIRType::Int64;
1712 	default:
1713 		SPIRV_CROSS_THROW("Invalid bit width.");
1714 	}
1715 }
1716 
to_unsigned_basetype(uint32_t width)1717 static inline SPIRType::BaseType to_unsigned_basetype(uint32_t width)
1718 {
1719 	switch (width)
1720 	{
1721 	case 8:
1722 		return SPIRType::UByte;
1723 	case 16:
1724 		return SPIRType::UShort;
1725 	case 32:
1726 		return SPIRType::UInt;
1727 	case 64:
1728 		return SPIRType::UInt64;
1729 	default:
1730 		SPIRV_CROSS_THROW("Invalid bit width.");
1731 	}
1732 }
1733 
1734 // Returns true if an arithmetic operation does not change behavior depending on signedness.
opcode_is_sign_invariant(spv::Op opcode)1735 static inline bool opcode_is_sign_invariant(spv::Op opcode)
1736 {
1737 	switch (opcode)
1738 	{
1739 	case spv::OpIEqual:
1740 	case spv::OpINotEqual:
1741 	case spv::OpISub:
1742 	case spv::OpIAdd:
1743 	case spv::OpIMul:
1744 	case spv::OpShiftLeftLogical:
1745 	case spv::OpBitwiseOr:
1746 	case spv::OpBitwiseXor:
1747 	case spv::OpBitwiseAnd:
1748 		return true;
1749 
1750 	default:
1751 		return false;
1752 	}
1753 }
1754 
1755 struct SetBindingPair
1756 {
1757 	uint32_t desc_set;
1758 	uint32_t binding;
1759 
operator ==SPIRV_CROSS_NAMESPACE::SetBindingPair1760 	inline bool operator==(const SetBindingPair &other) const
1761 	{
1762 		return desc_set == other.desc_set && binding == other.binding;
1763 	}
1764 
operator <SPIRV_CROSS_NAMESPACE::SetBindingPair1765 	inline bool operator<(const SetBindingPair &other) const
1766 	{
1767 		return desc_set < other.desc_set || (desc_set == other.desc_set && binding < other.binding);
1768 	}
1769 };
1770 
1771 struct LocationComponentPair
1772 {
1773 	uint32_t location;
1774 	uint32_t component;
1775 
operator ==SPIRV_CROSS_NAMESPACE::LocationComponentPair1776 	inline bool operator==(const LocationComponentPair &other) const
1777 	{
1778 		return location == other.location && component == other.component;
1779 	}
1780 
operator <SPIRV_CROSS_NAMESPACE::LocationComponentPair1781 	inline bool operator<(const LocationComponentPair &other) const
1782 	{
1783 		return location < other.location || (location == other.location && component < other.component);
1784 	}
1785 };
1786 
1787 struct StageSetBinding
1788 {
1789 	spv::ExecutionModel model;
1790 	uint32_t desc_set;
1791 	uint32_t binding;
1792 
operator ==SPIRV_CROSS_NAMESPACE::StageSetBinding1793 	inline bool operator==(const StageSetBinding &other) const
1794 	{
1795 		return model == other.model && desc_set == other.desc_set && binding == other.binding;
1796 	}
1797 };
1798 
1799 struct InternalHasher
1800 {
operator ()SPIRV_CROSS_NAMESPACE::InternalHasher1801 	inline size_t operator()(const SetBindingPair &value) const
1802 	{
1803 		// Quality of hash doesn't really matter here.
1804 		auto hash_set = std::hash<uint32_t>()(value.desc_set);
1805 		auto hash_binding = std::hash<uint32_t>()(value.binding);
1806 		return (hash_set * 0x10001b31) ^ hash_binding;
1807 	}
1808 
operator ()SPIRV_CROSS_NAMESPACE::InternalHasher1809 	inline size_t operator()(const LocationComponentPair &value) const
1810 	{
1811 		// Quality of hash doesn't really matter here.
1812 		auto hash_set = std::hash<uint32_t>()(value.location);
1813 		auto hash_binding = std::hash<uint32_t>()(value.component);
1814 		return (hash_set * 0x10001b31) ^ hash_binding;
1815 	}
1816 
operator ()SPIRV_CROSS_NAMESPACE::InternalHasher1817 	inline size_t operator()(const StageSetBinding &value) const
1818 	{
1819 		// Quality of hash doesn't really matter here.
1820 		auto hash_model = std::hash<uint32_t>()(value.model);
1821 		auto hash_set = std::hash<uint32_t>()(value.desc_set);
1822 		auto tmp_hash = (hash_model * 0x10001b31) ^ hash_set;
1823 		return (tmp_hash * 0x10001b31) ^ value.binding;
1824 	}
1825 };
1826 
1827 // Special constant used in a {MSL,HLSL}ResourceBinding desc_set
1828 // element to indicate the bindings for the push constants.
1829 static const uint32_t ResourceBindingPushConstantDescriptorSet = ~(0u);
1830 
1831 // Special constant used in a {MSL,HLSL}ResourceBinding binding
1832 // element to indicate the bindings for the push constants.
1833 static const uint32_t ResourceBindingPushConstantBinding = 0;
1834 } // namespace SPIRV_CROSS_NAMESPACE
1835 
1836 namespace std
1837 {
1838 template <SPIRV_CROSS_NAMESPACE::Types type>
1839 struct hash<SPIRV_CROSS_NAMESPACE::TypedID<type>>
1840 {
operator ()std::hash1841 	size_t operator()(const SPIRV_CROSS_NAMESPACE::TypedID<type> &value) const
1842 	{
1843 		return std::hash<uint32_t>()(value);
1844 	}
1845 };
1846 } // namespace std
1847 
1848 #endif
1849