Merge pull request #1146 from KhronosGroup/strong-types
Introduce stronger types for the SPIR-V ID in C++ API
This commit is contained in:
commit
ee0ac3b313
@ -465,6 +465,10 @@ if (SPIRV_CROSS_CLI)
|
||||
target_link_libraries(spirv-cross-msl-ycbcr-conversion-test spirv-cross-c)
|
||||
set_target_properties(spirv-cross-msl-ycbcr-conversion-test PROPERTIES LINK_FLAGS "${spirv-cross-link-flags}")
|
||||
|
||||
add_executable(spirv-cross-typed-id-test tests-other/typed_id_test.cpp)
|
||||
target_link_libraries(spirv-cross-typed-id-test spirv-cross-core)
|
||||
set_target_properties(spirv-cross-typed-id-test PROPERTIES LINK_FLAGS "${spirv-cross-link-flags}")
|
||||
|
||||
if (CMAKE_COMPILER_IS_GNUCXX OR (${CMAKE_CXX_COMPILER_ID} MATCHES "Clang"))
|
||||
target_compile_options(spirv-cross-c-api-test PRIVATE -std=c89 -Wall -Wextra)
|
||||
endif()
|
||||
@ -483,6 +487,8 @@ if (SPIRV_CROSS_CLI)
|
||||
COMMAND $<TARGET_FILE:spirv-cross-msl-ycbcr-conversion-test> ${CMAKE_CURRENT_SOURCE_DIR}/tests-other/msl_ycbcr_conversion_test.spv)
|
||||
add_test(NAME spirv-cross-msl-ycbcr-conversion-test-2
|
||||
COMMAND $<TARGET_FILE:spirv-cross-msl-ycbcr-conversion-test> ${CMAKE_CURRENT_SOURCE_DIR}/tests-other/msl_ycbcr_conversion_test_2.spv)
|
||||
add_test(NAME spirv-cross-typed-id-test
|
||||
COMMAND $<TARGET_FILE:spirv-cross-typed-id-test>)
|
||||
add_test(NAME spirv-cross-test
|
||||
COMMAND ${PYTHON_EXECUTABLE} ${CMAKE_CURRENT_SOURCE_DIR}/test_shaders.py --parallel
|
||||
${spirv-cross-externals}
|
||||
|
6
main.cpp
6
main.cpp
@ -246,7 +246,7 @@ static void print_resources(const Compiler &compiler, const char *tag, const Sma
|
||||
compiler.get_decoration_bitset(type.self).get(DecorationBufferBlock);
|
||||
bool is_sized_block = is_block && (compiler.get_storage_class(res.id) == StorageClassUniform ||
|
||||
compiler.get_storage_class(res.id) == StorageClassUniformConstant);
|
||||
uint32_t fallback_id = !is_push_constant && is_block ? res.base_type_id : res.id;
|
||||
ID fallback_id = !is_push_constant && is_block ? ID(res.base_type_id) : ID(res.id);
|
||||
|
||||
uint32_t block_size = 0;
|
||||
uint32_t runtime_array_stride = 0;
|
||||
@ -268,7 +268,7 @@ static void print_resources(const Compiler &compiler, const char *tag, const Sma
|
||||
for (auto arr : type.array)
|
||||
array = join("[", arr ? convert_to_string(arr) : "", "]") + array;
|
||||
|
||||
fprintf(stderr, " ID %03u : %s%s", res.id,
|
||||
fprintf(stderr, " ID %03u : %s%s", uint32_t(res.id),
|
||||
!res.name.empty() ? res.name.c_str() : compiler.get_fallback_name(fallback_id).c_str(), array.c_str());
|
||||
|
||||
if (mask.get(DecorationLocation))
|
||||
@ -442,7 +442,7 @@ static void print_spec_constants(const Compiler &compiler)
|
||||
fprintf(stderr, "Specialization constants\n");
|
||||
fprintf(stderr, "==================\n\n");
|
||||
for (auto &c : spec_constants)
|
||||
fprintf(stderr, "ID: %u, Spec ID: %u\n", c.id, c.constant_id);
|
||||
fprintf(stderr, "ID: %u, Spec ID: %u\n", uint32_t(c.id), c.constant_id);
|
||||
fprintf(stderr, "==================\n\n");
|
||||
}
|
||||
|
||||
|
@ -237,13 +237,13 @@ uint32_t CFG::find_loop_dominator(uint32_t block_id) const
|
||||
for (auto &pred : itr->second)
|
||||
{
|
||||
auto &pred_block = compiler.get<SPIRBlock>(pred);
|
||||
if (pred_block.merge == SPIRBlock::MergeLoop && pred_block.merge_block == block_id)
|
||||
if (pred_block.merge == SPIRBlock::MergeLoop && pred_block.merge_block == ID(block_id))
|
||||
{
|
||||
pred_block_id = pred;
|
||||
ignore_loop_header = true;
|
||||
break;
|
||||
}
|
||||
else if (pred_block.merge == SPIRBlock::MergeSelection && pred_block.next_block == block_id)
|
||||
else if (pred_block.merge == SPIRBlock::MergeSelection && pred_block.next_block == ID(block_id))
|
||||
{
|
||||
pred_block_id = pred;
|
||||
break;
|
||||
@ -268,14 +268,14 @@ uint32_t CFG::find_loop_dominator(uint32_t block_id) const
|
||||
return block_id;
|
||||
}
|
||||
|
||||
bool CFG::node_terminates_control_flow_in_sub_graph(uint32_t from, uint32_t to) const
|
||||
bool CFG::node_terminates_control_flow_in_sub_graph(BlockID from, BlockID to) const
|
||||
{
|
||||
// Walk backwards, starting from "to" block.
|
||||
// Only follow pred edges if they have a 1:1 relationship, or a merge relationship.
|
||||
// If we cannot find a path to "from", we must assume that to is inside control flow in some way.
|
||||
|
||||
auto &from_block = compiler.get<SPIRBlock>(from);
|
||||
uint32_t ignore_block_id = 0;
|
||||
BlockID ignore_block_id = 0;
|
||||
if (from_block.merge == SPIRBlock::MergeLoop)
|
||||
ignore_block_id = from_block.merge_block;
|
||||
|
||||
|
@ -97,7 +97,7 @@ public:
|
||||
|
||||
uint32_t find_loop_dominator(uint32_t block) const;
|
||||
|
||||
bool node_terminates_control_flow_in_sub_graph(uint32_t from, uint32_t to) const;
|
||||
bool node_terminates_control_flow_in_sub_graph(BlockID from, BlockID to) const;
|
||||
|
||||
private:
|
||||
struct VisitOrder
|
||||
|
329
spirv_common.hpp
329
spirv_common.hpp
@ -20,6 +20,7 @@
|
||||
#include "spirv.hpp"
|
||||
#include "spirv_cross_containers.hpp"
|
||||
#include "spirv_cross_error_handling.hpp"
|
||||
#include <functional>
|
||||
|
||||
// A bit crude, but allows projects which embed SPIRV-Cross statically to
|
||||
// effectively hide all the symbols from other projects.
|
||||
@ -270,20 +271,6 @@ struct Instruction
|
||||
uint32_t length = 0;
|
||||
};
|
||||
|
||||
// Helper for Variant interface.
|
||||
struct IVariant
|
||||
{
|
||||
virtual ~IVariant() = default;
|
||||
virtual IVariant *clone(ObjectPoolBase *pool) = 0;
|
||||
uint32_t self = 0;
|
||||
};
|
||||
|
||||
#define SPIRV_CROSS_DECLARE_CLONE(T) \
|
||||
IVariant *clone(ObjectPoolBase *pool) override \
|
||||
{ \
|
||||
return static_cast<ObjectPool<T> *>(pool)->allocate(*this); \
|
||||
}
|
||||
|
||||
enum Types
|
||||
{
|
||||
TypeNone,
|
||||
@ -303,6 +290,136 @@ enum Types
|
||||
TypeCount
|
||||
};
|
||||
|
||||
template <Types type>
|
||||
class TypedID;
|
||||
|
||||
template <>
|
||||
class TypedID<TypeNone>
|
||||
{
|
||||
public:
|
||||
TypedID() = default;
|
||||
TypedID(uint32_t id_)
|
||||
: id(id_)
|
||||
{
|
||||
}
|
||||
|
||||
template <Types U>
|
||||
TypedID(const TypedID<U> &other)
|
||||
{
|
||||
*this = other;
|
||||
}
|
||||
|
||||
template <Types U>
|
||||
TypedID &operator=(const TypedID<U> &other)
|
||||
{
|
||||
id = uint32_t(other);
|
||||
return *this;
|
||||
}
|
||||
|
||||
// Implicit conversion to u32 is desired here.
|
||||
// As long as we block implicit conversion between TypedID<A> and TypedID<B> we're good.
|
||||
operator uint32_t() const
|
||||
{
|
||||
return id;
|
||||
}
|
||||
|
||||
template <Types U>
|
||||
operator TypedID<U>() const
|
||||
{
|
||||
return TypedID<U>(*this);
|
||||
}
|
||||
|
||||
bool operator==(const TypedID &other) const
|
||||
{
|
||||
return id == other.id;
|
||||
}
|
||||
|
||||
bool operator!=(const TypedID &other) const
|
||||
{
|
||||
return id != other.id;
|
||||
}
|
||||
|
||||
template <Types type>
|
||||
bool operator==(const TypedID<type> &other) const
|
||||
{
|
||||
return id == uint32_t(other);
|
||||
}
|
||||
|
||||
template <Types type>
|
||||
bool operator!=(const TypedID<type> &other) const
|
||||
{
|
||||
return id != uint32_t(other);
|
||||
}
|
||||
|
||||
private:
|
||||
uint32_t id = 0;
|
||||
};
|
||||
|
||||
template <Types type>
|
||||
class TypedID
|
||||
{
|
||||
public:
|
||||
TypedID() = default;
|
||||
TypedID(uint32_t id_)
|
||||
: id(id_)
|
||||
{
|
||||
}
|
||||
|
||||
explicit TypedID(const TypedID<TypeNone> &other)
|
||||
: id(uint32_t(other))
|
||||
{
|
||||
}
|
||||
|
||||
operator uint32_t() const
|
||||
{
|
||||
return id;
|
||||
}
|
||||
|
||||
bool operator==(const TypedID &other) const
|
||||
{
|
||||
return id == other.id;
|
||||
}
|
||||
|
||||
bool operator!=(const TypedID &other) const
|
||||
{
|
||||
return id != other.id;
|
||||
}
|
||||
|
||||
bool operator==(const TypedID<TypeNone> &other) const
|
||||
{
|
||||
return id == uint32_t(other);
|
||||
}
|
||||
|
||||
bool operator!=(const TypedID<TypeNone> &other) const
|
||||
{
|
||||
return id != uint32_t(other);
|
||||
}
|
||||
|
||||
private:
|
||||
uint32_t id = 0;
|
||||
};
|
||||
|
||||
using VariableID = TypedID<TypeVariable>;
|
||||
using TypeID = TypedID<TypeType>;
|
||||
using ConstantID = TypedID<TypeConstant>;
|
||||
using FunctionID = TypedID<TypeFunction>;
|
||||
using BlockID = TypedID<TypeBlock>;
|
||||
using ID = TypedID<TypeNone>;
|
||||
|
||||
// Helper for Variant interface.
|
||||
struct IVariant
|
||||
{
|
||||
virtual ~IVariant() = default;
|
||||
virtual IVariant *clone(ObjectPoolBase *pool) = 0;
|
||||
ID self = 0;
|
||||
};
|
||||
|
||||
#define SPIRV_CROSS_DECLARE_CLONE(T) \
|
||||
IVariant *clone(ObjectPoolBase *pool) override \
|
||||
{ \
|
||||
return static_cast<ObjectPool<T> *>(pool)->allocate(*this); \
|
||||
}
|
||||
|
||||
struct SPIRUndef : IVariant
|
||||
{
|
||||
enum
|
||||
@ -310,11 +427,11 @@ struct SPIRUndef : IVariant
|
||||
type = TypeUndef
|
||||
};
|
||||
|
||||
explicit SPIRUndef(uint32_t basetype_)
|
||||
explicit SPIRUndef(TypeID basetype_)
|
||||
: basetype(basetype_)
|
||||
{
|
||||
}
|
||||
uint32_t basetype;
|
||||
TypeID basetype;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRUndef)
|
||||
};
|
||||
@ -344,15 +461,15 @@ struct SPIRCombinedImageSampler : IVariant
|
||||
{
|
||||
type = TypeCombinedImageSampler
|
||||
};
|
||||
SPIRCombinedImageSampler(uint32_t type_, uint32_t image_, uint32_t sampler_)
|
||||
SPIRCombinedImageSampler(TypeID type_, VariableID image_, VariableID sampler_)
|
||||
: combined_type(type_)
|
||||
, image(image_)
|
||||
, sampler(sampler_)
|
||||
{
|
||||
}
|
||||
uint32_t combined_type;
|
||||
uint32_t image;
|
||||
uint32_t sampler;
|
||||
TypeID combined_type;
|
||||
VariableID image;
|
||||
VariableID sampler;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRCombinedImageSampler)
|
||||
};
|
||||
@ -364,16 +481,18 @@ struct SPIRConstantOp : IVariant
|
||||
type = TypeConstantOp
|
||||
};
|
||||
|
||||
SPIRConstantOp(uint32_t result_type, spv::Op op, const uint32_t *args, uint32_t length)
|
||||
SPIRConstantOp(TypeID result_type, spv::Op op, const uint32_t *args, uint32_t length)
|
||||
: opcode(op)
|
||||
, arguments(args, args + length)
|
||||
, basetype(result_type)
|
||||
{
|
||||
arguments.reserve(length);
|
||||
for (uint32_t i = 0; i < length; i++)
|
||||
arguments.push_back(args[i]);
|
||||
}
|
||||
|
||||
spv::Op opcode;
|
||||
SmallVector<uint32_t> arguments;
|
||||
uint32_t basetype;
|
||||
TypeID basetype;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRConstantOp)
|
||||
};
|
||||
@ -436,11 +555,11 @@ struct SPIRType : IVariant
|
||||
|
||||
spv::StorageClass storage = spv::StorageClassGeneric;
|
||||
|
||||
SmallVector<uint32_t> member_types;
|
||||
SmallVector<TypeID> member_types;
|
||||
|
||||
struct ImageType
|
||||
{
|
||||
uint32_t type;
|
||||
TypeID type;
|
||||
spv::Dim dim;
|
||||
bool depth;
|
||||
bool arrayed;
|
||||
@ -453,11 +572,11 @@ struct SPIRType : IVariant
|
||||
// Structs can be declared multiple times if they are used as part of interface blocks.
|
||||
// We want to detect this so that we only emit the struct definition once.
|
||||
// Since we cannot rely on OpName to be equal, we need to figure out aliases.
|
||||
uint32_t type_alias = 0;
|
||||
TypeID type_alias = 0;
|
||||
|
||||
// Denotes the type which this type is based on.
|
||||
// Allows the backend to traverse how a complex type is built up during access chains.
|
||||
uint32_t parent_type = 0;
|
||||
TypeID parent_type = 0;
|
||||
|
||||
// Used in backends to avoid emitting members with conflicting names.
|
||||
std::unordered_set<std::string> member_name_cache;
|
||||
@ -496,7 +615,7 @@ struct SPIRExtension : IVariant
|
||||
// so in order to avoid conflicts, we can't stick them in the ids array.
|
||||
struct SPIREntryPoint
|
||||
{
|
||||
SPIREntryPoint(uint32_t self_, spv::ExecutionModel execution_model, const std::string &entry_name)
|
||||
SPIREntryPoint(FunctionID self_, spv::ExecutionModel execution_model, const std::string &entry_name)
|
||||
: self(self_)
|
||||
, name(entry_name)
|
||||
, orig_name(entry_name)
|
||||
@ -505,10 +624,10 @@ struct SPIREntryPoint
|
||||
}
|
||||
SPIREntryPoint() = default;
|
||||
|
||||
uint32_t self = 0;
|
||||
FunctionID self = 0;
|
||||
std::string name;
|
||||
std::string orig_name;
|
||||
SmallVector<uint32_t> interface_variables;
|
||||
SmallVector<VariableID> interface_variables;
|
||||
|
||||
Bitset flags;
|
||||
struct
|
||||
@ -529,7 +648,7 @@ struct SPIRExpression : IVariant
|
||||
};
|
||||
|
||||
// Only created by the backend target to avoid creating tons of temporaries.
|
||||
SPIRExpression(std::string expr, uint32_t expression_type_, bool immutable_)
|
||||
SPIRExpression(std::string expr, TypeID expression_type_, bool immutable_)
|
||||
: expression(move(expr))
|
||||
, expression_type(expression_type_)
|
||||
, immutable(immutable_)
|
||||
@ -539,14 +658,14 @@ struct SPIRExpression : IVariant
|
||||
// If non-zero, prepend expression with to_expression(base_expression).
|
||||
// Used in amortizing multiple calls to to_expression()
|
||||
// where in certain cases that would quickly force a temporary when not needed.
|
||||
uint32_t base_expression = 0;
|
||||
ID base_expression = 0;
|
||||
|
||||
std::string expression;
|
||||
uint32_t expression_type = 0;
|
||||
TypeID expression_type = 0;
|
||||
|
||||
// If this expression is a forwarded load,
|
||||
// allow us to reference the original variable.
|
||||
uint32_t loaded_from = 0;
|
||||
ID loaded_from = 0;
|
||||
|
||||
// If this expression will never change, we can avoid lots of temporaries
|
||||
// in high level source.
|
||||
@ -562,11 +681,11 @@ struct SPIRExpression : IVariant
|
||||
bool access_chain = false;
|
||||
|
||||
// A list of expressions which this expression depends on.
|
||||
SmallVector<uint32_t> expression_dependencies;
|
||||
SmallVector<ID> expression_dependencies;
|
||||
|
||||
// By reading this expression, we implicitly read these expressions as well.
|
||||
// Used by access chain Store and Load since we read multiple expressions in this case.
|
||||
SmallVector<uint32_t> implied_read_expressions;
|
||||
SmallVector<ID> implied_read_expressions;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRExpression)
|
||||
};
|
||||
@ -578,12 +697,12 @@ struct SPIRFunctionPrototype : IVariant
|
||||
type = TypeFunctionPrototype
|
||||
};
|
||||
|
||||
explicit SPIRFunctionPrototype(uint32_t return_type_)
|
||||
explicit SPIRFunctionPrototype(TypeID return_type_)
|
||||
: return_type(return_type_)
|
||||
{
|
||||
}
|
||||
|
||||
uint32_t return_type;
|
||||
TypeID return_type;
|
||||
SmallVector<uint32_t> parameter_types;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRFunctionPrototype)
|
||||
@ -658,23 +777,23 @@ struct SPIRBlock : IVariant
|
||||
Terminator terminator = Unknown;
|
||||
Merge merge = MergeNone;
|
||||
Hints hint = HintNone;
|
||||
uint32_t next_block = 0;
|
||||
uint32_t merge_block = 0;
|
||||
uint32_t continue_block = 0;
|
||||
BlockID next_block = 0;
|
||||
BlockID merge_block = 0;
|
||||
BlockID continue_block = 0;
|
||||
|
||||
uint32_t return_value = 0; // If 0, return nothing (void).
|
||||
uint32_t condition = 0;
|
||||
uint32_t true_block = 0;
|
||||
uint32_t false_block = 0;
|
||||
uint32_t default_block = 0;
|
||||
ID return_value = 0; // If 0, return nothing (void).
|
||||
ID condition = 0;
|
||||
BlockID true_block = 0;
|
||||
BlockID false_block = 0;
|
||||
BlockID default_block = 0;
|
||||
|
||||
SmallVector<Instruction> ops;
|
||||
|
||||
struct Phi
|
||||
{
|
||||
uint32_t local_variable; // flush local variable ...
|
||||
uint32_t parent; // If we're in from_block and want to branch into this block ...
|
||||
uint32_t function_variable; // to this function-global "phi" variable first.
|
||||
ID local_variable; // flush local variable ...
|
||||
BlockID parent; // If we're in from_block and want to branch into this block ...
|
||||
VariableID function_variable; // to this function-global "phi" variable first.
|
||||
};
|
||||
|
||||
// Before entering this block flush out local variables to magical "phi" variables.
|
||||
@ -682,16 +801,16 @@ struct SPIRBlock : IVariant
|
||||
|
||||
// Declare these temporaries before beginning the block.
|
||||
// Used for handling complex continue blocks which have side effects.
|
||||
SmallVector<std::pair<uint32_t, uint32_t>> declare_temporary;
|
||||
SmallVector<std::pair<TypeID, ID>> declare_temporary;
|
||||
|
||||
// Declare these temporaries, but only conditionally if this block turns out to be
|
||||
// a complex loop header.
|
||||
SmallVector<std::pair<uint32_t, uint32_t>> potential_declare_temporary;
|
||||
SmallVector<std::pair<TypeID, ID>> potential_declare_temporary;
|
||||
|
||||
struct Case
|
||||
{
|
||||
uint32_t value;
|
||||
uint32_t block;
|
||||
BlockID block;
|
||||
};
|
||||
SmallVector<Case> cases;
|
||||
|
||||
@ -707,25 +826,25 @@ struct SPIRBlock : IVariant
|
||||
|
||||
// If marked, we have explicitly handled Phi from this block, so skip any flushes related to that on a branch.
|
||||
// Used to handle an edge case with switch and case-label fallthrough where fall-through writes to Phi.
|
||||
uint32_t ignore_phi_from_block = 0;
|
||||
BlockID ignore_phi_from_block = 0;
|
||||
|
||||
// The dominating block which this block might be within.
|
||||
// Used in continue; blocks to determine if we really need to write continue.
|
||||
uint32_t loop_dominator = 0;
|
||||
BlockID loop_dominator = 0;
|
||||
|
||||
// All access to these variables are dominated by this block,
|
||||
// so before branching anywhere we need to make sure that we declare these variables.
|
||||
SmallVector<uint32_t> dominated_variables;
|
||||
SmallVector<VariableID> dominated_variables;
|
||||
|
||||
// These are variables which should be declared in a for loop header, if we
|
||||
// fail to use a classic for-loop,
|
||||
// we remove these variables, and fall back to regular variables outside the loop.
|
||||
SmallVector<uint32_t> loop_variables;
|
||||
SmallVector<VariableID> loop_variables;
|
||||
|
||||
// Some expressions are control-flow dependent, i.e. any instruction which relies on derivatives or
|
||||
// sub-group-like operations.
|
||||
// Make sure that we only use these expressions in the original block.
|
||||
SmallVector<uint32_t> invalidate_expressions;
|
||||
SmallVector<ID> invalidate_expressions;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRBlock)
|
||||
};
|
||||
@ -737,7 +856,7 @@ struct SPIRFunction : IVariant
|
||||
type = TypeFunction
|
||||
};
|
||||
|
||||
SPIRFunction(uint32_t return_type_, uint32_t function_type_)
|
||||
SPIRFunction(TypeID return_type_, TypeID function_type_)
|
||||
: return_type(return_type_)
|
||||
, function_type(function_type_)
|
||||
{
|
||||
@ -745,8 +864,8 @@ struct SPIRFunction : IVariant
|
||||
|
||||
struct Parameter
|
||||
{
|
||||
uint32_t type;
|
||||
uint32_t id;
|
||||
TypeID type;
|
||||
ID id;
|
||||
uint32_t read_count;
|
||||
uint32_t write_count;
|
||||
|
||||
@ -768,25 +887,25 @@ struct SPIRFunction : IVariant
|
||||
// or a global ID.
|
||||
struct CombinedImageSamplerParameter
|
||||
{
|
||||
uint32_t id;
|
||||
uint32_t image_id;
|
||||
uint32_t sampler_id;
|
||||
VariableID id;
|
||||
VariableID image_id;
|
||||
VariableID sampler_id;
|
||||
bool global_image;
|
||||
bool global_sampler;
|
||||
bool depth;
|
||||
};
|
||||
|
||||
uint32_t return_type;
|
||||
uint32_t function_type;
|
||||
TypeID return_type;
|
||||
TypeID function_type;
|
||||
SmallVector<Parameter> arguments;
|
||||
|
||||
// Can be used by backends to add magic arguments.
|
||||
// Currently used by combined image/sampler implementation.
|
||||
|
||||
SmallVector<Parameter> shadow_arguments;
|
||||
SmallVector<uint32_t> local_variables;
|
||||
uint32_t entry_block = 0;
|
||||
SmallVector<uint32_t> blocks;
|
||||
SmallVector<VariableID> local_variables;
|
||||
BlockID entry_block = 0;
|
||||
SmallVector<BlockID> blocks;
|
||||
SmallVector<CombinedImageSamplerParameter> combined_parameters;
|
||||
|
||||
struct EntryLine
|
||||
@ -796,12 +915,12 @@ struct SPIRFunction : IVariant
|
||||
};
|
||||
EntryLine entry_line;
|
||||
|
||||
void add_local_variable(uint32_t id)
|
||||
void add_local_variable(VariableID id)
|
||||
{
|
||||
local_variables.push_back(id);
|
||||
}
|
||||
|
||||
void add_parameter(uint32_t parameter_type, uint32_t id, bool alias_global_variable = false)
|
||||
void add_parameter(TypeID parameter_type, ID id, bool alias_global_variable = false)
|
||||
{
|
||||
// Arguments are read-only until proven otherwise.
|
||||
arguments.push_back({ parameter_type, id, 0u, 0u, alias_global_variable });
|
||||
@ -822,7 +941,7 @@ struct SPIRFunction : IVariant
|
||||
// On function entry, make sure to copy a constant array into thread addr space to work around
|
||||
// the case where we are passing a constant array by value to a function on backends which do not
|
||||
// consider arrays value types.
|
||||
SmallVector<uint32_t> constant_arrays_needed_on_stack;
|
||||
SmallVector<ID> constant_arrays_needed_on_stack;
|
||||
|
||||
bool active = false;
|
||||
bool flush_undeclared = true;
|
||||
@ -838,7 +957,7 @@ struct SPIRAccessChain : IVariant
|
||||
type = TypeAccessChain
|
||||
};
|
||||
|
||||
SPIRAccessChain(uint32_t basetype_, spv::StorageClass storage_, std::string base_, std::string dynamic_index_,
|
||||
SPIRAccessChain(TypeID basetype_, spv::StorageClass storage_, std::string base_, std::string dynamic_index_,
|
||||
int32_t static_index_)
|
||||
: basetype(basetype_)
|
||||
, storage(storage_)
|
||||
@ -853,20 +972,20 @@ struct SPIRAccessChain : IVariant
|
||||
// which has no usable buffer type ala GLSL SSBOs.
|
||||
// StructuredBuffer is too limited, so our only option is to deal with ByteAddressBuffer which works with raw addresses.
|
||||
|
||||
uint32_t basetype;
|
||||
TypeID basetype;
|
||||
spv::StorageClass storage;
|
||||
std::string base;
|
||||
std::string dynamic_index;
|
||||
int32_t static_index;
|
||||
|
||||
uint32_t loaded_from = 0;
|
||||
VariableID loaded_from = 0;
|
||||
uint32_t matrix_stride = 0;
|
||||
bool row_major_matrix = false;
|
||||
bool immutable = false;
|
||||
|
||||
// By reading this expression, we implicitly read these expressions as well.
|
||||
// Used by access chain Store and Load since we read multiple expressions in this case.
|
||||
SmallVector<uint32_t> implied_read_expressions;
|
||||
SmallVector<ID> implied_read_expressions;
|
||||
|
||||
SPIRV_CROSS_DECLARE_CLONE(SPIRAccessChain)
|
||||
};
|
||||
@ -879,7 +998,7 @@ struct SPIRVariable : IVariant
|
||||
};
|
||||
|
||||
SPIRVariable() = default;
|
||||
SPIRVariable(uint32_t basetype_, spv::StorageClass storage_, uint32_t initializer_ = 0, uint32_t basevariable_ = 0)
|
||||
SPIRVariable(TypeID basetype_, spv::StorageClass storage_, ID initializer_ = 0, VariableID basevariable_ = 0)
|
||||
: basetype(basetype_)
|
||||
, storage(storage_)
|
||||
, initializer(initializer_)
|
||||
@ -887,11 +1006,11 @@ struct SPIRVariable : IVariant
|
||||
{
|
||||
}
|
||||
|
||||
uint32_t basetype = 0;
|
||||
TypeID basetype = 0;
|
||||
spv::StorageClass storage = spv::StorageClassGeneric;
|
||||
uint32_t decoration = 0;
|
||||
uint32_t initializer = 0;
|
||||
uint32_t basevariable = 0;
|
||||
ID initializer = 0;
|
||||
VariableID basevariable = 0;
|
||||
|
||||
SmallVector<uint32_t> dereference_chain;
|
||||
bool compat_builtin = false;
|
||||
@ -901,10 +1020,10 @@ struct SPIRVariable : IVariant
|
||||
// When we read the variable as an expression, just forward
|
||||
// shadowed_id as the expression.
|
||||
bool statically_assigned = false;
|
||||
uint32_t static_expression = 0;
|
||||
ID static_expression = 0;
|
||||
|
||||
// Temporaries which can remain forwarded as long as this variable is not modified.
|
||||
SmallVector<uint32_t> dependees;
|
||||
SmallVector<ID> dependees;
|
||||
bool forwardable = true;
|
||||
|
||||
bool deferred_declaration = false;
|
||||
@ -917,7 +1036,7 @@ struct SPIRVariable : IVariant
|
||||
uint32_t remapped_components = 0;
|
||||
|
||||
// The block which dominates all access to this variable.
|
||||
uint32_t dominator = 0;
|
||||
BlockID dominator = 0;
|
||||
// If true, this variable is a loop variable, when accessing the variable
|
||||
// outside a loop,
|
||||
// we should statically forward it.
|
||||
@ -952,15 +1071,12 @@ struct SPIRConstant : IVariant
|
||||
{
|
||||
Constant r[4];
|
||||
// If != 0, this element is a specialization constant, and we should keep track of it as such.
|
||||
uint32_t id[4];
|
||||
ID id[4];
|
||||
uint32_t vecsize = 1;
|
||||
|
||||
// Workaround for MSVC 2013, initializing an array breaks.
|
||||
ConstantVector()
|
||||
{
|
||||
memset(r, 0, sizeof(r));
|
||||
for (unsigned i = 0; i < 4; i++)
|
||||
id[i] = 0;
|
||||
}
|
||||
};
|
||||
|
||||
@ -968,15 +1084,8 @@ struct SPIRConstant : IVariant
|
||||
{
|
||||
ConstantVector c[4];
|
||||
// If != 0, this column is a specialization constant, and we should keep track of it as such.
|
||||
uint32_t id[4];
|
||||
ID id[4];
|
||||
uint32_t columns = 1;
|
||||
|
||||
// Workaround for MSVC 2013, initializing an array breaks.
|
||||
ConstantMatrix()
|
||||
{
|
||||
for (unsigned i = 0; i < 4; i++)
|
||||
id[i] = 0;
|
||||
}
|
||||
};
|
||||
|
||||
static inline float f16_to_f32(uint16_t u16_value)
|
||||
@ -1141,16 +1250,18 @@ struct SPIRConstant : IVariant
|
||||
|
||||
SPIRConstant() = default;
|
||||
|
||||
SPIRConstant(uint32_t constant_type_, const uint32_t *elements, uint32_t num_elements, bool specialized)
|
||||
SPIRConstant(TypeID constant_type_, const uint32_t *elements, uint32_t num_elements, bool specialized)
|
||||
: constant_type(constant_type_)
|
||||
, specialization(specialized)
|
||||
{
|
||||
subconstants.insert(std::end(subconstants), elements, elements + num_elements);
|
||||
subconstants.reserve(num_elements);
|
||||
for (uint32_t i = 0; i < num_elements; i++)
|
||||
subconstants.push_back(elements[i]);
|
||||
specialization = specialized;
|
||||
}
|
||||
|
||||
// Construct scalar (32-bit).
|
||||
SPIRConstant(uint32_t constant_type_, uint32_t v0, bool specialized)
|
||||
SPIRConstant(TypeID constant_type_, uint32_t v0, bool specialized)
|
||||
: constant_type(constant_type_)
|
||||
, specialization(specialized)
|
||||
{
|
||||
@ -1160,7 +1271,7 @@ struct SPIRConstant : IVariant
|
||||
}
|
||||
|
||||
// Construct scalar (64-bit).
|
||||
SPIRConstant(uint32_t constant_type_, uint64_t v0, bool specialized)
|
||||
SPIRConstant(TypeID constant_type_, uint64_t v0, bool specialized)
|
||||
: constant_type(constant_type_)
|
||||
, specialization(specialized)
|
||||
{
|
||||
@ -1170,7 +1281,7 @@ struct SPIRConstant : IVariant
|
||||
}
|
||||
|
||||
// Construct vectors and matrices.
|
||||
SPIRConstant(uint32_t constant_type_, const SPIRConstant *const *vector_elements, uint32_t num_elements,
|
||||
SPIRConstant(TypeID constant_type_, const SPIRConstant *const *vector_elements, uint32_t num_elements,
|
||||
bool specialized)
|
||||
: constant_type(constant_type_)
|
||||
, specialization(specialized)
|
||||
@ -1202,7 +1313,7 @@ struct SPIRConstant : IVariant
|
||||
}
|
||||
}
|
||||
|
||||
uint32_t constant_type = 0;
|
||||
TypeID constant_type = 0;
|
||||
ConstantMatrix m;
|
||||
|
||||
// If this constant is a specialization constant (i.e. created with OpSpecConstant*).
|
||||
@ -1214,7 +1325,7 @@ struct SPIRConstant : IVariant
|
||||
bool is_used_as_lut = false;
|
||||
|
||||
// For composites which are constant arrays, etc.
|
||||
SmallVector<uint32_t> subconstants;
|
||||
SmallVector<ConstantID> subconstants;
|
||||
|
||||
// Non-Vulkan GLSL, HLSL and sometimes MSL emits defines for each specialization constant,
|
||||
// and uses them to initialize the constant. This allows the user
|
||||
@ -1349,9 +1460,9 @@ public:
|
||||
return type;
|
||||
}
|
||||
|
||||
uint32_t get_id() const
|
||||
ID get_id() const
|
||||
{
|
||||
return holder ? holder->self : 0;
|
||||
return holder ? holder->self : ID(0);
|
||||
}
|
||||
|
||||
bool empty() const
|
||||
@ -1591,4 +1702,16 @@ static inline bool opcode_is_sign_invariant(spv::Op opcode)
|
||||
}
|
||||
} // namespace SPIRV_CROSS_NAMESPACE
|
||||
|
||||
namespace std
|
||||
{
|
||||
template <SPIRV_CROSS_NAMESPACE::Types type>
|
||||
struct hash<SPIRV_CROSS_NAMESPACE::TypedID<type>>
|
||||
{
|
||||
size_t operator()(const SPIRV_CROSS_NAMESPACE::TypedID<type> &value) const
|
||||
{
|
||||
return std::hash<uint32_t>()(value);
|
||||
}
|
||||
};
|
||||
} // namespace std
|
||||
|
||||
#endif
|
||||
|
144
spirv_cross.cpp
144
spirv_cross.cpp
@ -578,7 +578,7 @@ ShaderResources Compiler::get_shader_resources() const
|
||||
return get_shader_resources(nullptr);
|
||||
}
|
||||
|
||||
ShaderResources Compiler::get_shader_resources(const unordered_set<uint32_t> &active_variables) const
|
||||
ShaderResources Compiler::get_shader_resources(const unordered_set<VariableID> &active_variables) const
|
||||
{
|
||||
return get_shader_resources(&active_variables);
|
||||
}
|
||||
@ -735,16 +735,16 @@ bool Compiler::InterfaceVariableAccessHandler::handle(Op opcode, const uint32_t
|
||||
return true;
|
||||
}
|
||||
|
||||
unordered_set<uint32_t> Compiler::get_active_interface_variables() const
|
||||
unordered_set<VariableID> Compiler::get_active_interface_variables() const
|
||||
{
|
||||
// Traverse the call graph and find all interface variables which are in use.
|
||||
unordered_set<uint32_t> variables;
|
||||
unordered_set<VariableID> variables;
|
||||
InterfaceVariableAccessHandler handler(*this, variables);
|
||||
traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
|
||||
|
||||
// Make sure we preserve output variables which are only initialized, but never accessed by any code.
|
||||
ir.for_each_typed_id<SPIRVariable>([&](uint32_t, const SPIRVariable &var) {
|
||||
if (var.storage == StorageClassOutput && var.initializer != 0)
|
||||
if (var.storage == StorageClassOutput && var.initializer != ID(0))
|
||||
variables.insert(var.self);
|
||||
});
|
||||
|
||||
@ -755,13 +755,13 @@ unordered_set<uint32_t> Compiler::get_active_interface_variables() const
|
||||
return variables;
|
||||
}
|
||||
|
||||
void Compiler::set_enabled_interface_variables(std::unordered_set<uint32_t> active_variables)
|
||||
void Compiler::set_enabled_interface_variables(std::unordered_set<VariableID> active_variables)
|
||||
{
|
||||
active_interface_variables = move(active_variables);
|
||||
check_active_interface_variables = true;
|
||||
}
|
||||
|
||||
ShaderResources Compiler::get_shader_resources(const unordered_set<uint32_t> *active_variables) const
|
||||
ShaderResources Compiler::get_shader_resources(const unordered_set<VariableID> *active_variables) const
|
||||
{
|
||||
ShaderResources res;
|
||||
|
||||
@ -978,17 +978,17 @@ void Compiler::update_name_cache(unordered_set<string> &cache, string &name)
|
||||
update_name_cache(cache, cache, name);
|
||||
}
|
||||
|
||||
void Compiler::set_name(uint32_t id, const std::string &name)
|
||||
void Compiler::set_name(ID id, const std::string &name)
|
||||
{
|
||||
ir.set_name(id, name);
|
||||
}
|
||||
|
||||
const SPIRType &Compiler::get_type(uint32_t id) const
|
||||
const SPIRType &Compiler::get_type(TypeID id) const
|
||||
{
|
||||
return get<SPIRType>(id);
|
||||
}
|
||||
|
||||
const SPIRType &Compiler::get_type_from_variable(uint32_t id) const
|
||||
const SPIRType &Compiler::get_type_from_variable(VariableID id) const
|
||||
{
|
||||
return get<SPIRType>(get<SPIRVariable>(id).basetype);
|
||||
}
|
||||
@ -1059,23 +1059,23 @@ bool Compiler::is_sampled_image_type(const SPIRType &type)
|
||||
type.image.dim != DimBuffer;
|
||||
}
|
||||
|
||||
void Compiler::set_member_decoration_string(uint32_t id, uint32_t index, spv::Decoration decoration,
|
||||
void Compiler::set_member_decoration_string(TypeID id, uint32_t index, spv::Decoration decoration,
|
||||
const std::string &argument)
|
||||
{
|
||||
ir.set_member_decoration_string(id, index, decoration, argument);
|
||||
}
|
||||
|
||||
void Compiler::set_member_decoration(uint32_t id, uint32_t index, Decoration decoration, uint32_t argument)
|
||||
void Compiler::set_member_decoration(TypeID id, uint32_t index, Decoration decoration, uint32_t argument)
|
||||
{
|
||||
ir.set_member_decoration(id, index, decoration, argument);
|
||||
}
|
||||
|
||||
void Compiler::set_member_name(uint32_t id, uint32_t index, const std::string &name)
|
||||
void Compiler::set_member_name(TypeID id, uint32_t index, const std::string &name)
|
||||
{
|
||||
ir.set_member_name(id, index, name);
|
||||
}
|
||||
|
||||
const std::string &Compiler::get_member_name(uint32_t id, uint32_t index) const
|
||||
const std::string &Compiler::get_member_name(TypeID id, uint32_t index) const
|
||||
{
|
||||
return ir.get_member_name(id, index);
|
||||
}
|
||||
@ -1091,7 +1091,7 @@ void Compiler::set_member_qualified_name(uint32_t type_id, uint32_t index, const
|
||||
ir.meta[type_id].members[index].qualified_alias = name;
|
||||
}
|
||||
|
||||
const string &Compiler::get_member_qualified_name(uint32_t type_id, uint32_t index) const
|
||||
const string &Compiler::get_member_qualified_name(TypeID type_id, uint32_t index) const
|
||||
{
|
||||
auto *m = ir.find_meta(type_id);
|
||||
if (m && index < m->members.size())
|
||||
@ -1100,32 +1100,32 @@ const string &Compiler::get_member_qualified_name(uint32_t type_id, uint32_t ind
|
||||
return ir.get_empty_string();
|
||||
}
|
||||
|
||||
uint32_t Compiler::get_member_decoration(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
uint32_t Compiler::get_member_decoration(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
return ir.get_member_decoration(id, index, decoration);
|
||||
}
|
||||
|
||||
const Bitset &Compiler::get_member_decoration_bitset(uint32_t id, uint32_t index) const
|
||||
const Bitset &Compiler::get_member_decoration_bitset(TypeID id, uint32_t index) const
|
||||
{
|
||||
return ir.get_member_decoration_bitset(id, index);
|
||||
}
|
||||
|
||||
bool Compiler::has_member_decoration(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
bool Compiler::has_member_decoration(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
return ir.has_member_decoration(id, index, decoration);
|
||||
}
|
||||
|
||||
void Compiler::unset_member_decoration(uint32_t id, uint32_t index, Decoration decoration)
|
||||
void Compiler::unset_member_decoration(TypeID id, uint32_t index, Decoration decoration)
|
||||
{
|
||||
ir.unset_member_decoration(id, index, decoration);
|
||||
}
|
||||
|
||||
void Compiler::set_decoration_string(uint32_t id, spv::Decoration decoration, const std::string &argument)
|
||||
void Compiler::set_decoration_string(ID id, spv::Decoration decoration, const std::string &argument)
|
||||
{
|
||||
ir.set_decoration_string(id, decoration, argument);
|
||||
}
|
||||
|
||||
void Compiler::set_decoration(uint32_t id, Decoration decoration, uint32_t argument)
|
||||
void Compiler::set_decoration(ID id, Decoration decoration, uint32_t argument)
|
||||
{
|
||||
ir.set_decoration(id, decoration, argument);
|
||||
}
|
||||
@ -1229,22 +1229,22 @@ void Compiler::unset_extended_member_decoration(uint32_t type, uint32_t index, E
|
||||
dec.extended.values[decoration] = 0;
|
||||
}
|
||||
|
||||
StorageClass Compiler::get_storage_class(uint32_t id) const
|
||||
StorageClass Compiler::get_storage_class(VariableID id) const
|
||||
{
|
||||
return get<SPIRVariable>(id).storage;
|
||||
}
|
||||
|
||||
const std::string &Compiler::get_name(uint32_t id) const
|
||||
const std::string &Compiler::get_name(ID id) const
|
||||
{
|
||||
return ir.get_name(id);
|
||||
}
|
||||
|
||||
const std::string Compiler::get_fallback_name(uint32_t id) const
|
||||
const std::string Compiler::get_fallback_name(ID id) const
|
||||
{
|
||||
return join("_", id);
|
||||
}
|
||||
|
||||
const std::string Compiler::get_block_fallback_name(uint32_t id) const
|
||||
const std::string Compiler::get_block_fallback_name(VariableID id) const
|
||||
{
|
||||
auto &var = get<SPIRVariable>(id);
|
||||
if (get_name(id).empty())
|
||||
@ -1253,37 +1253,37 @@ const std::string Compiler::get_block_fallback_name(uint32_t id) const
|
||||
return get_name(id);
|
||||
}
|
||||
|
||||
const Bitset &Compiler::get_decoration_bitset(uint32_t id) const
|
||||
const Bitset &Compiler::get_decoration_bitset(ID id) const
|
||||
{
|
||||
return ir.get_decoration_bitset(id);
|
||||
}
|
||||
|
||||
bool Compiler::has_decoration(uint32_t id, Decoration decoration) const
|
||||
bool Compiler::has_decoration(ID id, Decoration decoration) const
|
||||
{
|
||||
return ir.has_decoration(id, decoration);
|
||||
}
|
||||
|
||||
const string &Compiler::get_decoration_string(uint32_t id, Decoration decoration) const
|
||||
const string &Compiler::get_decoration_string(ID id, Decoration decoration) const
|
||||
{
|
||||
return ir.get_decoration_string(id, decoration);
|
||||
}
|
||||
|
||||
const string &Compiler::get_member_decoration_string(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
const string &Compiler::get_member_decoration_string(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
return ir.get_member_decoration_string(id, index, decoration);
|
||||
}
|
||||
|
||||
uint32_t Compiler::get_decoration(uint32_t id, Decoration decoration) const
|
||||
uint32_t Compiler::get_decoration(ID id, Decoration decoration) const
|
||||
{
|
||||
return ir.get_decoration(id, decoration);
|
||||
}
|
||||
|
||||
void Compiler::unset_decoration(uint32_t id, Decoration decoration)
|
||||
void Compiler::unset_decoration(ID id, Decoration decoration)
|
||||
{
|
||||
ir.unset_decoration(id, decoration);
|
||||
}
|
||||
|
||||
bool Compiler::get_binary_offset_for_decoration(uint32_t id, spv::Decoration decoration, uint32_t &word_offset) const
|
||||
bool Compiler::get_binary_offset_for_decoration(VariableID id, spv::Decoration decoration, uint32_t &word_offset) const
|
||||
{
|
||||
auto *m = ir.find_meta(id);
|
||||
if (!m)
|
||||
@ -1462,7 +1462,7 @@ SPIRBlock::ContinueBlockType Compiler::continue_block_type(const SPIRBlock &bloc
|
||||
if (block.merge == SPIRBlock::MergeLoop)
|
||||
return SPIRBlock::WhileLoop;
|
||||
|
||||
if (block.loop_dominator == SPIRBlock::NoDominator)
|
||||
if (block.loop_dominator == BlockID(SPIRBlock::NoDominator))
|
||||
{
|
||||
// Continue block is never reached from CFG.
|
||||
return SPIRBlock::ComplexLoop;
|
||||
@ -1729,7 +1729,7 @@ bool Compiler::BufferAccessHandler::handle(Op opcode, const uint32_t *args, uint
|
||||
return true;
|
||||
}
|
||||
|
||||
SmallVector<BufferRange> Compiler::get_active_buffer_ranges(uint32_t id) const
|
||||
SmallVector<BufferRange> Compiler::get_active_buffer_ranges(VariableID id) const
|
||||
{
|
||||
SmallVector<BufferRange> ranges;
|
||||
BufferAccessHandler handler(*this, ranges, id);
|
||||
@ -1822,19 +1822,19 @@ uint32_t Compiler::get_work_group_size_specialization_constants(SpecializationCo
|
||||
{
|
||||
auto &c = get<SPIRConstant>(execution.workgroup_size.constant);
|
||||
|
||||
if (c.m.c[0].id[0] != 0)
|
||||
if (c.m.c[0].id[0] != ID(0))
|
||||
{
|
||||
x.id = c.m.c[0].id[0];
|
||||
x.constant_id = get_decoration(c.m.c[0].id[0], DecorationSpecId);
|
||||
}
|
||||
|
||||
if (c.m.c[0].id[1] != 0)
|
||||
if (c.m.c[0].id[1] != ID(0))
|
||||
{
|
||||
y.id = c.m.c[0].id[1];
|
||||
y.constant_id = get_decoration(c.m.c[0].id[1], DecorationSpecId);
|
||||
}
|
||||
|
||||
if (c.m.c[0].id[2] != 0)
|
||||
if (c.m.c[0].id[2] != ID(0))
|
||||
{
|
||||
z.id = c.m.c[0].id[2];
|
||||
z.constant_id = get_decoration(c.m.c[0].id[2], DecorationSpecId);
|
||||
@ -1889,36 +1889,36 @@ bool Compiler::is_tessellation_shader() const
|
||||
return is_tessellation_shader(get_execution_model());
|
||||
}
|
||||
|
||||
void Compiler::set_remapped_variable_state(uint32_t id, bool remap_enable)
|
||||
void Compiler::set_remapped_variable_state(VariableID id, bool remap_enable)
|
||||
{
|
||||
get<SPIRVariable>(id).remapped_variable = remap_enable;
|
||||
}
|
||||
|
||||
bool Compiler::get_remapped_variable_state(uint32_t id) const
|
||||
bool Compiler::get_remapped_variable_state(VariableID id) const
|
||||
{
|
||||
return get<SPIRVariable>(id).remapped_variable;
|
||||
}
|
||||
|
||||
void Compiler::set_subpass_input_remapped_components(uint32_t id, uint32_t components)
|
||||
void Compiler::set_subpass_input_remapped_components(VariableID id, uint32_t components)
|
||||
{
|
||||
get<SPIRVariable>(id).remapped_components = components;
|
||||
}
|
||||
|
||||
uint32_t Compiler::get_subpass_input_remapped_components(uint32_t id) const
|
||||
uint32_t Compiler::get_subpass_input_remapped_components(VariableID id) const
|
||||
{
|
||||
return get<SPIRVariable>(id).remapped_components;
|
||||
}
|
||||
|
||||
void Compiler::add_implied_read_expression(SPIRExpression &e, uint32_t source)
|
||||
{
|
||||
auto itr = find(begin(e.implied_read_expressions), end(e.implied_read_expressions), source);
|
||||
auto itr = find(begin(e.implied_read_expressions), end(e.implied_read_expressions), ID(source));
|
||||
if (itr == end(e.implied_read_expressions))
|
||||
e.implied_read_expressions.push_back(source);
|
||||
}
|
||||
|
||||
void Compiler::add_implied_read_expression(SPIRAccessChain &e, uint32_t source)
|
||||
{
|
||||
auto itr = find(begin(e.implied_read_expressions), end(e.implied_read_expressions), source);
|
||||
auto itr = find(begin(e.implied_read_expressions), end(e.implied_read_expressions), ID(source));
|
||||
if (itr == end(e.implied_read_expressions))
|
||||
e.implied_read_expressions.push_back(source);
|
||||
}
|
||||
@ -2059,7 +2059,7 @@ bool Compiler::interface_variable_exists_in_entry_point(uint32_t id) const
|
||||
return true;
|
||||
|
||||
auto &execution = get_entry_point();
|
||||
return find(begin(execution.interface_variables), end(execution.interface_variables), id) !=
|
||||
return find(begin(execution.interface_variables), end(execution.interface_variables), VariableID(id)) !=
|
||||
end(execution.interface_variables);
|
||||
}
|
||||
|
||||
@ -2139,8 +2139,8 @@ bool Compiler::CombinedImageSamplerHandler::end_function_scope(const uint32_t *a
|
||||
{
|
||||
for (auto ¶m : params)
|
||||
{
|
||||
uint32_t image_id = param.global_image ? param.image_id : args[param.image_id];
|
||||
uint32_t sampler_id = param.global_sampler ? param.sampler_id : args[param.sampler_id];
|
||||
VariableID image_id = param.global_image ? param.image_id : VariableID(args[param.image_id]);
|
||||
VariableID sampler_id = param.global_sampler ? param.sampler_id : VariableID(args[param.sampler_id]);
|
||||
|
||||
auto *i = compiler.maybe_get_backing_variable(image_id);
|
||||
auto *s = compiler.maybe_get_backing_variable(sampler_id);
|
||||
@ -2157,8 +2157,8 @@ bool Compiler::CombinedImageSamplerHandler::end_function_scope(const uint32_t *a
|
||||
}
|
||||
|
||||
void Compiler::CombinedImageSamplerHandler::register_combined_image_sampler(SPIRFunction &caller,
|
||||
uint32_t combined_module_id,
|
||||
uint32_t image_id, uint32_t sampler_id,
|
||||
VariableID combined_module_id,
|
||||
VariableID image_id, VariableID sampler_id,
|
||||
bool depth)
|
||||
{
|
||||
// We now have a texture ID and a sampler ID which will either be found as a global
|
||||
@ -2445,8 +2445,8 @@ bool Compiler::CombinedImageSamplerHandler::handle(Op opcode, const uint32_t *ar
|
||||
// This information is statically known from the current place in the call stack.
|
||||
// Function parameters are not necessarily pointers, so if we don't have a backing variable, remapping will know
|
||||
// which backing variable the image/sample came from.
|
||||
uint32_t image_id = remap_parameter(args[2]);
|
||||
uint32_t sampler_id = is_fetch ? compiler.dummy_sampler_id : remap_parameter(args[3]);
|
||||
VariableID image_id = remap_parameter(args[2]);
|
||||
VariableID sampler_id = is_fetch ? compiler.dummy_sampler_id : remap_parameter(args[3]);
|
||||
|
||||
auto itr = find_if(begin(compiler.combined_image_samplers), end(compiler.combined_image_samplers),
|
||||
[image_id, sampler_id](const CombinedImageSampler &combined) {
|
||||
@ -2515,7 +2515,7 @@ bool Compiler::CombinedImageSamplerHandler::handle(Op opcode, const uint32_t *ar
|
||||
return true;
|
||||
}
|
||||
|
||||
uint32_t Compiler::build_dummy_sampler_for_combined_images()
|
||||
VariableID Compiler::build_dummy_sampler_for_combined_images()
|
||||
{
|
||||
DummySamplerForCombinedImageHandler handler(*this);
|
||||
traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
|
||||
@ -2569,12 +2569,12 @@ SmallVector<SpecializationConstant> Compiler::get_specialization_constants() con
|
||||
return spec_consts;
|
||||
}
|
||||
|
||||
SPIRConstant &Compiler::get_constant(uint32_t id)
|
||||
SPIRConstant &Compiler::get_constant(ConstantID id)
|
||||
{
|
||||
return get<SPIRConstant>(id);
|
||||
}
|
||||
|
||||
const SPIRConstant &Compiler::get_constant(uint32_t id) const
|
||||
const SPIRConstant &Compiler::get_constant(ConstantID id) const
|
||||
{
|
||||
return get<SPIRConstant>(id);
|
||||
}
|
||||
@ -2758,7 +2758,7 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
|
||||
if (length < 2)
|
||||
return false;
|
||||
|
||||
uint32_t ptr = args[0];
|
||||
ID ptr = args[0];
|
||||
auto *var = compiler.maybe_get_backing_variable(ptr);
|
||||
|
||||
// If we store through an access chain, we have a partial write.
|
||||
@ -2803,7 +2803,7 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
|
||||
// The result of an access chain is a fixed expression and is not really considered a temporary.
|
||||
auto &e = compiler.set<SPIRExpression>(args[1], "", args[0], true);
|
||||
auto *backing_variable = compiler.maybe_get_backing_variable(ptr);
|
||||
e.loaded_from = backing_variable ? backing_variable->self : 0;
|
||||
e.loaded_from = backing_variable ? VariableID(backing_variable->self) : VariableID(0);
|
||||
|
||||
// Other backends might use SPIRAccessChain for this later.
|
||||
compiler.ir.ids[args[1]].set_allow_type_rewrite();
|
||||
@ -2816,8 +2816,8 @@ bool Compiler::AnalyzeVariableScopeAccessHandler::handle(spv::Op op, const uint3
|
||||
if (length < 2)
|
||||
return false;
|
||||
|
||||
uint32_t lhs = args[0];
|
||||
uint32_t rhs = args[1];
|
||||
ID lhs = args[0];
|
||||
ID rhs = args[1];
|
||||
auto *var = compiler.maybe_get_backing_variable(lhs);
|
||||
|
||||
// If we store through an access chain, we have a partial write.
|
||||
@ -3182,7 +3182,8 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
|
||||
for (auto &var : handler.accessed_variables_to_block)
|
||||
{
|
||||
// Only deal with variables which are considered local variables in this function.
|
||||
if (find(begin(entry.local_variables), end(entry.local_variables), var.first) == end(entry.local_variables))
|
||||
if (find(begin(entry.local_variables), end(entry.local_variables), VariableID(var.first)) ==
|
||||
end(entry.local_variables))
|
||||
continue;
|
||||
|
||||
DominatorBuilder builder(cfg);
|
||||
@ -3223,7 +3224,7 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
|
||||
builder.lift_continue_block_dominator();
|
||||
|
||||
// Add it to a per-block list of variables.
|
||||
uint32_t dominating_block = builder.get_dominator();
|
||||
BlockID dominating_block = builder.get_dominator();
|
||||
|
||||
// For variables whose dominating block is inside a loop, there is a risk that these variables
|
||||
// actually need to be preserved across loop iterations. We can express this by adding
|
||||
@ -3241,7 +3242,7 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
|
||||
if (preserve)
|
||||
{
|
||||
// Find the outermost loop scope.
|
||||
while (block->loop_dominator != SPIRBlock::NoDominator)
|
||||
while (block->loop_dominator != BlockID(SPIRBlock::NoDominator))
|
||||
block = &get<SPIRBlock>(block->loop_dominator);
|
||||
|
||||
if (block->self != dominating_block)
|
||||
@ -3361,17 +3362,17 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry, AnalyzeVariableScopeA
|
||||
{
|
||||
auto &var = get<SPIRVariable>(loop_variable.first);
|
||||
auto dominator = var.dominator;
|
||||
auto block = loop_variable.second;
|
||||
BlockID block = loop_variable.second;
|
||||
|
||||
// The variable was accessed in multiple continue blocks, ignore.
|
||||
if (block == ~(0u) || block == 0)
|
||||
if (block == BlockID(~(0u)) || block == BlockID(0))
|
||||
continue;
|
||||
|
||||
// Dead code.
|
||||
if (dominator == 0)
|
||||
if (dominator == ID(0))
|
||||
continue;
|
||||
|
||||
uint32_t header = 0;
|
||||
BlockID header = 0;
|
||||
|
||||
// Find the loop header for this block if we are a continue block.
|
||||
{
|
||||
@ -3522,7 +3523,7 @@ bool Compiler::may_read_undefined_variable_in_block(const SPIRBlock &block, uint
|
||||
return true;
|
||||
}
|
||||
|
||||
Bitset Compiler::get_buffer_block_flags(uint32_t id) const
|
||||
Bitset Compiler::get_buffer_block_flags(VariableID id) const
|
||||
{
|
||||
return ir.get_buffer_block_flags(get<SPIRVariable>(id));
|
||||
}
|
||||
@ -3961,13 +3962,13 @@ bool Compiler::CombinedImageSamplerUsageHandler::handle(Op opcode, const uint32_
|
||||
return true;
|
||||
}
|
||||
|
||||
bool Compiler::buffer_is_hlsl_counter_buffer(uint32_t id) const
|
||||
bool Compiler::buffer_is_hlsl_counter_buffer(VariableID id) const
|
||||
{
|
||||
auto *m = ir.find_meta(id);
|
||||
return m && m->hlsl_is_magic_counter_buffer;
|
||||
}
|
||||
|
||||
bool Compiler::buffer_get_hlsl_counter_buffer(uint32_t id, uint32_t &counter_id) const
|
||||
bool Compiler::buffer_get_hlsl_counter_buffer(VariableID id, uint32_t &counter_id) const
|
||||
{
|
||||
auto *m = ir.find_meta(id);
|
||||
|
||||
@ -4032,7 +4033,7 @@ const SmallVector<std::string> &Compiler::get_declared_extensions() const
|
||||
return ir.declared_extensions;
|
||||
}
|
||||
|
||||
std::string Compiler::get_remapped_declared_block_name(uint32_t id) const
|
||||
std::string Compiler::get_remapped_declared_block_name(VariableID id) const
|
||||
{
|
||||
return get_remapped_declared_block_name(id, false);
|
||||
}
|
||||
@ -4325,8 +4326,7 @@ bool Compiler::InterlockedResourceAccessHandler::end_function_scope(const uint32
|
||||
|
||||
void Compiler::InterlockedResourceAccessHandler::access_potential_resource(uint32_t id)
|
||||
{
|
||||
if ((use_critical_section && in_crit_sec) ||
|
||||
(control_flow_interlock && call_stack_is_interlocked) ||
|
||||
if ((use_critical_section && in_crit_sec) || (control_flow_interlock && call_stack_is_interlocked) ||
|
||||
split_function_case)
|
||||
{
|
||||
compiler.interlocked_resources.insert(id);
|
||||
@ -4563,8 +4563,8 @@ void Compiler::analyze_interlocked_resource_usage()
|
||||
traverse_all_reachable_opcodes(get<SPIRFunction>(ir.default_entry_point), handler);
|
||||
|
||||
// For GLSL. If we hit any of these cases, we have to fall back to conservative approach.
|
||||
interlocked_is_complex = !handler.use_critical_section ||
|
||||
handler.interlock_function_id != ir.default_entry_point;
|
||||
interlocked_is_complex =
|
||||
!handler.use_critical_section || handler.interlock_function_id != ir.default_entry_point;
|
||||
}
|
||||
}
|
||||
|
||||
@ -4577,7 +4577,7 @@ bool Compiler::type_is_array_of_pointers(const SPIRType &type) const
|
||||
return type.pointer_depth == get<SPIRType>(type.parent_type).pointer_depth;
|
||||
}
|
||||
|
||||
bool Compiler::flush_phi_required(uint32_t from, uint32_t to) const
|
||||
bool Compiler::flush_phi_required(BlockID from, BlockID to) const
|
||||
{
|
||||
auto &child = get<SPIRBlock>(to);
|
||||
for (auto &phi : child.phi_variables)
|
||||
|
116
spirv_cross.hpp
116
spirv_cross.hpp
@ -27,18 +27,18 @@ struct Resource
|
||||
{
|
||||
// Resources are identified with their SPIR-V ID.
|
||||
// This is the ID of the OpVariable.
|
||||
uint32_t id;
|
||||
ID id;
|
||||
|
||||
// The type ID of the variable which includes arrays and all type modifications.
|
||||
// This type ID is not suitable for parsing OpMemberDecoration of a struct and other decorations in general
|
||||
// since these modifications typically happen on the base_type_id.
|
||||
uint32_t type_id;
|
||||
TypeID type_id;
|
||||
|
||||
// The base type of the declared resource.
|
||||
// This type is the base type which ignores pointers and arrays of the type_id.
|
||||
// This is mostly useful to parse decorations of the underlying type.
|
||||
// base_type_id can also be obtained with get_type(get_type(type_id).self).
|
||||
uint32_t base_type_id;
|
||||
TypeID base_type_id;
|
||||
|
||||
// The declared name (OpName) of the resource.
|
||||
// For Buffer blocks, the name actually reflects the externally
|
||||
@ -77,17 +77,17 @@ struct ShaderResources
|
||||
struct CombinedImageSampler
|
||||
{
|
||||
// The ID of the sampler2D variable.
|
||||
uint32_t combined_id;
|
||||
VariableID combined_id;
|
||||
// The ID of the texture2D variable.
|
||||
uint32_t image_id;
|
||||
VariableID image_id;
|
||||
// The ID of the sampler variable.
|
||||
uint32_t sampler_id;
|
||||
VariableID sampler_id;
|
||||
};
|
||||
|
||||
struct SpecializationConstant
|
||||
{
|
||||
// The ID of the specialization constant.
|
||||
uint32_t id;
|
||||
ConstantID id;
|
||||
// The constant ID of the constant, used in Vulkan during pipeline creation.
|
||||
uint32_t constant_id;
|
||||
};
|
||||
@ -142,81 +142,81 @@ public:
|
||||
virtual std::string compile();
|
||||
|
||||
// Gets the identifier (OpName) of an ID. If not defined, an empty string will be returned.
|
||||
const std::string &get_name(uint32_t id) const;
|
||||
const std::string &get_name(ID id) const;
|
||||
|
||||
// Applies a decoration to an ID. Effectively injects OpDecorate.
|
||||
void set_decoration(uint32_t id, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_decoration_string(uint32_t id, spv::Decoration decoration, const std::string &argument);
|
||||
void set_decoration(ID id, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_decoration_string(ID id, spv::Decoration decoration, const std::string &argument);
|
||||
|
||||
// Overrides the identifier OpName of an ID.
|
||||
// Identifiers beginning with underscores or identifiers which contain double underscores
|
||||
// are reserved by the implementation.
|
||||
void set_name(uint32_t id, const std::string &name);
|
||||
void set_name(ID id, const std::string &name);
|
||||
|
||||
// Gets a bitmask for the decorations which are applied to ID.
|
||||
// I.e. (1ull << spv::DecorationFoo) | (1ull << spv::DecorationBar)
|
||||
const Bitset &get_decoration_bitset(uint32_t id) const;
|
||||
const Bitset &get_decoration_bitset(ID id) const;
|
||||
|
||||
// Returns whether the decoration has been applied to the ID.
|
||||
bool has_decoration(uint32_t id, spv::Decoration decoration) const;
|
||||
bool has_decoration(ID id, spv::Decoration decoration) const;
|
||||
|
||||
// Gets the value for decorations which take arguments.
|
||||
// If the decoration is a boolean (i.e. spv::DecorationNonWritable),
|
||||
// 1 will be returned.
|
||||
// If decoration doesn't exist or decoration is not recognized,
|
||||
// 0 will be returned.
|
||||
uint32_t get_decoration(uint32_t id, spv::Decoration decoration) const;
|
||||
const std::string &get_decoration_string(uint32_t id, spv::Decoration decoration) const;
|
||||
uint32_t get_decoration(ID id, spv::Decoration decoration) const;
|
||||
const std::string &get_decoration_string(ID id, spv::Decoration decoration) const;
|
||||
|
||||
// Removes the decoration for an ID.
|
||||
void unset_decoration(uint32_t id, spv::Decoration decoration);
|
||||
void unset_decoration(ID id, spv::Decoration decoration);
|
||||
|
||||
// Gets the SPIR-V type associated with ID.
|
||||
// Mostly used with Resource::type_id and Resource::base_type_id to parse the underlying type of a resource.
|
||||
const SPIRType &get_type(uint32_t id) const;
|
||||
const SPIRType &get_type(TypeID id) const;
|
||||
|
||||
// Gets the SPIR-V type of a variable.
|
||||
const SPIRType &get_type_from_variable(uint32_t id) const;
|
||||
const SPIRType &get_type_from_variable(VariableID id) const;
|
||||
|
||||
// Gets the underlying storage class for an OpVariable.
|
||||
spv::StorageClass get_storage_class(uint32_t id) const;
|
||||
spv::StorageClass get_storage_class(VariableID id) const;
|
||||
|
||||
// If get_name() is an empty string, get the fallback name which will be used
|
||||
// instead in the disassembled source.
|
||||
virtual const std::string get_fallback_name(uint32_t id) const;
|
||||
virtual const std::string get_fallback_name(ID id) const;
|
||||
|
||||
// If get_name() of a Block struct is an empty string, get the fallback name.
|
||||
// This needs to be per-variable as multiple variables can use the same block type.
|
||||
virtual const std::string get_block_fallback_name(uint32_t id) const;
|
||||
virtual const std::string get_block_fallback_name(VariableID id) const;
|
||||
|
||||
// Given an OpTypeStruct in ID, obtain the identifier for member number "index".
|
||||
// This may be an empty string.
|
||||
const std::string &get_member_name(uint32_t id, uint32_t index) const;
|
||||
const std::string &get_member_name(TypeID id, uint32_t index) const;
|
||||
|
||||
// Given an OpTypeStruct in ID, obtain the OpMemberDecoration for member number "index".
|
||||
uint32_t get_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
const std::string &get_member_decoration_string(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
uint32_t get_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
const std::string &get_member_decoration_string(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
|
||||
// Sets the member identifier for OpTypeStruct ID, member number "index".
|
||||
void set_member_name(uint32_t id, uint32_t index, const std::string &name);
|
||||
void set_member_name(TypeID id, uint32_t index, const std::string &name);
|
||||
|
||||
// Returns the qualified member identifier for OpTypeStruct ID, member number "index",
|
||||
// or an empty string if no qualified alias exists
|
||||
const std::string &get_member_qualified_name(uint32_t type_id, uint32_t index) const;
|
||||
const std::string &get_member_qualified_name(TypeID type_id, uint32_t index) const;
|
||||
|
||||
// Gets the decoration mask for a member of a struct, similar to get_decoration_mask.
|
||||
const Bitset &get_member_decoration_bitset(uint32_t id, uint32_t index) const;
|
||||
const Bitset &get_member_decoration_bitset(TypeID id, uint32_t index) const;
|
||||
|
||||
// Returns whether the decoration has been applied to a member of a struct.
|
||||
bool has_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
bool has_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
|
||||
// Similar to set_decoration, but for struct members.
|
||||
void set_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_member_decoration_string(uint32_t id, uint32_t index, spv::Decoration decoration,
|
||||
void set_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_member_decoration_string(TypeID id, uint32_t index, spv::Decoration decoration,
|
||||
const std::string &argument);
|
||||
|
||||
// Unsets a member decoration, similar to unset_decoration.
|
||||
void unset_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration);
|
||||
void unset_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration);
|
||||
|
||||
// Gets the fallback name for a member, similar to get_fallback_name.
|
||||
virtual const std::string get_fallback_member_name(uint32_t index) const
|
||||
@ -228,7 +228,7 @@ public:
|
||||
// SPIR-V shader. The granularity of this analysis is per-member of a struct.
|
||||
// This can be used for Buffer (UBO), BufferBlock/StorageBuffer (SSBO) and PushConstant blocks.
|
||||
// ID is the Resource::id obtained from get_shader_resources().
|
||||
SmallVector<BufferRange> get_active_buffer_ranges(uint32_t id) const;
|
||||
SmallVector<BufferRange> get_active_buffer_ranges(VariableID id) const;
|
||||
|
||||
// Returns the effective size of a buffer block.
|
||||
size_t get_declared_struct_size(const SPIRType &struct_type) const;
|
||||
@ -256,12 +256,12 @@ public:
|
||||
//
|
||||
// To use the returned set as the filter for which variables are used during compilation,
|
||||
// this set can be moved to set_enabled_interface_variables().
|
||||
std::unordered_set<uint32_t> get_active_interface_variables() const;
|
||||
std::unordered_set<VariableID> get_active_interface_variables() const;
|
||||
|
||||
// Sets the interface variables which are used during compilation.
|
||||
// By default, all variables are used.
|
||||
// Once set, compile() will only consider the set in active_variables.
|
||||
void set_enabled_interface_variables(std::unordered_set<uint32_t> active_variables);
|
||||
void set_enabled_interface_variables(std::unordered_set<VariableID> active_variables);
|
||||
|
||||
// Query shader resources, use ids with reflection interface to modify or query binding points, etc.
|
||||
ShaderResources get_shader_resources() const;
|
||||
@ -269,19 +269,19 @@ public:
|
||||
// Query shader resources, but only return the variables which are part of active_variables.
|
||||
// E.g.: get_shader_resources(get_active_variables()) to only return the variables which are statically
|
||||
// accessed.
|
||||
ShaderResources get_shader_resources(const std::unordered_set<uint32_t> &active_variables) const;
|
||||
ShaderResources get_shader_resources(const std::unordered_set<VariableID> &active_variables) const;
|
||||
|
||||
// Remapped variables are considered built-in variables and a backend will
|
||||
// not emit a declaration for this variable.
|
||||
// This is mostly useful for making use of builtins which are dependent on extensions.
|
||||
void set_remapped_variable_state(uint32_t id, bool remap_enable);
|
||||
bool get_remapped_variable_state(uint32_t id) const;
|
||||
void set_remapped_variable_state(VariableID id, bool remap_enable);
|
||||
bool get_remapped_variable_state(VariableID id) const;
|
||||
|
||||
// For subpassInput variables which are remapped to plain variables,
|
||||
// the number of components in the remapped
|
||||
// variable must be specified as the backing type of subpass inputs are opaque.
|
||||
void set_subpass_input_remapped_components(uint32_t id, uint32_t components);
|
||||
uint32_t get_subpass_input_remapped_components(uint32_t id) const;
|
||||
void set_subpass_input_remapped_components(VariableID id, uint32_t components);
|
||||
uint32_t get_subpass_input_remapped_components(VariableID id) const;
|
||||
|
||||
// All operations work on the current entry point.
|
||||
// Entry points can be swapped out with set_entry_point().
|
||||
@ -362,7 +362,7 @@ public:
|
||||
// If the returned ID is non-zero, it can be decorated with set/bindings as desired before calling compile().
|
||||
// Calling this function also invalidates get_active_interface_variables(), so this should be called
|
||||
// before that function.
|
||||
uint32_t build_dummy_sampler_for_combined_images();
|
||||
VariableID build_dummy_sampler_for_combined_images();
|
||||
|
||||
// Analyzes all separate image and samplers used from the currently selected entry point,
|
||||
// and re-routes them all to a combined image sampler instead.
|
||||
@ -411,8 +411,8 @@ public:
|
||||
// constant_type is the SPIRType for the specialization constant,
|
||||
// which can be queried to determine which fields in the unions should be poked at.
|
||||
SmallVector<SpecializationConstant> get_specialization_constants() const;
|
||||
SPIRConstant &get_constant(uint32_t id);
|
||||
const SPIRConstant &get_constant(uint32_t id) const;
|
||||
SPIRConstant &get_constant(ConstantID id);
|
||||
const SPIRConstant &get_constant(ConstantID id) const;
|
||||
|
||||
uint32_t get_current_id_bound() const
|
||||
{
|
||||
@ -435,7 +435,7 @@ public:
|
||||
// If the decoration was declared, sets the word_offset to an offset into the provided SPIR-V binary buffer and returns true,
|
||||
// otherwise, returns false.
|
||||
// If the decoration does not have any value attached to it (e.g. DecorationRelaxedPrecision), this function will also return false.
|
||||
bool get_binary_offset_for_decoration(uint32_t id, spv::Decoration decoration, uint32_t &word_offset) const;
|
||||
bool get_binary_offset_for_decoration(VariableID id, spv::Decoration decoration, uint32_t &word_offset) const;
|
||||
|
||||
// HLSL counter buffer reflection interface.
|
||||
// Append/Consume/Increment/Decrement in HLSL is implemented as two "neighbor" buffer objects where
|
||||
@ -450,7 +450,7 @@ public:
|
||||
// only return true if OpSource was reported HLSL.
|
||||
// To rely on this functionality, ensure that the SPIR-V module is not stripped.
|
||||
|
||||
bool buffer_is_hlsl_counter_buffer(uint32_t id) const;
|
||||
bool buffer_is_hlsl_counter_buffer(VariableID id) const;
|
||||
|
||||
// Queries if a buffer object has a neighbor "counter" buffer.
|
||||
// If so, the ID of that counter buffer will be returned in counter_id.
|
||||
@ -458,7 +458,7 @@ public:
|
||||
// Otherwise, this query is purely based on OpName identifiers as found in the SPIR-V module, and will
|
||||
// only return true if OpSource was reported HLSL.
|
||||
// To rely on this functionality, ensure that the SPIR-V module is not stripped.
|
||||
bool buffer_get_hlsl_counter_buffer(uint32_t id, uint32_t &counter_id) const;
|
||||
bool buffer_get_hlsl_counter_buffer(VariableID id, uint32_t &counter_id) const;
|
||||
|
||||
// Gets the list of all SPIR-V Capabilities which were declared in the SPIR-V module.
|
||||
const SmallVector<spv::Capability> &get_declared_capabilities() const;
|
||||
@ -479,13 +479,13 @@ public:
|
||||
// ID is the name of a variable as returned by Resource::id, and must be a variable with a Block-like type.
|
||||
//
|
||||
// This also applies to HLSL cbuffers.
|
||||
std::string get_remapped_declared_block_name(uint32_t id) const;
|
||||
std::string get_remapped_declared_block_name(VariableID id) const;
|
||||
|
||||
// For buffer block variables, get the decorations for that variable.
|
||||
// Sometimes, decorations for buffer blocks are found in member decorations instead
|
||||
// of direct decorations on the variable itself.
|
||||
// The most common use here is to check if a buffer is readonly or writeonly.
|
||||
Bitset get_buffer_block_flags(uint32_t id) const;
|
||||
Bitset get_buffer_block_flags(VariableID id) const;
|
||||
|
||||
protected:
|
||||
const uint32_t *stream(const Instruction &instr) const
|
||||
@ -509,7 +509,7 @@ protected:
|
||||
|
||||
SPIRFunction *current_function = nullptr;
|
||||
SPIRBlock *current_block = nullptr;
|
||||
std::unordered_set<uint32_t> active_interface_variables;
|
||||
std::unordered_set<VariableID> active_interface_variables;
|
||||
bool check_active_interface_variables = false;
|
||||
|
||||
// If our IDs are out of range here as part of opcodes, throw instead of
|
||||
@ -549,7 +549,9 @@ protected:
|
||||
template <typename T>
|
||||
const T *maybe_get(uint32_t id) const
|
||||
{
|
||||
if (ir.ids[id].get_type() == static_cast<Types>(T::type))
|
||||
if (id >= ir.ids.size())
|
||||
return nullptr;
|
||||
else if (ir.ids[id].get_type() == static_cast<Types>(T::type))
|
||||
return &get<T>(id);
|
||||
else
|
||||
return nullptr;
|
||||
@ -618,7 +620,7 @@ protected:
|
||||
inline bool is_single_block_loop(uint32_t next) const
|
||||
{
|
||||
auto &block = get<SPIRBlock>(next);
|
||||
return block.merge == SPIRBlock::MergeLoop && block.continue_block == next;
|
||||
return block.merge == SPIRBlock::MergeLoop && block.continue_block == ID(next);
|
||||
}
|
||||
|
||||
inline bool is_break(uint32_t next) const
|
||||
@ -748,7 +750,7 @@ protected:
|
||||
|
||||
struct InterfaceVariableAccessHandler : OpcodeHandler
|
||||
{
|
||||
InterfaceVariableAccessHandler(const Compiler &compiler_, std::unordered_set<uint32_t> &variables_)
|
||||
InterfaceVariableAccessHandler(const Compiler &compiler_, std::unordered_set<VariableID> &variables_)
|
||||
: compiler(compiler_)
|
||||
, variables(variables_)
|
||||
{
|
||||
@ -757,7 +759,7 @@ protected:
|
||||
bool handle(spv::Op opcode, const uint32_t *args, uint32_t length) override;
|
||||
|
||||
const Compiler &compiler;
|
||||
std::unordered_set<uint32_t> &variables;
|
||||
std::unordered_set<VariableID> &variables;
|
||||
};
|
||||
|
||||
struct CombinedImageSamplerHandler : OpcodeHandler
|
||||
@ -779,8 +781,8 @@ protected:
|
||||
uint32_t remap_parameter(uint32_t id);
|
||||
void push_remap_parameters(const SPIRFunction &func, const uint32_t *args, uint32_t length);
|
||||
void pop_remap_parameters();
|
||||
void register_combined_image_sampler(SPIRFunction &caller, uint32_t combined_id, uint32_t texture_id,
|
||||
uint32_t sampler_id, bool depth);
|
||||
void register_combined_image_sampler(SPIRFunction &caller, VariableID combined_id, VariableID texture_id,
|
||||
VariableID sampler_id, bool depth);
|
||||
};
|
||||
|
||||
struct DummySamplerForCombinedImageHandler : OpcodeHandler
|
||||
@ -813,7 +815,7 @@ protected:
|
||||
// This must be an ordered data structure so we always pick the same type aliases.
|
||||
SmallVector<uint32_t> global_struct_cache;
|
||||
|
||||
ShaderResources get_shader_resources(const std::unordered_set<uint32_t> *active_variables) const;
|
||||
ShaderResources get_shader_resources(const std::unordered_set<VariableID> *active_variables) const;
|
||||
|
||||
VariableTypeRemapCallback variable_remap_callback;
|
||||
|
||||
@ -985,7 +987,7 @@ protected:
|
||||
struct InterlockedResourceAccessPrepassHandler : OpcodeHandler
|
||||
{
|
||||
InterlockedResourceAccessPrepassHandler(Compiler &compiler_, uint32_t entry_point_id)
|
||||
: compiler(compiler_)
|
||||
: compiler(compiler_)
|
||||
{
|
||||
call_stack.push_back(entry_point_id);
|
||||
}
|
||||
@ -1038,7 +1040,7 @@ protected:
|
||||
bool reflection_ssbo_instance_name_is_significant() const;
|
||||
std::string get_remapped_declared_block_name(uint32_t id, bool fallback_prefer_instance_name) const;
|
||||
|
||||
bool flush_phi_required(uint32_t from, uint32_t to) const;
|
||||
bool flush_phi_required(BlockID from, BlockID to) const;
|
||||
|
||||
private:
|
||||
// Used only to implement the old deprecated get_entry_point() interface.
|
||||
|
@ -162,7 +162,7 @@ struct spvc_compiler_options_s : ScratchMemoryAllocation
|
||||
|
||||
struct spvc_set_s : ScratchMemoryAllocation
|
||||
{
|
||||
std::unordered_set<uint32_t> set;
|
||||
std::unordered_set<VariableID> set;
|
||||
};
|
||||
|
||||
// Dummy-inherit to we can keep our opaque type handle type safe in C-land as well,
|
||||
|
@ -162,7 +162,7 @@ static string ensure_valid_identifier(const string &name, bool member)
|
||||
return str;
|
||||
}
|
||||
|
||||
const string &ParsedIR::get_name(uint32_t id) const
|
||||
const string &ParsedIR::get_name(ID id) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (m)
|
||||
@ -171,7 +171,7 @@ const string &ParsedIR::get_name(uint32_t id) const
|
||||
return empty_string;
|
||||
}
|
||||
|
||||
const string &ParsedIR::get_member_name(uint32_t id, uint32_t index) const
|
||||
const string &ParsedIR::get_member_name(TypeID id, uint32_t index) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (m)
|
||||
@ -184,7 +184,7 @@ const string &ParsedIR::get_member_name(uint32_t id, uint32_t index) const
|
||||
return empty_string;
|
||||
}
|
||||
|
||||
void ParsedIR::set_name(uint32_t id, const string &name)
|
||||
void ParsedIR::set_name(ID id, const string &name)
|
||||
{
|
||||
auto &str = meta[id].decoration.alias;
|
||||
str.clear();
|
||||
@ -199,7 +199,7 @@ void ParsedIR::set_name(uint32_t id, const string &name)
|
||||
str = ensure_valid_identifier(name, false);
|
||||
}
|
||||
|
||||
void ParsedIR::set_member_name(uint32_t id, uint32_t index, const string &name)
|
||||
void ParsedIR::set_member_name(TypeID id, uint32_t index, const string &name)
|
||||
{
|
||||
meta[id].members.resize(max(meta[id].members.size(), size_t(index) + 1));
|
||||
|
||||
@ -215,7 +215,7 @@ void ParsedIR::set_member_name(uint32_t id, uint32_t index, const string &name)
|
||||
str = ensure_valid_identifier(name, true);
|
||||
}
|
||||
|
||||
void ParsedIR::set_decoration_string(uint32_t id, Decoration decoration, const string &argument)
|
||||
void ParsedIR::set_decoration_string(ID id, Decoration decoration, const string &argument)
|
||||
{
|
||||
auto &dec = meta[id].decoration;
|
||||
dec.decoration_flags.set(decoration);
|
||||
@ -231,7 +231,7 @@ void ParsedIR::set_decoration_string(uint32_t id, Decoration decoration, const s
|
||||
}
|
||||
}
|
||||
|
||||
void ParsedIR::set_decoration(uint32_t id, Decoration decoration, uint32_t argument)
|
||||
void ParsedIR::set_decoration(ID id, Decoration decoration, uint32_t argument)
|
||||
{
|
||||
auto &dec = meta[id].decoration;
|
||||
dec.decoration_flags.set(decoration);
|
||||
@ -297,7 +297,7 @@ void ParsedIR::set_decoration(uint32_t id, Decoration decoration, uint32_t argum
|
||||
}
|
||||
}
|
||||
|
||||
void ParsedIR::set_member_decoration(uint32_t id, uint32_t index, Decoration decoration, uint32_t argument)
|
||||
void ParsedIR::set_member_decoration(TypeID id, uint32_t index, Decoration decoration, uint32_t argument)
|
||||
{
|
||||
meta[id].members.resize(max(meta[id].members.size(), size_t(index) + 1));
|
||||
auto &dec = meta[id].members[index];
|
||||
@ -345,7 +345,7 @@ void ParsedIR::set_member_decoration(uint32_t id, uint32_t index, Decoration dec
|
||||
|
||||
// Recursively marks any constants referenced by the specified constant instruction as being used
|
||||
// as an array length. The id must be a constant instruction (SPIRConstant or SPIRConstantOp).
|
||||
void ParsedIR::mark_used_as_array_length(uint32_t id)
|
||||
void ParsedIR::mark_used_as_array_length(ID id)
|
||||
{
|
||||
switch (ids[id].get_type())
|
||||
{
|
||||
@ -401,7 +401,7 @@ Bitset ParsedIR::get_buffer_block_flags(const SPIRVariable &var) const
|
||||
return base_flags;
|
||||
}
|
||||
|
||||
const Bitset &ParsedIR::get_member_decoration_bitset(uint32_t id, uint32_t index) const
|
||||
const Bitset &ParsedIR::get_member_decoration_bitset(TypeID id, uint32_t index) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (m)
|
||||
@ -414,12 +414,12 @@ const Bitset &ParsedIR::get_member_decoration_bitset(uint32_t id, uint32_t index
|
||||
return cleared_bitset;
|
||||
}
|
||||
|
||||
bool ParsedIR::has_decoration(uint32_t id, Decoration decoration) const
|
||||
bool ParsedIR::has_decoration(ID id, Decoration decoration) const
|
||||
{
|
||||
return get_decoration_bitset(id).get(decoration);
|
||||
}
|
||||
|
||||
uint32_t ParsedIR::get_decoration(uint32_t id, Decoration decoration) const
|
||||
uint32_t ParsedIR::get_decoration(ID id, Decoration decoration) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (!m)
|
||||
@ -460,7 +460,7 @@ uint32_t ParsedIR::get_decoration(uint32_t id, Decoration decoration) const
|
||||
}
|
||||
}
|
||||
|
||||
const string &ParsedIR::get_decoration_string(uint32_t id, Decoration decoration) const
|
||||
const string &ParsedIR::get_decoration_string(ID id, Decoration decoration) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (!m)
|
||||
@ -481,7 +481,7 @@ const string &ParsedIR::get_decoration_string(uint32_t id, Decoration decoration
|
||||
}
|
||||
}
|
||||
|
||||
void ParsedIR::unset_decoration(uint32_t id, Decoration decoration)
|
||||
void ParsedIR::unset_decoration(ID id, Decoration decoration)
|
||||
{
|
||||
auto &dec = meta[id].decoration;
|
||||
dec.decoration_flags.clear(decoration);
|
||||
@ -543,12 +543,12 @@ void ParsedIR::unset_decoration(uint32_t id, Decoration decoration)
|
||||
}
|
||||
}
|
||||
|
||||
bool ParsedIR::has_member_decoration(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
bool ParsedIR::has_member_decoration(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
return get_member_decoration_bitset(id, index).get(decoration);
|
||||
}
|
||||
|
||||
uint32_t ParsedIR::get_member_decoration(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
uint32_t ParsedIR::get_member_decoration(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (!m)
|
||||
@ -582,7 +582,7 @@ uint32_t ParsedIR::get_member_decoration(uint32_t id, uint32_t index, Decoration
|
||||
}
|
||||
}
|
||||
|
||||
const Bitset &ParsedIR::get_decoration_bitset(uint32_t id) const
|
||||
const Bitset &ParsedIR::get_decoration_bitset(ID id) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (m)
|
||||
@ -594,7 +594,7 @@ const Bitset &ParsedIR::get_decoration_bitset(uint32_t id) const
|
||||
return cleared_bitset;
|
||||
}
|
||||
|
||||
void ParsedIR::set_member_decoration_string(uint32_t id, uint32_t index, Decoration decoration, const string &argument)
|
||||
void ParsedIR::set_member_decoration_string(TypeID id, uint32_t index, Decoration decoration, const string &argument)
|
||||
{
|
||||
meta[id].members.resize(max(meta[id].members.size(), size_t(index) + 1));
|
||||
auto &dec = meta[id].members[index];
|
||||
@ -611,7 +611,7 @@ void ParsedIR::set_member_decoration_string(uint32_t id, uint32_t index, Decorat
|
||||
}
|
||||
}
|
||||
|
||||
const string &ParsedIR::get_member_decoration_string(uint32_t id, uint32_t index, Decoration decoration) const
|
||||
const string &ParsedIR::get_member_decoration_string(TypeID id, uint32_t index, Decoration decoration) const
|
||||
{
|
||||
auto *m = find_meta(id);
|
||||
if (m)
|
||||
@ -634,7 +634,7 @@ const string &ParsedIR::get_member_decoration_string(uint32_t id, uint32_t index
|
||||
return empty_string;
|
||||
}
|
||||
|
||||
void ParsedIR::unset_member_decoration(uint32_t id, uint32_t index, Decoration decoration)
|
||||
void ParsedIR::unset_member_decoration(TypeID id, uint32_t index, Decoration decoration)
|
||||
{
|
||||
auto &m = meta[id];
|
||||
if (index >= m.members.size())
|
||||
@ -687,7 +687,7 @@ uint32_t ParsedIR::increase_bound_by(uint32_t incr_amount)
|
||||
return uint32_t(curr_bound);
|
||||
}
|
||||
|
||||
void ParsedIR::remove_typed_id(Types type, uint32_t id)
|
||||
void ParsedIR::remove_typed_id(Types type, ID id)
|
||||
{
|
||||
auto &type_ids = ids_for_type[type];
|
||||
type_ids.erase(remove(begin(type_ids), end(type_ids), id), end(type_ids));
|
||||
@ -702,7 +702,7 @@ void ParsedIR::reset_all_of_type(Types type)
|
||||
ids_for_type[type].clear();
|
||||
}
|
||||
|
||||
void ParsedIR::add_typed_id(Types type, uint32_t id)
|
||||
void ParsedIR::add_typed_id(Types type, ID id)
|
||||
{
|
||||
if (loop_iteration_depth_hard != 0)
|
||||
SPIRV_CROSS_THROW("Cannot add typed ID while looping over it.");
|
||||
@ -748,7 +748,7 @@ void ParsedIR::add_typed_id(Types type, uint32_t id)
|
||||
}
|
||||
}
|
||||
|
||||
const Meta *ParsedIR::find_meta(uint32_t id) const
|
||||
const Meta *ParsedIR::find_meta(ID id) const
|
||||
{
|
||||
auto itr = meta.find(id);
|
||||
if (itr != end(meta))
|
||||
@ -757,7 +757,7 @@ const Meta *ParsedIR::find_meta(uint32_t id) const
|
||||
return nullptr;
|
||||
}
|
||||
|
||||
Meta *ParsedIR::find_meta(uint32_t id)
|
||||
Meta *ParsedIR::find_meta(ID id)
|
||||
{
|
||||
auto itr = meta.find(id);
|
||||
if (itr != end(meta))
|
||||
|
@ -57,19 +57,19 @@ public:
|
||||
SmallVector<Variant> ids;
|
||||
|
||||
// Various meta data for IDs, decorations, names, etc.
|
||||
std::unordered_map<uint32_t, Meta> meta;
|
||||
std::unordered_map<ID, Meta> meta;
|
||||
|
||||
// Holds all IDs which have a certain type.
|
||||
// This is needed so we can iterate through a specific kind of resource quickly,
|
||||
// and in-order of module declaration.
|
||||
SmallVector<uint32_t> ids_for_type[TypeCount];
|
||||
SmallVector<ID> ids_for_type[TypeCount];
|
||||
|
||||
// Special purpose lists which contain a union of types.
|
||||
// This is needed so we can declare specialization constants and structs in an interleaved fashion,
|
||||
// among other things.
|
||||
// Constants can be of struct type, and struct array sizes can use specialization constants.
|
||||
SmallVector<uint32_t> ids_for_constant_or_type;
|
||||
SmallVector<uint32_t> ids_for_constant_or_variable;
|
||||
SmallVector<ID> ids_for_constant_or_type;
|
||||
SmallVector<ID> ids_for_constant_or_variable;
|
||||
|
||||
// Declared capabilities and extensions in the SPIR-V module.
|
||||
// Not really used except for reflection at the moment.
|
||||
@ -88,12 +88,12 @@ public:
|
||||
};
|
||||
using BlockMetaFlags = uint8_t;
|
||||
SmallVector<BlockMetaFlags> block_meta;
|
||||
std::unordered_map<uint32_t, uint32_t> continue_block_to_loop_header;
|
||||
std::unordered_map<BlockID, BlockID> continue_block_to_loop_header;
|
||||
|
||||
// Normally, we'd stick SPIREntryPoint in ids array, but it conflicts with SPIRFunction.
|
||||
// Entry points can therefore be seen as some sort of meta structure.
|
||||
std::unordered_map<uint32_t, SPIREntryPoint> entry_points;
|
||||
uint32_t default_entry_point = 0;
|
||||
std::unordered_map<FunctionID, SPIREntryPoint> entry_points;
|
||||
FunctionID default_entry_point = 0;
|
||||
|
||||
struct Source
|
||||
{
|
||||
@ -114,34 +114,34 @@ public:
|
||||
// Can be useful for simple "raw" reflection.
|
||||
// However, most members are here because the Parser needs most of these,
|
||||
// and might as well just have the whole suite of decoration/name handling in one place.
|
||||
void set_name(uint32_t id, const std::string &name);
|
||||
const std::string &get_name(uint32_t id) const;
|
||||
void set_decoration(uint32_t id, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_decoration_string(uint32_t id, spv::Decoration decoration, const std::string &argument);
|
||||
bool has_decoration(uint32_t id, spv::Decoration decoration) const;
|
||||
uint32_t get_decoration(uint32_t id, spv::Decoration decoration) const;
|
||||
const std::string &get_decoration_string(uint32_t id, spv::Decoration decoration) const;
|
||||
const Bitset &get_decoration_bitset(uint32_t id) const;
|
||||
void unset_decoration(uint32_t id, spv::Decoration decoration);
|
||||
void set_name(ID id, const std::string &name);
|
||||
const std::string &get_name(ID id) const;
|
||||
void set_decoration(ID id, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_decoration_string(ID id, spv::Decoration decoration, const std::string &argument);
|
||||
bool has_decoration(ID id, spv::Decoration decoration) const;
|
||||
uint32_t get_decoration(ID id, spv::Decoration decoration) const;
|
||||
const std::string &get_decoration_string(ID id, spv::Decoration decoration) const;
|
||||
const Bitset &get_decoration_bitset(ID id) const;
|
||||
void unset_decoration(ID id, spv::Decoration decoration);
|
||||
|
||||
// Decoration handling methods (for members of a struct).
|
||||
void set_member_name(uint32_t id, uint32_t index, const std::string &name);
|
||||
const std::string &get_member_name(uint32_t id, uint32_t index) const;
|
||||
void set_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_member_decoration_string(uint32_t id, uint32_t index, spv::Decoration decoration,
|
||||
void set_member_name(TypeID id, uint32_t index, const std::string &name);
|
||||
const std::string &get_member_name(TypeID id, uint32_t index) const;
|
||||
void set_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration, uint32_t argument = 0);
|
||||
void set_member_decoration_string(TypeID id, uint32_t index, spv::Decoration decoration,
|
||||
const std::string &argument);
|
||||
uint32_t get_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
const std::string &get_member_decoration_string(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
bool has_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration) const;
|
||||
const Bitset &get_member_decoration_bitset(uint32_t id, uint32_t index) const;
|
||||
void unset_member_decoration(uint32_t id, uint32_t index, spv::Decoration decoration);
|
||||
uint32_t get_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
const std::string &get_member_decoration_string(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
bool has_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration) const;
|
||||
const Bitset &get_member_decoration_bitset(TypeID id, uint32_t index) const;
|
||||
void unset_member_decoration(TypeID id, uint32_t index, spv::Decoration decoration);
|
||||
|
||||
void mark_used_as_array_length(uint32_t id);
|
||||
void mark_used_as_array_length(ID id);
|
||||
uint32_t increase_bound_by(uint32_t count);
|
||||
Bitset get_buffer_block_flags(const SPIRVariable &var) const;
|
||||
|
||||
void add_typed_id(Types type, uint32_t id);
|
||||
void remove_typed_id(Types type, uint32_t id);
|
||||
void add_typed_id(Types type, ID id);
|
||||
void remove_typed_id(Types type, ID id);
|
||||
|
||||
class LoopLock
|
||||
{
|
||||
@ -198,8 +198,8 @@ public:
|
||||
|
||||
void reset_all_of_type(Types type);
|
||||
|
||||
Meta *find_meta(uint32_t id);
|
||||
const Meta *find_meta(uint32_t id) const;
|
||||
Meta *find_meta(ID id);
|
||||
const Meta *find_meta(ID id) const;
|
||||
|
||||
const std::string &get_empty_string() const
|
||||
{
|
||||
|
@ -767,7 +767,8 @@ void CompilerGLSL::emit_header()
|
||||
|
||||
// If there are any spec constants on legacy GLSL, defer declaration, we need to set up macro
|
||||
// declarations before we can emit the work group size.
|
||||
if (options.vulkan_semantics || ((wg_x.id == 0) && (wg_y.id == 0) && (wg_z.id == 0)))
|
||||
if (options.vulkan_semantics ||
|
||||
((wg_x.id == ConstantID(0)) && (wg_y.id == ConstantID(0)) && (wg_z.id == ConstantID(0))))
|
||||
build_workgroup_size(inputs, wg_x, wg_y, wg_z);
|
||||
}
|
||||
else
|
||||
@ -864,7 +865,8 @@ void CompilerGLSL::emit_struct(SPIRType &type)
|
||||
// Type-punning with these types is legal, which complicates things
|
||||
// when we are storing struct and array types in an SSBO for example.
|
||||
// If the type master is packed however, we can no longer assume that the struct declaration will be redundant.
|
||||
if (type.type_alias != 0 && !has_extended_decoration(type.type_alias, SPIRVCrossDecorationBufferBlockRepacked))
|
||||
if (type.type_alias != TypeID(0) &&
|
||||
!has_extended_decoration(type.type_alias, SPIRVCrossDecorationBufferBlockRepacked))
|
||||
return;
|
||||
|
||||
add_resource_name(type.self);
|
||||
@ -2132,7 +2134,7 @@ void CompilerGLSL::emit_constant(const SPIRConstant &constant)
|
||||
auto name = to_name(constant.self);
|
||||
|
||||
SpecializationConstant wg_x, wg_y, wg_z;
|
||||
uint32_t workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
ID workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
|
||||
// This specialization constant is implicitly declared by emitting layout() in;
|
||||
if (constant.self == workgroup_size_id)
|
||||
@ -2141,7 +2143,8 @@ void CompilerGLSL::emit_constant(const SPIRConstant &constant)
|
||||
// These specialization constants are implicitly declared by emitting layout() in;
|
||||
// In legacy GLSL, we will still need to emit macros for these, so a layout() in; declaration
|
||||
// later can use macro overrides for work group size.
|
||||
bool is_workgroup_size_constant = constant.self == wg_x.id || constant.self == wg_y.id || constant.self == wg_z.id;
|
||||
bool is_workgroup_size_constant = ConstantID(constant.self) == wg_x.id || ConstantID(constant.self) == wg_y.id ||
|
||||
ConstantID(constant.self) == wg_z.id;
|
||||
|
||||
if (options.vulkan_semantics && is_workgroup_size_constant)
|
||||
{
|
||||
@ -2491,7 +2494,7 @@ void CompilerGLSL::declare_undefined_values()
|
||||
|
||||
bool CompilerGLSL::variable_is_lut(const SPIRVariable &var) const
|
||||
{
|
||||
bool statically_assigned = var.statically_assigned && var.static_expression != 0 && var.remapped_variable;
|
||||
bool statically_assigned = var.statically_assigned && var.static_expression != ID(0) && var.remapped_variable;
|
||||
|
||||
if (statically_assigned)
|
||||
{
|
||||
@ -2620,7 +2623,7 @@ void CompilerGLSL::emit_resources()
|
||||
SpecializationConstant wg_x, wg_y, wg_z;
|
||||
get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
|
||||
if ((wg_x.id != 0) || (wg_y.id != 0) || (wg_z.id != 0))
|
||||
if ((wg_x.id != ConstantID(0)) || (wg_y.id != ConstantID(0)) || (wg_z.id != ConstantID(0)))
|
||||
{
|
||||
SmallVector<string> inputs;
|
||||
build_workgroup_size(inputs, wg_x, wg_y, wg_z);
|
||||
@ -4732,7 +4735,7 @@ void CompilerGLSL::emit_mix_op(uint32_t result_type, uint32_t id, uint32_t left,
|
||||
emit_trinary_func_op(result_type, id, left, right, lerp, "mix");
|
||||
}
|
||||
|
||||
string CompilerGLSL::to_combined_image_sampler(uint32_t image_id, uint32_t samp_id)
|
||||
string CompilerGLSL::to_combined_image_sampler(VariableID image_id, VariableID samp_id)
|
||||
{
|
||||
// Keep track of the array indices we have used to load the image.
|
||||
// We'll need to use the same array index into the combined image sampler array.
|
||||
@ -4754,18 +4757,18 @@ string CompilerGLSL::to_combined_image_sampler(uint32_t image_id, uint32_t samp_
|
||||
samp_id = samp->self;
|
||||
|
||||
auto image_itr = find_if(begin(args), end(args),
|
||||
[image_id](const SPIRFunction::Parameter ¶m) { return param.id == image_id; });
|
||||
[image_id](const SPIRFunction::Parameter ¶m) { return image_id == param.id; });
|
||||
|
||||
auto sampler_itr = find_if(begin(args), end(args),
|
||||
[samp_id](const SPIRFunction::Parameter ¶m) { return param.id == samp_id; });
|
||||
[samp_id](const SPIRFunction::Parameter ¶m) { return samp_id == param.id; });
|
||||
|
||||
if (image_itr != end(args) || sampler_itr != end(args))
|
||||
{
|
||||
// If any parameter originates from a parameter, we will find it in our argument list.
|
||||
bool global_image = image_itr == end(args);
|
||||
bool global_sampler = sampler_itr == end(args);
|
||||
uint32_t iid = global_image ? image_id : uint32_t(image_itr - begin(args));
|
||||
uint32_t sid = global_sampler ? samp_id : uint32_t(sampler_itr - begin(args));
|
||||
VariableID iid = global_image ? image_id : VariableID(image_itr - begin(args));
|
||||
VariableID sid = global_sampler ? samp_id : VariableID(sampler_itr - begin(args));
|
||||
|
||||
auto &combined = current_function->combined_parameters;
|
||||
auto itr = find_if(begin(combined), end(combined), [=](const SPIRFunction::CombinedImageSamplerParameter &p) {
|
||||
@ -4879,7 +4882,7 @@ std::string CompilerGLSL::to_texture_op(const Instruction &i, bool *forward,
|
||||
uint32_t length = i.length;
|
||||
|
||||
uint32_t result_type_id = ops[0];
|
||||
uint32_t img = ops[2];
|
||||
VariableID img = ops[2];
|
||||
uint32_t coord = ops[3];
|
||||
uint32_t dref = 0;
|
||||
uint32_t comp = 0;
|
||||
@ -5036,7 +5039,7 @@ std::string CompilerGLSL::to_texture_op(const Instruction &i, bool *forward,
|
||||
{
|
||||
bool image_is_depth = false;
|
||||
const auto *combined = maybe_get<SPIRCombinedImageSampler>(img);
|
||||
uint32_t image_id = combined ? combined->image : img;
|
||||
VariableID image_id = combined ? combined->image : img;
|
||||
|
||||
if (combined && image_is_comparison(imgtype, combined->image))
|
||||
image_is_depth = true;
|
||||
@ -5078,7 +5081,7 @@ bool CompilerGLSL::expression_is_constant_null(uint32_t id) const
|
||||
|
||||
// Returns the function name for a texture sampling function for the specified image and sampling characteristics.
|
||||
// For some subclasses, the function is a method on the specified image.
|
||||
string CompilerGLSL::to_function_name(uint32_t tex, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
string CompilerGLSL::to_function_name(VariableID tex, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
bool is_proj, bool has_array_offsets, bool has_offset, bool has_grad, bool,
|
||||
uint32_t lod, uint32_t minlod)
|
||||
{
|
||||
@ -5169,7 +5172,7 @@ std::string CompilerGLSL::convert_separate_image_to_expression(uint32_t id)
|
||||
}
|
||||
|
||||
// Returns the function args for a texture sampling function for the specified image and sampling characteristics.
|
||||
string CompilerGLSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
string CompilerGLSL::to_function_args(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
bool is_proj, uint32_t coord, uint32_t coord_components, uint32_t dref,
|
||||
uint32_t grad_x, uint32_t grad_y, uint32_t lod, uint32_t coffset, uint32_t offset,
|
||||
uint32_t bias, uint32_t comp, uint32_t sample, uint32_t /*minlod*/,
|
||||
@ -7587,7 +7590,7 @@ bool CompilerGLSL::remove_unity_swizzle(uint32_t base, string &op)
|
||||
|
||||
string CompilerGLSL::build_composite_combiner(uint32_t return_type, const uint32_t *elems, uint32_t length)
|
||||
{
|
||||
uint32_t base = 0;
|
||||
ID base = 0;
|
||||
string op;
|
||||
string subop;
|
||||
|
||||
@ -7648,7 +7651,7 @@ string CompilerGLSL::build_composite_combiner(uint32_t return_type, const uint32
|
||||
subop = to_composite_constructor_expression(elems[i]);
|
||||
}
|
||||
|
||||
base = e ? e->base_expression : 0;
|
||||
base = e ? e->base_expression : ID(0);
|
||||
}
|
||||
|
||||
if (swizzle_optimization)
|
||||
@ -8010,7 +8013,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
auto &expr = set<SPIRExpression>(ops[1], move(e), ops[0], should_forward(ops[2]));
|
||||
|
||||
auto *backing_variable = maybe_get_backing_variable(ops[2]);
|
||||
expr.loaded_from = backing_variable ? backing_variable->self : ops[2];
|
||||
expr.loaded_from = backing_variable ? backing_variable->self : ID(ops[2]);
|
||||
expr.need_transpose = meta.need_transpose;
|
||||
expr.access_chain = true;
|
||||
|
||||
@ -8144,8 +8147,8 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
|
||||
for (auto &combined : callee.combined_parameters)
|
||||
{
|
||||
uint32_t image_id = combined.global_image ? combined.image_id : arg[combined.image_id];
|
||||
uint32_t sampler_id = combined.global_sampler ? combined.sampler_id : arg[combined.sampler_id];
|
||||
auto image_id = combined.global_image ? combined.image_id : VariableID(arg[combined.image_id]);
|
||||
auto sampler_id = combined.global_sampler ? combined.sampler_id : VariableID(arg[combined.sampler_id]);
|
||||
arglist.push_back(to_combined_image_sampler(image_id, sampler_id));
|
||||
}
|
||||
|
||||
@ -8452,7 +8455,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
if (pointer)
|
||||
{
|
||||
auto *var = maybe_get_backing_variable(rhs);
|
||||
e.loaded_from = var ? var->self : 0;
|
||||
e.loaded_from = var ? var->self : ID(0);
|
||||
}
|
||||
|
||||
// If we're copying an access chain, need to inherit the read expressions.
|
||||
@ -8507,7 +8510,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
// a value we might not need, and bog down codegen.
|
||||
SPIRConstant c;
|
||||
c.constant_type = type0.parent_type;
|
||||
assert(type0.parent_type != 0);
|
||||
assert(type0.parent_type != ID(0));
|
||||
args.push_back(constant_expression(c));
|
||||
}
|
||||
else if (elems[i] >= type0.vecsize)
|
||||
@ -9388,7 +9391,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
|
||||
// When using the image, we need to know which variable it is actually loaded from.
|
||||
auto *var = maybe_get_backing_variable(ops[2]);
|
||||
e.loaded_from = var ? var->self : 0;
|
||||
e.loaded_from = var ? var->self : ID(0);
|
||||
break;
|
||||
}
|
||||
|
||||
@ -9611,7 +9614,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
|
||||
|
||||
// When using the pointer, we need to know which variable it is actually loaded from.
|
||||
auto *var = maybe_get_backing_variable(ops[2]);
|
||||
e.loaded_from = var ? var->self : 0;
|
||||
e.loaded_from = var ? var->self : ID(0);
|
||||
break;
|
||||
}
|
||||
|
||||
@ -10983,7 +10986,7 @@ void CompilerGLSL::require_extension_internal(const string &ext)
|
||||
}
|
||||
}
|
||||
|
||||
void CompilerGLSL::flatten_buffer_block(uint32_t id)
|
||||
void CompilerGLSL::flatten_buffer_block(VariableID id)
|
||||
{
|
||||
auto &var = get<SPIRVariable>(id);
|
||||
auto &type = get<SPIRType>(var.basetype);
|
||||
@ -11304,7 +11307,7 @@ void CompilerGLSL::emit_fixup()
|
||||
}
|
||||
}
|
||||
|
||||
void CompilerGLSL::flush_phi(uint32_t from, uint32_t to)
|
||||
void CompilerGLSL::flush_phi(BlockID from, BlockID to)
|
||||
{
|
||||
auto &child = get<SPIRBlock>(to);
|
||||
if (child.ignore_phi_from_block == from)
|
||||
@ -11333,7 +11336,7 @@ void CompilerGLSL::flush_phi(uint32_t from, uint32_t to)
|
||||
// This is judged to be extremely rare, so deal with it here using a simple, but suboptimal algorithm.
|
||||
bool need_saved_temporary =
|
||||
find_if(itr + 1, end(child.phi_variables), [&](const SPIRBlock::Phi &future_phi) -> bool {
|
||||
return future_phi.local_variable == phi.function_variable && future_phi.parent == from;
|
||||
return future_phi.local_variable == ID(phi.function_variable) && future_phi.parent == from;
|
||||
}) != end(child.phi_variables);
|
||||
|
||||
if (need_saved_temporary)
|
||||
@ -11368,7 +11371,7 @@ void CompilerGLSL::flush_phi(uint32_t from, uint32_t to)
|
||||
}
|
||||
}
|
||||
|
||||
void CompilerGLSL::branch_to_continue(uint32_t from, uint32_t to)
|
||||
void CompilerGLSL::branch_to_continue(BlockID from, BlockID to)
|
||||
{
|
||||
auto &to_block = get<SPIRBlock>(to);
|
||||
if (from == to)
|
||||
@ -11398,7 +11401,7 @@ void CompilerGLSL::branch_to_continue(uint32_t from, uint32_t to)
|
||||
// so just use "self" here.
|
||||
loop_dominator = from;
|
||||
}
|
||||
else if (from_block.loop_dominator != SPIRBlock::NoDominator)
|
||||
else if (from_block.loop_dominator != BlockID(SPIRBlock::NoDominator))
|
||||
{
|
||||
loop_dominator = from_block.loop_dominator;
|
||||
}
|
||||
@ -11422,7 +11425,7 @@ void CompilerGLSL::branch_to_continue(uint32_t from, uint32_t to)
|
||||
}
|
||||
}
|
||||
|
||||
void CompilerGLSL::branch(uint32_t from, uint32_t to)
|
||||
void CompilerGLSL::branch(BlockID from, BlockID to)
|
||||
{
|
||||
flush_phi(from, to);
|
||||
flush_control_dependent_expressions(from);
|
||||
@ -11444,7 +11447,8 @@ void CompilerGLSL::branch(uint32_t from, uint32_t to)
|
||||
// Only sensible solution is to make a ladder variable, which we declare at the top of the switch block,
|
||||
// write to the ladder here, and defer the break.
|
||||
// The loop we're breaking out of must dominate the switch block, or there is no ladder breaking case.
|
||||
if (current_emitting_switch && is_loop_break(to) && current_emitting_switch->loop_dominator != ~0u &&
|
||||
if (current_emitting_switch && is_loop_break(to) &&
|
||||
current_emitting_switch->loop_dominator != BlockID(SPIRBlock::NoDominator) &&
|
||||
get<SPIRBlock>(current_emitting_switch->loop_dominator).merge_block == to)
|
||||
{
|
||||
if (!current_emitting_switch->need_ladder_break)
|
||||
@ -11486,10 +11490,10 @@ void CompilerGLSL::branch(uint32_t from, uint32_t to)
|
||||
// Inner scope always takes precedence.
|
||||
}
|
||||
|
||||
void CompilerGLSL::branch(uint32_t from, uint32_t cond, uint32_t true_block, uint32_t false_block)
|
||||
void CompilerGLSL::branch(BlockID from, uint32_t cond, BlockID true_block, BlockID false_block)
|
||||
{
|
||||
auto &from_block = get<SPIRBlock>(from);
|
||||
uint32_t merge_block = from_block.merge == SPIRBlock::MergeSelection ? from_block.next_block : 0;
|
||||
BlockID merge_block = from_block.merge == SPIRBlock::MergeSelection ? from_block.next_block : BlockID(0);
|
||||
|
||||
// If we branch directly to a selection merge target, we don't need a code path.
|
||||
// This covers both merge out of if () / else () as well as a break for switch blocks.
|
||||
@ -11888,12 +11892,12 @@ void CompilerGLSL::flush_undeclared_variables(SPIRBlock &block)
|
||||
flush_variable_declaration(v);
|
||||
}
|
||||
|
||||
void CompilerGLSL::emit_hoisted_temporaries(SmallVector<pair<uint32_t, uint32_t>> &temporaries)
|
||||
void CompilerGLSL::emit_hoisted_temporaries(SmallVector<pair<TypeID, ID>> &temporaries)
|
||||
{
|
||||
// If we need to force temporaries for certain IDs due to continue blocks, do it before starting loop header.
|
||||
// Need to sort these to ensure that reference output is stable.
|
||||
sort(begin(temporaries), end(temporaries),
|
||||
[](const pair<uint32_t, uint32_t> &a, const pair<uint32_t, uint32_t> &b) { return a.second < b.second; });
|
||||
[](const pair<TypeID, ID> &a, const pair<TypeID, ID> &b) { return a.second < b.second; });
|
||||
|
||||
for (auto &tmp : temporaries)
|
||||
{
|
||||
@ -12347,7 +12351,7 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block)
|
||||
}
|
||||
|
||||
if (!cfg.node_terminates_control_flow_in_sub_graph(current_function->entry_block, block.self) ||
|
||||
block.loop_dominator != SPIRBlock::NoDominator)
|
||||
block.loop_dominator != BlockID(SPIRBlock::NoDominator))
|
||||
{
|
||||
statement("return;");
|
||||
}
|
||||
@ -12360,7 +12364,7 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block)
|
||||
}
|
||||
}
|
||||
else if (!cfg.node_terminates_control_flow_in_sub_graph(current_function->entry_block, block.self) ||
|
||||
block.loop_dominator != SPIRBlock::NoDominator)
|
||||
block.loop_dominator != BlockID(SPIRBlock::NoDominator))
|
||||
{
|
||||
// If this block is the very final block and not called from control flow,
|
||||
// we do not need an explicit return which looks out of place. Just end the function here.
|
||||
@ -12407,7 +12411,7 @@ void CompilerGLSL::emit_block_chain(SPIRBlock &block)
|
||||
assert(block.merge == SPIRBlock::MergeSelection);
|
||||
branch_to_continue(block.self, block.next_block);
|
||||
}
|
||||
else if (block.self != block.next_block)
|
||||
else if (BlockID(block.self) != block.next_block)
|
||||
emit_block_chain(get<SPIRBlock>(block.next_block));
|
||||
}
|
||||
}
|
||||
@ -12782,10 +12786,11 @@ void CompilerGLSL::reorder_type_alias()
|
||||
for (auto alias_itr = begin(type_ids); alias_itr != end(type_ids); ++alias_itr)
|
||||
{
|
||||
auto &type = get<SPIRType>(*alias_itr);
|
||||
if (type.type_alias != 0 && !has_extended_decoration(type.type_alias, SPIRVCrossDecorationBufferBlockRepacked))
|
||||
if (type.type_alias != TypeID(0) &&
|
||||
!has_extended_decoration(type.type_alias, SPIRVCrossDecorationBufferBlockRepacked))
|
||||
{
|
||||
// We will skip declaring this type, so make sure the type_alias type comes before.
|
||||
auto master_itr = find(begin(type_ids), end(type_ids), type.type_alias);
|
||||
auto master_itr = find(begin(type_ids), end(type_ids), ID(type.type_alias));
|
||||
assert(master_itr != end(type_ids));
|
||||
|
||||
if (alias_itr < master_itr)
|
||||
|
@ -209,7 +209,7 @@ public:
|
||||
// For this to work, all types in the block must be the same basic type, e.g. mixing vec2 and vec4 is fine, but
|
||||
// mixing int and float is not.
|
||||
// The name of the uniform array will be the same as the interface block name.
|
||||
void flatten_buffer_block(uint32_t id);
|
||||
void flatten_buffer_block(VariableID id);
|
||||
|
||||
protected:
|
||||
void reset();
|
||||
@ -259,10 +259,10 @@ protected:
|
||||
virtual void emit_fixup();
|
||||
virtual std::string variable_decl(const SPIRType &type, const std::string &name, uint32_t id = 0);
|
||||
virtual std::string to_func_call_arg(const SPIRFunction::Parameter &arg, uint32_t id);
|
||||
virtual std::string to_function_name(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
virtual std::string to_function_name(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
bool is_proj, bool has_array_offsets, bool has_offset, bool has_grad,
|
||||
bool has_dref, uint32_t lod, uint32_t minlod);
|
||||
virtual std::string to_function_args(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
virtual std::string to_function_args(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
bool is_proj, uint32_t coord, uint32_t coord_components, uint32_t dref,
|
||||
uint32_t grad_x, uint32_t grad_y, uint32_t lod, uint32_t coffset,
|
||||
uint32_t offset, uint32_t bias, uint32_t comp, uint32_t sample,
|
||||
@ -428,17 +428,17 @@ protected:
|
||||
void emit_interface_block(const SPIRVariable &type);
|
||||
void emit_flattened_io_block(const SPIRVariable &var, const char *qual);
|
||||
void emit_block_chain(SPIRBlock &block);
|
||||
void emit_hoisted_temporaries(SmallVector<std::pair<uint32_t, uint32_t>> &temporaries);
|
||||
void emit_hoisted_temporaries(SmallVector<std::pair<TypeID, ID>> &temporaries);
|
||||
std::string constant_value_macro_name(uint32_t id);
|
||||
void emit_constant(const SPIRConstant &constant);
|
||||
void emit_specialization_constant_op(const SPIRConstantOp &constant);
|
||||
std::string emit_continue_block(uint32_t continue_block, bool follow_true_block, bool follow_false_block);
|
||||
bool attempt_emit_loop_header(SPIRBlock &block, SPIRBlock::Method method);
|
||||
|
||||
void branch(uint32_t from, uint32_t to);
|
||||
void branch_to_continue(uint32_t from, uint32_t to);
|
||||
void branch(uint32_t from, uint32_t cond, uint32_t true_block, uint32_t false_block);
|
||||
void flush_phi(uint32_t from, uint32_t to);
|
||||
void branch(BlockID from, BlockID to);
|
||||
void branch_to_continue(BlockID from, BlockID to);
|
||||
void branch(BlockID from, uint32_t cond, BlockID true_block, BlockID false_block);
|
||||
void flush_phi(BlockID from, BlockID to);
|
||||
void flush_variable_declaration(uint32_t id);
|
||||
void flush_undeclared_variables(SPIRBlock &block);
|
||||
void emit_variable_temporary_copies(const SPIRVariable &var);
|
||||
@ -543,7 +543,7 @@ protected:
|
||||
virtual std::string layout_for_member(const SPIRType &type, uint32_t index);
|
||||
virtual std::string to_interpolation_qualifiers(const Bitset &flags);
|
||||
std::string layout_for_variable(const SPIRVariable &variable);
|
||||
std::string to_combined_image_sampler(uint32_t image_id, uint32_t samp_id);
|
||||
std::string to_combined_image_sampler(VariableID image_id, VariableID samp_id);
|
||||
virtual bool skip_argument(uint32_t id) const;
|
||||
virtual void emit_array_copy(const std::string &lhs, uint32_t rhs_id, spv::StorageClass lhs_storage,
|
||||
spv::StorageClass rhs_storage);
|
||||
|
@ -1045,7 +1045,7 @@ void CompilerHLSL::emit_specialization_constants_and_structs()
|
||||
{
|
||||
bool emitted = false;
|
||||
SpecializationConstant wg_x, wg_y, wg_z;
|
||||
uint32_t workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
ID workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
|
||||
auto loop_lock = ir.create_loop_hard_lock();
|
||||
for (auto &id_ : ir.ids_for_constant_or_type)
|
||||
@ -2489,7 +2489,7 @@ void CompilerHLSL::emit_texture_op(const Instruction &i)
|
||||
|
||||
uint32_t result_type = ops[0];
|
||||
uint32_t id = ops[1];
|
||||
uint32_t img = ops[2];
|
||||
VariableID img = ops[2];
|
||||
uint32_t coord = ops[3];
|
||||
uint32_t dref = 0;
|
||||
uint32_t comp = 0;
|
||||
@ -3724,7 +3724,7 @@ void CompilerHLSL::emit_access_chain(const Instruction &instruction)
|
||||
e.row_major_matrix = row_major_matrix;
|
||||
e.matrix_stride = matrix_stride;
|
||||
e.immutable = should_forward(ops[2]);
|
||||
e.loaded_from = backing_variable ? backing_variable->self : 0;
|
||||
e.loaded_from = backing_variable ? backing_variable->self : ID(0);
|
||||
|
||||
if (chain)
|
||||
{
|
||||
@ -4494,7 +4494,7 @@ void CompilerHLSL::emit_instruction(const Instruction &instruction)
|
||||
|
||||
// When using the pointer, we need to know which variable it is actually loaded from.
|
||||
auto *var = maybe_get_backing_variable(ops[2]);
|
||||
e.loaded_from = var ? var->self : 0;
|
||||
e.loaded_from = var ? var->self : ID(0);
|
||||
break;
|
||||
}
|
||||
|
||||
@ -4764,7 +4764,7 @@ void CompilerHLSL::add_vertex_attribute_remap(const HLSLVertexAttributeRemap &ve
|
||||
remap_vertex_attributes.push_back(vertex_attributes);
|
||||
}
|
||||
|
||||
uint32_t CompilerHLSL::remap_num_workgroups_builtin()
|
||||
VariableID CompilerHLSL::remap_num_workgroups_builtin()
|
||||
{
|
||||
update_active_builtins();
|
||||
|
||||
|
@ -114,7 +114,7 @@ public:
|
||||
// If non-zero, this returns the variable ID of a cbuffer which corresponds to
|
||||
// the cbuffer declared above. By default, no binding or descriptor set decoration is set,
|
||||
// so the calling application should declare explicit bindings on this ID before calling compile().
|
||||
uint32_t remap_num_workgroups_builtin();
|
||||
VariableID remap_num_workgroups_builtin();
|
||||
|
||||
private:
|
||||
std::string type_to_glsl(const SPIRType &type, uint32_t id = 0) override;
|
||||
|
@ -1433,7 +1433,7 @@ void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, co
|
||||
else if (!strip_array)
|
||||
ir.meta[var.self].decoration.qualified_alias = qual_var_name;
|
||||
|
||||
if (var.storage == StorageClassOutput && var.initializer != 0)
|
||||
if (var.storage == StorageClassOutput && var.initializer != ID(0))
|
||||
{
|
||||
entry_func.fixup_hooks_in.push_back(
|
||||
[=, &var]() { statement(qual_var_name, " = ", to_expression(var.initializer), ";"); });
|
||||
@ -2653,7 +2653,7 @@ void CompilerMSL::mark_scalar_layout_structs(const SPIRType &type)
|
||||
void CompilerMSL::align_struct(SPIRType &ib_type, unordered_set<uint32_t> &aligned_structs)
|
||||
{
|
||||
// We align structs recursively, so stop any redundant work.
|
||||
uint32_t &ib_type_id = ib_type.self;
|
||||
ID &ib_type_id = ib_type.self;
|
||||
if (aligned_structs.count(ib_type_id))
|
||||
return;
|
||||
aligned_structs.insert(ib_type_id);
|
||||
@ -2932,8 +2932,9 @@ void CompilerMSL::emit_store_statement(uint32_t lhs_expression, uint32_t rhs_exp
|
||||
// Special handling when storing to a remapped physical type.
|
||||
// This is mostly to deal with std140 padded matrices or vectors.
|
||||
|
||||
uint32_t physical_type_id =
|
||||
lhs_remapped_type ? get_extended_decoration(lhs_expression, SPIRVCrossDecorationPhysicalTypeID) : type.self;
|
||||
TypeID physical_type_id = lhs_remapped_type ?
|
||||
ID(get_extended_decoration(lhs_expression, SPIRVCrossDecorationPhysicalTypeID)) :
|
||||
type.self;
|
||||
|
||||
auto &physical_type = get<SPIRType>(physical_type_id);
|
||||
|
||||
@ -4555,7 +4556,7 @@ void CompilerMSL::emit_resources()
|
||||
void CompilerMSL::emit_specialization_constants_and_structs()
|
||||
{
|
||||
SpecializationConstant wg_x, wg_y, wg_z;
|
||||
uint32_t workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
ID workgroup_size_id = get_work_group_size_specialization_constants(wg_x, wg_y, wg_z);
|
||||
bool emitted = false;
|
||||
|
||||
unordered_set<uint32_t> declared_structs;
|
||||
@ -4649,7 +4650,7 @@ void CompilerMSL::emit_specialization_constants_and_structs()
|
||||
// Output non-builtin interface structs. These include local function structs
|
||||
// and structs nested within uniform and read-write buffers.
|
||||
auto &type = id.get<SPIRType>();
|
||||
uint32_t type_id = type.self;
|
||||
TypeID type_id = type.self;
|
||||
|
||||
bool is_struct = (type.basetype == SPIRType::Struct) && type.array.empty();
|
||||
bool is_block =
|
||||
@ -4885,7 +4886,7 @@ bool CompilerMSL::emit_tessellation_access_chain(const uint32_t *ops, uint32_t l
|
||||
// expression so we don't try to dereference it as a variable pointer.
|
||||
// Don't do this if the index is a constant 1, though. We need to drop stores
|
||||
// to that one.
|
||||
auto *m = ir.find_meta(var ? var->self : 0);
|
||||
auto *m = ir.find_meta(var ? var->self : ID(0));
|
||||
if (get_execution_model() == ExecutionModelTessellationControl && var && m &&
|
||||
m->decoration.builtin_type == BuiltInTessLevelInner && get_entry_point().flags.get(ExecutionModeTriangles))
|
||||
{
|
||||
@ -6219,7 +6220,7 @@ void CompilerMSL::emit_function_prototype(SPIRFunction &func, const Bitset &)
|
||||
|
||||
local_variable_names = resource_names;
|
||||
|
||||
processing_entry_point = (func.self == ir.default_entry_point);
|
||||
processing_entry_point = func.self == ir.default_entry_point;
|
||||
|
||||
string decl = processing_entry_point ? "" : "inline ";
|
||||
|
||||
@ -6264,7 +6265,7 @@ void CompilerMSL::emit_function_prototype(SPIRFunction &func, const Bitset &)
|
||||
for (auto var_id : vars_needing_early_declaration)
|
||||
{
|
||||
auto &ed_var = get<SPIRVariable>(var_id);
|
||||
uint32_t &initializer = ed_var.initializer;
|
||||
ID &initializer = ed_var.initializer;
|
||||
if (!initializer)
|
||||
initializer = ir.increase_bound_by(1);
|
||||
|
||||
@ -6341,14 +6342,14 @@ static bool needs_chroma_reconstruction(const MSLConstexprSampler *constexpr_sam
|
||||
}
|
||||
|
||||
// Returns the texture sampling function string for the specified image and sampling characteristics.
|
||||
string CompilerMSL::to_function_name(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool, bool,
|
||||
string CompilerMSL::to_function_name(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool, bool,
|
||||
bool, bool, bool has_dref, uint32_t, uint32_t)
|
||||
{
|
||||
const MSLConstexprSampler *constexpr_sampler = nullptr;
|
||||
bool is_dynamic_img_sampler = false;
|
||||
if (auto *var = maybe_get_backing_variable(img))
|
||||
{
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : var->self);
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : VariableID(var->self));
|
||||
is_dynamic_img_sampler = has_extended_decoration(var->self, SPIRVCrossDecorationDynamicImageSampler);
|
||||
}
|
||||
|
||||
@ -6504,16 +6505,16 @@ static inline bool sampling_type_needs_f32_conversion(const SPIRType &type)
|
||||
}
|
||||
|
||||
// Returns the function args for a texture sampling function for the specified image and sampling characteristics.
|
||||
string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool is_proj,
|
||||
uint32_t coord, uint32_t, uint32_t dref, uint32_t grad_x, uint32_t grad_y,
|
||||
uint32_t lod, uint32_t coffset, uint32_t offset, uint32_t bias, uint32_t comp,
|
||||
uint32_t sample, uint32_t minlod, bool *p_forward)
|
||||
string CompilerMSL::to_function_args(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather,
|
||||
bool is_proj, uint32_t coord, uint32_t, uint32_t dref, uint32_t grad_x,
|
||||
uint32_t grad_y, uint32_t lod, uint32_t coffset, uint32_t offset, uint32_t bias,
|
||||
uint32_t comp, uint32_t sample, uint32_t minlod, bool *p_forward)
|
||||
{
|
||||
const MSLConstexprSampler *constexpr_sampler = nullptr;
|
||||
bool is_dynamic_img_sampler = false;
|
||||
if (auto *var = maybe_get_backing_variable(img))
|
||||
{
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : var->self);
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : VariableID(var->self));
|
||||
is_dynamic_img_sampler = has_extended_decoration(var->self, SPIRVCrossDecorationDynamicImageSampler);
|
||||
}
|
||||
|
||||
@ -6959,7 +6960,7 @@ string CompilerMSL::to_texture_op(const Instruction &i, bool *forward, SmallVect
|
||||
bool is_dynamic_img_sampler = false;
|
||||
if (auto *var = maybe_get_backing_variable(img))
|
||||
{
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : var->self);
|
||||
constexpr_sampler = find_constexpr_sampler(var->basevariable ? var->basevariable : VariableID(var->self));
|
||||
is_dynamic_img_sampler = has_extended_decoration(var->self, SPIRVCrossDecorationDynamicImageSampler);
|
||||
}
|
||||
|
||||
@ -7166,7 +7167,7 @@ string CompilerMSL::to_func_call_arg(const SPIRFunction::Parameter &arg, uint32_
|
||||
// so just create a thread local copy in the current function.
|
||||
arg_str = join("_", id, "_array_copy");
|
||||
auto &constants = current_function->constant_arrays_needed_on_stack;
|
||||
auto itr = find(begin(constants), end(constants), id);
|
||||
auto itr = find(begin(constants), end(constants), ID(id));
|
||||
if (itr == end(constants))
|
||||
{
|
||||
force_recompile();
|
||||
@ -7284,7 +7285,7 @@ string CompilerMSL::to_func_call_arg(const SPIRFunction::Parameter &arg, uint32_
|
||||
string CompilerMSL::to_sampler_expression(uint32_t id)
|
||||
{
|
||||
auto *combined = maybe_get<SPIRCombinedImageSampler>(id);
|
||||
auto expr = to_expression(combined ? combined->image : id);
|
||||
auto expr = to_expression(combined ? combined->image : VariableID(id));
|
||||
auto index = expr.find_first_of('[');
|
||||
|
||||
uint32_t samp_id = 0;
|
||||
@ -7305,7 +7306,7 @@ string CompilerMSL::to_swizzle_expression(uint32_t id)
|
||||
{
|
||||
auto *combined = maybe_get<SPIRCombinedImageSampler>(id);
|
||||
|
||||
auto expr = to_expression(combined ? combined->image : id);
|
||||
auto expr = to_expression(combined ? combined->image : VariableID(id));
|
||||
auto index = expr.find_first_of('[');
|
||||
|
||||
// If an image is part of an argument buffer translate this to a legal identifier.
|
||||
@ -10904,7 +10905,7 @@ CompilerMSL::MemberSorter::MemberSorter(SPIRType &t, Meta &m, SortAspect sa)
|
||||
meta.members.resize(max(type.member_types.size(), meta.members.size()));
|
||||
}
|
||||
|
||||
void CompilerMSL::remap_constexpr_sampler(uint32_t id, const MSLConstexprSampler &sampler)
|
||||
void CompilerMSL::remap_constexpr_sampler(VariableID id, const MSLConstexprSampler &sampler)
|
||||
{
|
||||
auto &type = get<SPIRType>(get<SPIRVariable>(id).basetype);
|
||||
if (type.basetype != SPIRType::SampledImage && type.basetype != SPIRType::Sampler)
|
||||
|
@ -363,21 +363,21 @@ public:
|
||||
// buffer if the shader needs it.
|
||||
bool needs_output_buffer() const
|
||||
{
|
||||
return capture_output_to_buffer && stage_out_var_id != 0;
|
||||
return capture_output_to_buffer && stage_out_var_id != ID(0);
|
||||
}
|
||||
|
||||
// Provide feedback to calling API to allow it to pass a patch output
|
||||
// buffer if the shader needs it.
|
||||
bool needs_patch_output_buffer() const
|
||||
{
|
||||
return capture_output_to_buffer && patch_stage_out_var_id != 0;
|
||||
return capture_output_to_buffer && patch_stage_out_var_id != ID(0);
|
||||
}
|
||||
|
||||
// Provide feedback to calling API to allow it to pass an input threadgroup
|
||||
// buffer if the shader needs it.
|
||||
bool needs_input_threadgroup_mem() const
|
||||
{
|
||||
return capture_output_to_buffer && stage_in_var_id != 0;
|
||||
return capture_output_to_buffer && stage_in_var_id != ID(0);
|
||||
}
|
||||
|
||||
explicit CompilerMSL(std::vector<uint32_t> spirv);
|
||||
@ -450,7 +450,7 @@ public:
|
||||
// This can be used on both combined image/samplers (sampler2D) or standalone samplers.
|
||||
// The remapped sampler must not be an array of samplers.
|
||||
// Prefer remap_constexpr_sampler_by_binding unless you're also doing reflection anyways.
|
||||
void remap_constexpr_sampler(uint32_t id, const MSLConstexprSampler &sampler);
|
||||
void remap_constexpr_sampler(VariableID id, const MSLConstexprSampler &sampler);
|
||||
|
||||
// Same as remap_constexpr_sampler, except you provide set/binding, rather than variable ID.
|
||||
// Remaps based on ID take priority over set/binding remaps.
|
||||
@ -551,10 +551,10 @@ protected:
|
||||
std::string builtin_to_glsl(spv::BuiltIn builtin, spv::StorageClass storage) override;
|
||||
std::string to_func_call_arg(const SPIRFunction::Parameter &arg, uint32_t id) override;
|
||||
std::string to_name(uint32_t id, bool allow_alias = true) const override;
|
||||
std::string to_function_name(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool is_proj,
|
||||
std::string to_function_name(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool is_proj,
|
||||
bool has_array_offsets, bool has_offset, bool has_grad, bool has_dref, uint32_t lod,
|
||||
uint32_t minlod) override;
|
||||
std::string to_function_args(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool is_proj,
|
||||
std::string to_function_args(VariableID img, const SPIRType &imgtype, bool is_fetch, bool is_gather, bool is_proj,
|
||||
uint32_t coord, uint32_t coord_components, uint32_t dref, uint32_t grad_x,
|
||||
uint32_t grad_y, uint32_t lod, uint32_t coffset, uint32_t offset, uint32_t bias,
|
||||
uint32_t comp, uint32_t sample, uint32_t minlod, bool *p_forward) override;
|
||||
@ -752,12 +752,12 @@ protected:
|
||||
// Intentionally uninitialized, works around MSVC 2013 bug.
|
||||
uint32_t next_metal_resource_ids[kMaxArgumentBuffers];
|
||||
|
||||
uint32_t stage_in_var_id = 0;
|
||||
uint32_t stage_out_var_id = 0;
|
||||
uint32_t patch_stage_in_var_id = 0;
|
||||
uint32_t patch_stage_out_var_id = 0;
|
||||
uint32_t stage_in_ptr_var_id = 0;
|
||||
uint32_t stage_out_ptr_var_id = 0;
|
||||
VariableID stage_in_var_id = 0;
|
||||
VariableID stage_out_var_id = 0;
|
||||
VariableID patch_stage_in_var_id = 0;
|
||||
VariableID patch_stage_out_var_id = 0;
|
||||
VariableID stage_in_ptr_var_id = 0;
|
||||
VariableID stage_out_ptr_var_id = 0;
|
||||
bool has_sampled_images = false;
|
||||
bool needs_vertex_idx_arg = false;
|
||||
bool needs_instance_idx_arg = false;
|
||||
|
@ -278,7 +278,9 @@ void Parser::parse(const Instruction &instruction)
|
||||
|
||||
// Strings need nul-terminator and consume the whole word.
|
||||
uint32_t strlen_words = uint32_t((e.name.size() + 1 + 3) >> 2);
|
||||
e.interface_variables.insert(end(e.interface_variables), ops + strlen_words + 2, ops + instruction.length);
|
||||
|
||||
for (uint32_t i = strlen_words + 2; i < instruction.length; i++)
|
||||
e.interface_variables.push_back(ops[i]);
|
||||
|
||||
// Set the name of the entry point in case OpName is not provided later.
|
||||
ir.set_name(ops[1], e.name);
|
||||
@ -658,7 +660,7 @@ void Parser::parse(const Instruction &instruction)
|
||||
}
|
||||
}
|
||||
|
||||
if (type.type_alias == 0)
|
||||
if (type.type_alias == TypeID(0))
|
||||
global_struct_cache.push_back(id);
|
||||
}
|
||||
break;
|
||||
@ -1008,12 +1010,12 @@ void Parser::parse(const Instruction &instruction)
|
||||
ir.block_meta[current_block->self] |= ParsedIR::BLOCK_META_LOOP_HEADER_BIT;
|
||||
ir.block_meta[current_block->merge_block] |= ParsedIR::BLOCK_META_LOOP_MERGE_BIT;
|
||||
|
||||
ir.continue_block_to_loop_header[current_block->continue_block] = current_block->self;
|
||||
ir.continue_block_to_loop_header[current_block->continue_block] = BlockID(current_block->self);
|
||||
|
||||
// Don't add loop headers to continue blocks,
|
||||
// which would make it impossible branch into the loop header since
|
||||
// they are treated as continues.
|
||||
if (current_block->continue_block != current_block->self)
|
||||
if (current_block->continue_block != BlockID(current_block->self))
|
||||
ir.block_meta[current_block->continue_block] |= ParsedIR::BLOCK_META_CONTINUE_BIT;
|
||||
|
||||
if (length >= 3)
|
||||
|
@ -285,7 +285,7 @@ void CompilerReflection::emit_type(const SPIRType &type, bool &emitted_open_tag)
|
||||
{
|
||||
auto name = type_to_glsl(type);
|
||||
|
||||
if (type.type_alias != 0)
|
||||
if (type.type_alias != TypeID(0))
|
||||
return;
|
||||
|
||||
if (!emitted_open_tag)
|
||||
@ -468,7 +468,7 @@ void CompilerReflection::emit_resources(const char *tag, const SmallVector<Resou
|
||||
bool is_block = get_decoration_bitset(type.self).get(DecorationBlock) ||
|
||||
get_decoration_bitset(type.self).get(DecorationBufferBlock);
|
||||
|
||||
uint32_t fallback_id = !is_push_constant && is_block ? res.base_type_id : res.id;
|
||||
ID fallback_id = !is_push_constant && is_block ? ID(res.base_type_id) : ID(res.id);
|
||||
|
||||
json_stream->begin_json_object();
|
||||
|
||||
|
49
tests-other/typed_id_test.cpp
Normal file
49
tests-other/typed_id_test.cpp
Normal file
@ -0,0 +1,49 @@
|
||||
#include "spirv_common.hpp"
|
||||
|
||||
using namespace SPIRV_CROSS_NAMESPACE;
|
||||
|
||||
int main()
|
||||
{
|
||||
// Construct from uint32_t.
|
||||
VariableID var_id = 10;
|
||||
TypeID type_id = 20;
|
||||
ConstantID constant_id = 30;
|
||||
|
||||
// Assign from uint32_t.
|
||||
var_id = 100;
|
||||
type_id = 40;
|
||||
constant_id = 60;
|
||||
|
||||
// Construct generic ID.
|
||||
ID generic_var_id = var_id;
|
||||
ID generic_type_id = type_id;
|
||||
ID generic_constant_id = constant_id;
|
||||
|
||||
// Assign generic id.
|
||||
generic_var_id = var_id;
|
||||
generic_type_id = type_id;
|
||||
generic_constant_id = constant_id;
|
||||
|
||||
// Assign generic ID to typed ID
|
||||
var_id = generic_var_id;
|
||||
type_id = generic_type_id;
|
||||
constant_id = generic_constant_id;
|
||||
|
||||
// Implicit conversion to uint32_t.
|
||||
uint32_t a;
|
||||
a = var_id;
|
||||
a = type_id;
|
||||
a = constant_id;
|
||||
a = generic_var_id;
|
||||
a = generic_type_id;
|
||||
a = generic_constant_id;
|
||||
|
||||
// Copy assignment.
|
||||
var_id = VariableID(10);
|
||||
type_id = TypeID(10);
|
||||
constant_id = ConstantID(10);
|
||||
|
||||
// These operations are blocked, assign or construction from mismatched types.
|
||||
//var_id = type_id;
|
||||
//var_id = TypeID(100);
|
||||
}
|
Loading…
Reference in New Issue
Block a user