The bones of what will become a Metal backend.

This isn't anywhere close to usable (not even a blank screen will work), let alone complete. It's about 20% done.
This commit is contained in:
Alex Szpakowski
2020-01-26 15:46:33 -04:00
parent 23bc1a7aae
commit 74cf9b8236
58 changed files with 56720 additions and 15 deletions
+131
View File
@@ -0,0 +1,131 @@
/*
** Copyright (c) 2014-2016 The Khronos Group Inc.
**
** Permission is hereby granted, free of charge, to any person obtaining a copy
** of this software and/or associated documentation files (the "Materials"),
** to deal in the Materials without restriction, including without limitation
** the rights to use, copy, modify, merge, publish, distribute, sublicense,
** and/or sell copies of the Materials, and to permit persons to whom the
** Materials are furnished to do so, subject to the following conditions:
**
** The above copyright notice and this permission notice shall be included in
** all copies or substantial portions of the Materials.
**
** MODIFICATIONS TO THIS FILE MAY MEAN IT NO LONGER ACCURATELY REFLECTS KHRONOS
** STANDARDS. THE UNMODIFIED, NORMATIVE VERSIONS OF KHRONOS SPECIFICATIONS AND
** HEADER INFORMATION ARE LOCATED AT https://www.khronos.org/registry/
**
** THE MATERIALS ARE PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
** OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
** FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL
** THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
** LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING
** FROM,OUT OF OR IN CONNECTION WITH THE MATERIALS OR THE USE OR OTHER DEALINGS
** IN THE MATERIALS.
*/
#ifndef GLSLstd450_H
#define GLSLstd450_H
static const int GLSLstd450Version = 100;
static const int GLSLstd450Revision = 3;
enum GLSLstd450 {
GLSLstd450Bad = 0, // Don't use
GLSLstd450Round = 1,
GLSLstd450RoundEven = 2,
GLSLstd450Trunc = 3,
GLSLstd450FAbs = 4,
GLSLstd450SAbs = 5,
GLSLstd450FSign = 6,
GLSLstd450SSign = 7,
GLSLstd450Floor = 8,
GLSLstd450Ceil = 9,
GLSLstd450Fract = 10,
GLSLstd450Radians = 11,
GLSLstd450Degrees = 12,
GLSLstd450Sin = 13,
GLSLstd450Cos = 14,
GLSLstd450Tan = 15,
GLSLstd450Asin = 16,
GLSLstd450Acos = 17,
GLSLstd450Atan = 18,
GLSLstd450Sinh = 19,
GLSLstd450Cosh = 20,
GLSLstd450Tanh = 21,
GLSLstd450Asinh = 22,
GLSLstd450Acosh = 23,
GLSLstd450Atanh = 24,
GLSLstd450Atan2 = 25,
GLSLstd450Pow = 26,
GLSLstd450Exp = 27,
GLSLstd450Log = 28,
GLSLstd450Exp2 = 29,
GLSLstd450Log2 = 30,
GLSLstd450Sqrt = 31,
GLSLstd450InverseSqrt = 32,
GLSLstd450Determinant = 33,
GLSLstd450MatrixInverse = 34,
GLSLstd450Modf = 35, // second operand needs an OpVariable to write to
GLSLstd450ModfStruct = 36, // no OpVariable operand
GLSLstd450FMin = 37,
GLSLstd450UMin = 38,
GLSLstd450SMin = 39,
GLSLstd450FMax = 40,
GLSLstd450UMax = 41,
GLSLstd450SMax = 42,
GLSLstd450FClamp = 43,
GLSLstd450UClamp = 44,
GLSLstd450SClamp = 45,
GLSLstd450FMix = 46,
GLSLstd450IMix = 47, // Reserved
GLSLstd450Step = 48,
GLSLstd450SmoothStep = 49,
GLSLstd450Fma = 50,
GLSLstd450Frexp = 51, // second operand needs an OpVariable to write to
GLSLstd450FrexpStruct = 52, // no OpVariable operand
GLSLstd450Ldexp = 53,
GLSLstd450PackSnorm4x8 = 54,
GLSLstd450PackUnorm4x8 = 55,
GLSLstd450PackSnorm2x16 = 56,
GLSLstd450PackUnorm2x16 = 57,
GLSLstd450PackHalf2x16 = 58,
GLSLstd450PackDouble2x32 = 59,
GLSLstd450UnpackSnorm2x16 = 60,
GLSLstd450UnpackUnorm2x16 = 61,
GLSLstd450UnpackHalf2x16 = 62,
GLSLstd450UnpackSnorm4x8 = 63,
GLSLstd450UnpackUnorm4x8 = 64,
GLSLstd450UnpackDouble2x32 = 65,
GLSLstd450Length = 66,
GLSLstd450Distance = 67,
GLSLstd450Cross = 68,
GLSLstd450Normalize = 69,
GLSLstd450FaceForward = 70,
GLSLstd450Reflect = 71,
GLSLstd450Refract = 72,
GLSLstd450FindILsb = 73,
GLSLstd450FindSMsb = 74,
GLSLstd450FindUMsb = 75,
GLSLstd450InterpolateAtCentroid = 76,
GLSLstd450InterpolateAtSample = 77,
GLSLstd450InterpolateAtOffset = 78,
GLSLstd450NMin = 79,
GLSLstd450NMax = 80,
GLSLstd450NClamp = 81,
GLSLstd450Count
};
#endif // #ifndef GLSLstd450_H
@@ -0,0 +1,79 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_BARRIER_HPP
#define SPIRV_CROSS_BARRIER_HPP
#include <atomic>
#include <thread>
namespace spirv_cross
{
class Barrier
{
public:
Barrier()
{
count.store(0);
iteration.store(0);
}
void set_release_divisor(unsigned divisor)
{
this->divisor = divisor;
}
static inline void memoryBarrier()
{
std::atomic_thread_fence(std::memory_order_seq_cst);
}
void reset_counter()
{
count.store(0);
iteration.store(0);
}
void wait()
{
unsigned target_iteration = iteration.load(std::memory_order_relaxed) + 1;
// Overflows cleanly.
unsigned target_count = divisor * target_iteration;
// Barriers don't enforce memory ordering.
// Be as relaxed about the barrier as we possibly can!
unsigned c = count.fetch_add(1u, std::memory_order_relaxed);
if (c + 1 == target_count)
{
iteration.store(target_iteration, std::memory_order_relaxed);
}
else
{
// If we have more threads than the CPU, don't hog the CPU for very long periods of time.
while (iteration.load(std::memory_order_relaxed) != target_iteration)
std::this_thread::yield();
}
}
private:
unsigned divisor = 1;
std::atomic<unsigned> count;
std::atomic<unsigned> iteration;
};
}
#endif
@@ -0,0 +1,126 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_EXTERNAL_INTERFACE_H
#define SPIRV_CROSS_EXTERNAL_INTERFACE_H
#ifdef __cplusplus
extern "C" {
#endif
#include <stddef.h>
typedef struct spirv_cross_shader spirv_cross_shader_t;
struct spirv_cross_interface
{
spirv_cross_shader_t *(*construct)(void);
void (*destruct)(spirv_cross_shader_t *thiz);
void (*invoke)(spirv_cross_shader_t *thiz);
};
void spirv_cross_set_stage_input(spirv_cross_shader_t *thiz, unsigned location, void *data, size_t size);
void spirv_cross_set_stage_output(spirv_cross_shader_t *thiz, unsigned location, void *data, size_t size);
void spirv_cross_set_push_constant(spirv_cross_shader_t *thiz, void *data, size_t size);
void spirv_cross_set_uniform_constant(spirv_cross_shader_t *thiz, unsigned location, void *data, size_t size);
void spirv_cross_set_resource(spirv_cross_shader_t *thiz, unsigned set, unsigned binding, void **data, size_t size);
const struct spirv_cross_interface *spirv_cross_get_interface(void);
typedef enum spirv_cross_builtin {
SPIRV_CROSS_BUILTIN_POSITION = 0,
SPIRV_CROSS_BUILTIN_FRAG_COORD = 1,
SPIRV_CROSS_BUILTIN_WORK_GROUP_ID = 2,
SPIRV_CROSS_BUILTIN_NUM_WORK_GROUPS = 3,
SPIRV_CROSS_NUM_BUILTINS
} spirv_cross_builtin;
void spirv_cross_set_builtin(spirv_cross_shader_t *thiz, spirv_cross_builtin builtin, void *data, size_t size);
#define SPIRV_CROSS_NUM_DESCRIPTOR_SETS 4
#define SPIRV_CROSS_NUM_DESCRIPTOR_BINDINGS 16
#define SPIRV_CROSS_NUM_STAGE_INPUTS 16
#define SPIRV_CROSS_NUM_STAGE_OUTPUTS 16
#define SPIRV_CROSS_NUM_UNIFORM_CONSTANTS 32
enum spirv_cross_format
{
SPIRV_CROSS_FORMAT_R8_UNORM = 0,
SPIRV_CROSS_FORMAT_R8G8_UNORM = 1,
SPIRV_CROSS_FORMAT_R8G8B8_UNORM = 2,
SPIRV_CROSS_FORMAT_R8G8B8A8_UNORM = 3,
SPIRV_CROSS_NUM_FORMATS
};
enum spirv_cross_wrap
{
SPIRV_CROSS_WRAP_CLAMP_TO_EDGE = 0,
SPIRV_CROSS_WRAP_REPEAT = 1,
SPIRV_CROSS_NUM_WRAP
};
enum spirv_cross_filter
{
SPIRV_CROSS_FILTER_NEAREST = 0,
SPIRV_CROSS_FILTER_LINEAR = 1,
SPIRV_CROSS_NUM_FILTER
};
enum spirv_cross_mipfilter
{
SPIRV_CROSS_MIPFILTER_BASE = 0,
SPIRV_CROSS_MIPFILTER_NEAREST = 1,
SPIRV_CROSS_MIPFILTER_LINEAR = 2,
SPIRV_CROSS_NUM_MIPFILTER
};
struct spirv_cross_miplevel
{
const void *data;
unsigned width, height;
size_t stride;
};
struct spirv_cross_sampler_info
{
const struct spirv_cross_miplevel *mipmaps;
unsigned num_mipmaps;
enum spirv_cross_format format;
enum spirv_cross_wrap wrap_s;
enum spirv_cross_wrap wrap_t;
enum spirv_cross_filter min_filter;
enum spirv_cross_filter mag_filter;
enum spirv_cross_mipfilter mip_filter;
};
typedef struct spirv_cross_sampler_2d spirv_cross_sampler_2d_t;
spirv_cross_sampler_2d_t *spirv_cross_create_sampler_2d(const struct spirv_cross_sampler_info *info);
void spirv_cross_destroy_sampler_2d(spirv_cross_sampler_2d_t *samp);
#ifdef __cplusplus
}
#endif
#endif
@@ -0,0 +1,62 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_IMAGE_HPP
#define SPIRV_CROSS_IMAGE_HPP
#ifndef GLM_SWIZZLE
#define GLM_SWIZZLE
#endif
#ifndef GLM_FORCE_RADIANS
#define GLM_FORCE_RADIANS
#endif
#include <glm/glm.hpp>
namespace spirv_cross
{
template <typename T>
struct image2DBase
{
virtual ~image2DBase() = default;
inline virtual T load(glm::ivec2 coord) const
{
return T(0, 0, 0, 1);
}
inline virtual void store(glm::ivec2 coord, const T &v)
{
}
};
typedef image2DBase<glm::vec4> image2D;
typedef image2DBase<glm::ivec4> iimage2D;
typedef image2DBase<glm::uvec4> uimage2D;
template <typename T>
inline T imageLoad(const image2DBase<T> &image, glm::ivec2 coord)
{
return image.load(coord);
}
template <typename T>
void imageStore(image2DBase<T> &image, glm::ivec2 coord, const T &value)
{
image.store(coord, value);
}
}
#endif
@@ -0,0 +1,603 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_INTERNAL_INTERFACE_HPP
#define SPIRV_CROSS_INTERNAL_INTERFACE_HPP
// This file must only be included by the shader generated by spirv-cross!
#ifndef GLM_FORCE_SWIZZLE
#define GLM_FORCE_SWIZZLE
#endif
#ifndef GLM_FORCE_RADIANS
#define GLM_FORCE_RADIANS
#endif
#include <glm/glm.hpp>
#include "barrier.hpp"
#include "external_interface.h"
#include "image.hpp"
#include "sampler.hpp"
#include "thread_group.hpp"
#include <assert.h>
#include <stdint.h>
namespace internal
{
// Adaptor helpers to adapt GLSL access chain syntax to C++.
// Don't bother with arrays of arrays on uniforms ...
// Would likely need horribly complex variadic template munging.
template <typename T>
struct Interface
{
enum
{
ArraySize = 1,
Size = sizeof(T)
};
Interface()
: ptr(0)
{
}
T &get()
{
assert(ptr);
return *ptr;
}
T *ptr;
};
// For array types, return a pointer instead.
template <typename T, unsigned U>
struct Interface<T[U]>
{
enum
{
ArraySize = U,
Size = U * sizeof(T)
};
Interface()
: ptr(0)
{
}
T *get()
{
assert(ptr);
return ptr;
}
T *ptr;
};
// For case when array size is 1, avoid double dereference.
template <typename T>
struct PointerInterface
{
enum
{
ArraySize = 1,
Size = sizeof(T *)
};
enum
{
PreDereference = true
};
PointerInterface()
: ptr(0)
{
}
T &get()
{
assert(ptr);
return *ptr;
}
T *ptr;
};
// Automatically converts a pointer down to reference to match GLSL syntax.
template <typename T>
struct DereferenceAdaptor
{
DereferenceAdaptor(T **ptr)
: ptr(ptr)
{
}
T &operator[](unsigned index) const
{
return *(ptr[index]);
}
T **ptr;
};
// We can't have a linear array of T* since T* can be an abstract type in case of samplers.
// We also need a list of pointers since we can have run-time length SSBOs.
template <typename T, unsigned U>
struct PointerInterface<T[U]>
{
enum
{
ArraySize = U,
Size = sizeof(T *) * U
};
enum
{
PreDereference = false
};
PointerInterface()
: ptr(0)
{
}
DereferenceAdaptor<T> get()
{
assert(ptr);
return DereferenceAdaptor<T>(ptr);
}
T **ptr;
};
// Resources can be more abstract and be unsized,
// so we need to have an array of pointers for those cases.
template <typename T>
struct Resource : PointerInterface<T>
{
};
// POD with no unknown sizes, so we can express these as flat arrays.
template <typename T>
struct UniformConstant : Interface<T>
{
};
template <typename T>
struct StageInput : Interface<T>
{
};
template <typename T>
struct StageOutput : Interface<T>
{
};
template <typename T>
struct PushConstant : Interface<T>
{
};
}
struct spirv_cross_shader
{
struct PPSize
{
PPSize()
: ptr(0)
, size(0)
{
}
void **ptr;
size_t size;
};
struct PPSizeResource
{
PPSizeResource()
: ptr(0)
, size(0)
, pre_dereference(false)
{
}
void **ptr;
size_t size;
bool pre_dereference;
};
PPSizeResource resources[SPIRV_CROSS_NUM_DESCRIPTOR_SETS][SPIRV_CROSS_NUM_DESCRIPTOR_BINDINGS];
PPSize stage_inputs[SPIRV_CROSS_NUM_STAGE_INPUTS];
PPSize stage_outputs[SPIRV_CROSS_NUM_STAGE_OUTPUTS];
PPSize uniform_constants[SPIRV_CROSS_NUM_UNIFORM_CONSTANTS];
PPSize push_constant;
PPSize builtins[SPIRV_CROSS_NUM_BUILTINS];
template <typename U>
void register_builtin(spirv_cross_builtin builtin, const U &value)
{
assert(!builtins[builtin].ptr);
builtins[builtin].ptr = (void **)&value.ptr;
builtins[builtin].size = sizeof(*value.ptr) * U::ArraySize;
}
void set_builtin(spirv_cross_builtin builtin, void *data, size_t size)
{
assert(builtins[builtin].ptr);
assert(size >= builtins[builtin].size);
*builtins[builtin].ptr = data;
}
template <typename U>
void register_resource(const internal::Resource<U> &value, unsigned set, unsigned binding)
{
assert(set < SPIRV_CROSS_NUM_DESCRIPTOR_SETS);
assert(binding < SPIRV_CROSS_NUM_DESCRIPTOR_BINDINGS);
assert(!resources[set][binding].ptr);
resources[set][binding].ptr = (void **)&value.ptr;
resources[set][binding].size = internal::Resource<U>::Size;
resources[set][binding].pre_dereference = internal::Resource<U>::PreDereference;
}
template <typename U>
void register_stage_input(const internal::StageInput<U> &value, unsigned location)
{
assert(location < SPIRV_CROSS_NUM_STAGE_INPUTS);
assert(!stage_inputs[location].ptr);
stage_inputs[location].ptr = (void **)&value.ptr;
stage_inputs[location].size = internal::StageInput<U>::Size;
}
template <typename U>
void register_stage_output(const internal::StageOutput<U> &value, unsigned location)
{
assert(location < SPIRV_CROSS_NUM_STAGE_OUTPUTS);
assert(!stage_outputs[location].ptr);
stage_outputs[location].ptr = (void **)&value.ptr;
stage_outputs[location].size = internal::StageOutput<U>::Size;
}
template <typename U>
void register_uniform_constant(const internal::UniformConstant<U> &value, unsigned location)
{
assert(location < SPIRV_CROSS_NUM_UNIFORM_CONSTANTS);
assert(!uniform_constants[location].ptr);
uniform_constants[location].ptr = (void **)&value.ptr;
uniform_constants[location].size = internal::UniformConstant<U>::Size;
}
template <typename U>
void register_push_constant(const internal::PushConstant<U> &value)
{
assert(!push_constant.ptr);
push_constant.ptr = (void **)&value.ptr;
push_constant.size = internal::PushConstant<U>::Size;
}
void set_stage_input(unsigned location, void *data, size_t size)
{
assert(location < SPIRV_CROSS_NUM_STAGE_INPUTS);
assert(stage_inputs[location].ptr);
assert(size >= stage_inputs[location].size);
*stage_inputs[location].ptr = data;
}
void set_stage_output(unsigned location, void *data, size_t size)
{
assert(location < SPIRV_CROSS_NUM_STAGE_OUTPUTS);
assert(stage_outputs[location].ptr);
assert(size >= stage_outputs[location].size);
*stage_outputs[location].ptr = data;
}
void set_uniform_constant(unsigned location, void *data, size_t size)
{
assert(location < SPIRV_CROSS_NUM_UNIFORM_CONSTANTS);
assert(uniform_constants[location].ptr);
assert(size >= uniform_constants[location].size);
*uniform_constants[location].ptr = data;
}
void set_push_constant(void *data, size_t size)
{
assert(push_constant.ptr);
assert(size >= push_constant.size);
*push_constant.ptr = data;
}
void set_resource(unsigned set, unsigned binding, void **data, size_t size)
{
assert(set < SPIRV_CROSS_NUM_DESCRIPTOR_SETS);
assert(binding < SPIRV_CROSS_NUM_DESCRIPTOR_BINDINGS);
assert(resources[set][binding].ptr);
assert(size >= resources[set][binding].size);
// We're using the regular PointerInterface, dereference ahead of time.
if (resources[set][binding].pre_dereference)
*resources[set][binding].ptr = *data;
else
*resources[set][binding].ptr = data;
}
};
namespace spirv_cross
{
template <typename T>
struct BaseShader : spirv_cross_shader
{
void invoke()
{
static_cast<T *>(this)->main();
}
};
struct FragmentResources
{
internal::StageOutput<glm::vec4> gl_FragCoord;
void init(spirv_cross_shader &s)
{
s.register_builtin(SPIRV_CROSS_BUILTIN_FRAG_COORD, gl_FragCoord);
}
#define gl_FragCoord __res->gl_FragCoord.get()
};
template <typename T, typename Res>
struct FragmentShader : BaseShader<FragmentShader<T, Res>>
{
inline void main()
{
impl.main();
}
FragmentShader()
{
resources.init(*this);
impl.__res = &resources;
}
T impl;
Res resources;
};
struct VertexResources
{
internal::StageOutput<glm::vec4> gl_Position;
void init(spirv_cross_shader &s)
{
s.register_builtin(SPIRV_CROSS_BUILTIN_POSITION, gl_Position);
}
#define gl_Position __res->gl_Position.get()
};
template <typename T, typename Res>
struct VertexShader : BaseShader<VertexShader<T, Res>>
{
inline void main()
{
impl.main();
}
VertexShader()
{
resources.init(*this);
impl.__res = &resources;
}
T impl;
Res resources;
};
struct TessEvaluationResources
{
inline void init(spirv_cross_shader &)
{
}
};
template <typename T, typename Res>
struct TessEvaluationShader : BaseShader<TessEvaluationShader<T, Res>>
{
inline void main()
{
impl.main();
}
TessEvaluationShader()
{
resources.init(*this);
impl.__res = &resources;
}
T impl;
Res resources;
};
struct TessControlResources
{
inline void init(spirv_cross_shader &)
{
}
};
template <typename T, typename Res>
struct TessControlShader : BaseShader<TessControlShader<T, Res>>
{
inline void main()
{
impl.main();
}
TessControlShader()
{
resources.init(*this);
impl.__res = &resources;
}
T impl;
Res resources;
};
struct GeometryResources
{
inline void init(spirv_cross_shader &)
{
}
};
template <typename T, typename Res>
struct GeometryShader : BaseShader<GeometryShader<T, Res>>
{
inline void main()
{
impl.main();
}
GeometryShader()
{
resources.init(*this);
impl.__res = &resources;
}
T impl;
Res resources;
};
struct ComputeResources
{
internal::StageInput<glm::uvec3> gl_WorkGroupID__;
internal::StageInput<glm::uvec3> gl_NumWorkGroups__;
void init(spirv_cross_shader &s)
{
s.register_builtin(SPIRV_CROSS_BUILTIN_WORK_GROUP_ID, gl_WorkGroupID__);
s.register_builtin(SPIRV_CROSS_BUILTIN_NUM_WORK_GROUPS, gl_NumWorkGroups__);
}
#define gl_WorkGroupID __res->gl_WorkGroupID__.get()
#define gl_NumWorkGroups __res->gl_NumWorkGroups__.get()
Barrier barrier__;
#define barrier() __res->barrier__.wait()
};
struct ComputePrivateResources
{
uint32_t gl_LocalInvocationIndex__;
#define gl_LocalInvocationIndex __priv_res.gl_LocalInvocationIndex__
glm::uvec3 gl_LocalInvocationID__;
#define gl_LocalInvocationID __priv_res.gl_LocalInvocationID__
glm::uvec3 gl_GlobalInvocationID__;
#define gl_GlobalInvocationID __priv_res.gl_GlobalInvocationID__
};
template <typename T, typename Res, unsigned WorkGroupX, unsigned WorkGroupY, unsigned WorkGroupZ>
struct ComputeShader : BaseShader<ComputeShader<T, Res, WorkGroupX, WorkGroupY, WorkGroupZ>>
{
inline void main()
{
resources.barrier__.reset_counter();
for (unsigned z = 0; z < WorkGroupZ; z++)
for (unsigned y = 0; y < WorkGroupY; y++)
for (unsigned x = 0; x < WorkGroupX; x++)
impl[z][y][x].__priv_res.gl_GlobalInvocationID__ =
glm::uvec3(WorkGroupX, WorkGroupY, WorkGroupZ) * resources.gl_WorkGroupID__.get() +
glm::uvec3(x, y, z);
group.run();
group.wait();
}
ComputeShader()
: group(&impl[0][0][0])
{
resources.init(*this);
resources.barrier__.set_release_divisor(WorkGroupX * WorkGroupY * WorkGroupZ);
unsigned i = 0;
for (unsigned z = 0; z < WorkGroupZ; z++)
{
for (unsigned y = 0; y < WorkGroupY; y++)
{
for (unsigned x = 0; x < WorkGroupX; x++)
{
impl[z][y][x].__priv_res.gl_LocalInvocationID__ = glm::uvec3(x, y, z);
impl[z][y][x].__priv_res.gl_LocalInvocationIndex__ = i++;
impl[z][y][x].__res = &resources;
}
}
}
}
T impl[WorkGroupZ][WorkGroupY][WorkGroupX];
ThreadGroup<T, WorkGroupX * WorkGroupY * WorkGroupZ> group;
Res resources;
};
inline void memoryBarrierShared()
{
Barrier::memoryBarrier();
}
inline void memoryBarrier()
{
Barrier::memoryBarrier();
}
// TODO: Rest of the barriers.
// Atomics
template <typename T>
inline T atomicAdd(T &v, T a)
{
static_assert(sizeof(std::atomic<T>) == sizeof(T), "Cannot cast properly to std::atomic<T>.");
// We need explicit memory barriers in GLSL to enfore any ordering.
// FIXME: Can we really cast this? There is no other way I think ...
return std::atomic_fetch_add_explicit(reinterpret_cast<std::atomic<T> *>(&v), a, std::memory_order_relaxed);
}
}
void spirv_cross_set_stage_input(spirv_cross_shader_t *shader, unsigned location, void *data, size_t size)
{
shader->set_stage_input(location, data, size);
}
void spirv_cross_set_stage_output(spirv_cross_shader_t *shader, unsigned location, void *data, size_t size)
{
shader->set_stage_output(location, data, size);
}
void spirv_cross_set_uniform_constant(spirv_cross_shader_t *shader, unsigned location, void *data, size_t size)
{
shader->set_uniform_constant(location, data, size);
}
void spirv_cross_set_resource(spirv_cross_shader_t *shader, unsigned set, unsigned binding, void **data, size_t size)
{
shader->set_resource(set, binding, data, size);
}
void spirv_cross_set_push_constant(spirv_cross_shader_t *shader, void *data, size_t size)
{
shader->set_push_constant(data, size);
}
void spirv_cross_set_builtin(spirv_cross_shader_t *shader, spirv_cross_builtin builtin, void *data, size_t size)
{
shader->set_builtin(builtin, data, size);
}
#endif
@@ -0,0 +1,105 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_SAMPLER_HPP
#define SPIRV_CROSS_SAMPLER_HPP
#include <vector>
namespace spirv_cross
{
struct spirv_cross_sampler_2d
{
inline virtual ~spirv_cross_sampler_2d()
{
}
};
template <typename T>
struct sampler2DBase : spirv_cross_sampler_2d
{
sampler2DBase(const spirv_cross_sampler_info *info)
{
mips.insert(mips.end(), info->mipmaps, info->mipmaps + info->num_mipmaps);
format = info->format;
wrap_s = info->wrap_s;
wrap_t = info->wrap_t;
min_filter = info->min_filter;
mag_filter = info->mag_filter;
mip_filter = info->mip_filter;
}
inline virtual T sample(glm::vec2 uv, float bias)
{
return sampleLod(uv, bias);
}
inline virtual T sampleLod(glm::vec2 uv, float lod)
{
if (mag_filter == SPIRV_CROSS_FILTER_NEAREST)
{
uv.x = wrap(uv.x, wrap_s, mips[0].width);
uv.y = wrap(uv.y, wrap_t, mips[0].height);
glm::vec2 uv_full = uv * glm::vec2(mips[0].width, mips[0].height);
int x = int(uv_full.x);
int y = int(uv_full.y);
return sample(x, y, 0);
}
else
{
return T(0, 0, 0, 1);
}
}
inline float wrap(float v, spirv_cross_wrap wrap, unsigned size)
{
switch (wrap)
{
case SPIRV_CROSS_WRAP_REPEAT:
return v - glm::floor(v);
case SPIRV_CROSS_WRAP_CLAMP_TO_EDGE:
{
float half = 0.5f / size;
return glm::clamp(v, half, 1.0f - half);
}
default:
return 0.0f;
}
}
std::vector<spirv_cross_miplevel> mips;
spirv_cross_format format;
spirv_cross_wrap wrap_s;
spirv_cross_format wrap_t;
spirv_cross_filter min_filter;
spirv_cross_filter mag_filter;
spirv_cross_mipfilter mip_filter;
};
typedef sampler2DBase<glm::vec4> sampler2D;
typedef sampler2DBase<glm::ivec4> isampler2D;
typedef sampler2DBase<glm::uvec4> usampler2D;
template <typename T>
inline T texture(const sampler2DBase<T> &samp, const glm::vec2 &uv, float bias = 0.0f)
{
return samp.sample(uv, bias);
}
}
#endif
@@ -0,0 +1,113 @@
/*
* Copyright 2015-2017 ARM Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_THREAD_GROUP_HPP
#define SPIRV_CROSS_THREAD_GROUP_HPP
#include <condition_variable>
#include <mutex>
#include <thread>
namespace spirv_cross
{
template <typename T, unsigned Size>
class ThreadGroup
{
public:
ThreadGroup(T *impl)
{
for (unsigned i = 0; i < Size; i++)
workers[i].start(&impl[i]);
}
void run()
{
for (auto &worker : workers)
worker.run();
}
void wait()
{
for (auto &worker : workers)
worker.wait();
}
private:
struct Thread
{
enum State
{
Idle,
Running,
Dying
};
State state = Idle;
void start(T *impl)
{
worker = std::thread([impl, this] {
for (;;)
{
{
std::unique_lock<std::mutex> l{ lock };
cond.wait(l, [this] { return state != Idle; });
if (state == Dying)
break;
}
impl->main();
std::lock_guard<std::mutex> l{ lock };
state = Idle;
cond.notify_one();
}
});
}
void wait()
{
std::unique_lock<std::mutex> l{ lock };
cond.wait(l, [this] { return state == Idle; });
}
void run()
{
std::lock_guard<std::mutex> l{ lock };
state = Running;
cond.notify_one();
}
~Thread()
{
if (worker.joinable())
{
{
std::lock_guard<std::mutex> l{ lock };
state = Dying;
cond.notify_one();
}
worker.join();
}
}
std::thread worker;
std::condition_variable cond;
std::mutex lock;
};
Thread workers[Size];
};
}
#endif
File diff suppressed because it is too large Load Diff
File diff suppressed because it is too large Load Diff
+397
View File
@@ -0,0 +1,397 @@
/*
* Copyright 2016-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "spirv_cfg.hpp"
#include "spirv_cross.hpp"
#include <algorithm>
#include <assert.h>
using namespace std;
namespace SPIRV_CROSS_NAMESPACE
{
CFG::CFG(Compiler &compiler_, const SPIRFunction &func_)
: compiler(compiler_)
, func(func_)
{
build_post_order_visit_order();
build_immediate_dominators();
}
uint32_t CFG::find_common_dominator(uint32_t a, uint32_t b) const
{
while (a != b)
{
if (get_visit_order(a) < get_visit_order(b))
a = get_immediate_dominator(a);
else
b = get_immediate_dominator(b);
}
return a;
}
void CFG::build_immediate_dominators()
{
// Traverse the post-order in reverse and build up the immediate dominator tree.
immediate_dominators.clear();
immediate_dominators[func.entry_block] = func.entry_block;
for (auto i = post_order.size(); i; i--)
{
uint32_t block = post_order[i - 1];
auto &pred = preceding_edges[block];
if (pred.empty()) // This is for the entry block, but we've already set up the dominators.
continue;
for (auto &edge : pred)
{
if (immediate_dominators[block])
{
assert(immediate_dominators[edge]);
immediate_dominators[block] = find_common_dominator(immediate_dominators[block], edge);
}
else
immediate_dominators[block] = edge;
}
}
}
bool CFG::is_back_edge(uint32_t to) const
{
// We have a back edge if the visit order is set with the temporary magic value 0.
// Crossing edges will have already been recorded with a visit order.
auto itr = visit_order.find(to);
return itr != end(visit_order) && itr->second.get() == 0;
}
bool CFG::has_visited_forward_edge(uint32_t to) const
{
// If > 0, we have visited the edge already, and this is not a back edge branch.
auto itr = visit_order.find(to);
return itr != end(visit_order) && itr->second.get() > 0;
}
bool CFG::post_order_visit(uint32_t block_id)
{
// If we have already branched to this block (back edge), stop recursion.
// If our branches are back-edges, we do not record them.
// We have to record crossing edges however.
if (has_visited_forward_edge(block_id))
return true;
else if (is_back_edge(block_id))
return false;
// Block back-edges from recursively revisiting ourselves.
visit_order[block_id].get() = 0;
auto &block = compiler.get<SPIRBlock>(block_id);
// If this is a loop header, add an implied branch to the merge target.
// This is needed to avoid annoying cases with do { ... } while(false) loops often generated by inliners.
// To the CFG, this is linear control flow, but we risk picking the do/while scope as our dominating block.
// This makes sure that if we are accessing a variable outside the do/while, we choose the loop header as dominator.
// We could use has_visited_forward_edge, but this break code-gen where the merge block is unreachable in the CFG.
// Make a point out of visiting merge target first. This is to make sure that post visit order outside the loop
// is lower than inside the loop, which is going to be key for some traversal algorithms like post-dominance analysis.
// For selection constructs true/false blocks will end up visiting the merge block directly and it works out fine,
// but for loops, only the header might end up actually branching to merge block.
if (block.merge == SPIRBlock::MergeLoop && post_order_visit(block.merge_block))
add_branch(block_id, block.merge_block);
// First visit our branch targets.
switch (block.terminator)
{
case SPIRBlock::Direct:
if (post_order_visit(block.next_block))
add_branch(block_id, block.next_block);
break;
case SPIRBlock::Select:
if (post_order_visit(block.true_block))
add_branch(block_id, block.true_block);
if (post_order_visit(block.false_block))
add_branch(block_id, block.false_block);
break;
case SPIRBlock::MultiSelect:
for (auto &target : block.cases)
{
if (post_order_visit(target.block))
add_branch(block_id, target.block);
}
if (block.default_block && post_order_visit(block.default_block))
add_branch(block_id, block.default_block);
break;
default:
break;
}
// If this is a selection merge, add an implied branch to the merge target.
// This is needed to avoid cases where an inner branch dominates the outer branch.
// This can happen if one of the branches exit early, e.g.:
// if (cond) { ...; break; } else { var = 100 } use_var(var);
// We can use the variable without a Phi since there is only one possible parent here.
// However, in this case, we need to hoist out the inner variable to outside the branch.
// Use same strategy as loops.
if (block.merge == SPIRBlock::MergeSelection && post_order_visit(block.next_block))
{
// If there is only one preceding edge to the merge block and it's not ourselves, we need a fixup.
// Add a fake branch so any dominator in either the if (), or else () block, or a lone case statement
// will be hoisted out to outside the selection merge.
// If size > 1, the variable will be automatically hoisted, so we should not mess with it.
// The exception here is switch blocks, where we can have multiple edges to merge block,
// all coming from same scope, so be more conservative in this case.
// Adding fake branches unconditionally breaks parameter preservation analysis,
// which looks at how variables are accessed through the CFG.
auto pred_itr = preceding_edges.find(block.next_block);
if (pred_itr != end(preceding_edges))
{
auto &pred = pred_itr->second;
auto succ_itr = succeeding_edges.find(block_id);
size_t num_succeeding_edges = 0;
if (succ_itr != end(succeeding_edges))
num_succeeding_edges = succ_itr->second.size();
if (block.terminator == SPIRBlock::MultiSelect && num_succeeding_edges == 1)
{
// Multiple branches can come from the same scope due to "break;", so we need to assume that all branches
// come from same case scope in worst case, even if there are multiple preceding edges.
// If we have more than one succeeding edge from the block header, it should be impossible
// to have a dominator be inside the block.
// Only case this can go wrong is if we have 2 or more edges from block header and
// 2 or more edges to merge block, and still have dominator be inside a case label.
if (!pred.empty())
add_branch(block_id, block.next_block);
}
else
{
if (pred.size() == 1 && *pred.begin() != block_id)
add_branch(block_id, block.next_block);
}
}
else
{
// If the merge block does not have any preceding edges, i.e. unreachable, hallucinate it.
// We're going to do code-gen for it, and domination analysis requires that we have at least one preceding edge.
add_branch(block_id, block.next_block);
}
}
// Then visit ourselves. Start counting at one, to let 0 be a magic value for testing back vs. crossing edges.
visit_order[block_id].get() = ++visit_count;
post_order.push_back(block_id);
return true;
}
void CFG::build_post_order_visit_order()
{
uint32_t block = func.entry_block;
visit_count = 0;
visit_order.clear();
post_order.clear();
post_order_visit(block);
}
void CFG::add_branch(uint32_t from, uint32_t to)
{
const auto add_unique = [](SmallVector<uint32_t> &l, uint32_t value) {
auto itr = find(begin(l), end(l), value);
if (itr == end(l))
l.push_back(value);
};
add_unique(preceding_edges[to], from);
add_unique(succeeding_edges[from], to);
}
uint32_t CFG::find_loop_dominator(uint32_t block_id) const
{
while (block_id != SPIRBlock::NoDominator)
{
auto itr = preceding_edges.find(block_id);
if (itr == end(preceding_edges))
return SPIRBlock::NoDominator;
if (itr->second.empty())
return SPIRBlock::NoDominator;
uint32_t pred_block_id = SPIRBlock::NoDominator;
bool ignore_loop_header = false;
// If we are a merge block, go directly to the header block.
// Only consider a loop dominator if we are branching from inside a block to a loop header.
// NOTE: In the CFG we forced an edge from header to merge block always to support variable scopes properly.
for (auto &pred : itr->second)
{
auto &pred_block = compiler.get<SPIRBlock>(pred);
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 == ID(block_id))
{
pred_block_id = pred;
break;
}
}
// No merge block means we can just pick any edge. Loop headers dominate the inner loop, so any path we
// take will lead there.
if (pred_block_id == SPIRBlock::NoDominator)
pred_block_id = itr->second.front();
block_id = pred_block_id;
if (!ignore_loop_header && block_id)
{
auto &block = compiler.get<SPIRBlock>(block_id);
if (block.merge == SPIRBlock::MergeLoop)
return block_id;
}
}
return block_id;
}
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);
BlockID ignore_block_id = 0;
if (from_block.merge == SPIRBlock::MergeLoop)
ignore_block_id = from_block.merge_block;
while (to != from)
{
auto pred_itr = preceding_edges.find(to);
if (pred_itr == end(preceding_edges))
return false;
DominatorBuilder builder(*this);
for (auto &edge : pred_itr->second)
builder.add_block(edge);
uint32_t dominator = builder.get_dominator();
if (dominator == 0)
return false;
auto &dom = compiler.get<SPIRBlock>(dominator);
bool true_path_ignore = false;
bool false_path_ignore = false;
if (ignore_block_id && dom.terminator == SPIRBlock::Select)
{
auto &true_block = compiler.get<SPIRBlock>(dom.true_block);
auto &false_block = compiler.get<SPIRBlock>(dom.false_block);
auto &ignore_block = compiler.get<SPIRBlock>(ignore_block_id);
true_path_ignore = compiler.execution_is_branchless(true_block, ignore_block);
false_path_ignore = compiler.execution_is_branchless(false_block, ignore_block);
}
if ((dom.merge == SPIRBlock::MergeSelection && dom.next_block == to) ||
(dom.merge == SPIRBlock::MergeLoop && dom.merge_block == to) ||
(dom.terminator == SPIRBlock::Direct && dom.next_block == to) ||
(dom.terminator == SPIRBlock::Select && dom.true_block == to && false_path_ignore) ||
(dom.terminator == SPIRBlock::Select && dom.false_block == to && true_path_ignore))
{
// Allow walking selection constructs if the other branch reaches out of a loop construct.
// It cannot be in-scope anymore.
to = dominator;
}
else
return false;
}
return true;
}
DominatorBuilder::DominatorBuilder(const CFG &cfg_)
: cfg(cfg_)
{
}
void DominatorBuilder::add_block(uint32_t block)
{
if (!cfg.get_immediate_dominator(block))
{
// Unreachable block via the CFG, we will never emit this code anyways.
return;
}
if (!dominator)
{
dominator = block;
return;
}
if (block != dominator)
dominator = cfg.find_common_dominator(block, dominator);
}
void DominatorBuilder::lift_continue_block_dominator()
{
// It is possible for a continue block to be the dominator of a variable is only accessed inside the while block of a do-while loop.
// We cannot safely declare variables inside a continue block, so move any variable declared
// in a continue block to the entry block to simplify.
// It makes very little sense for a continue block to ever be a dominator, so fall back to the simplest
// solution.
if (!dominator)
return;
auto &block = cfg.get_compiler().get<SPIRBlock>(dominator);
auto post_order = cfg.get_visit_order(dominator);
// If we are branching to a block with a higher post-order traversal index (continue blocks), we have a problem
// since we cannot create sensible GLSL code for this, fallback to entry block.
bool back_edge_dominator = false;
switch (block.terminator)
{
case SPIRBlock::Direct:
if (cfg.get_visit_order(block.next_block) > post_order)
back_edge_dominator = true;
break;
case SPIRBlock::Select:
if (cfg.get_visit_order(block.true_block) > post_order)
back_edge_dominator = true;
if (cfg.get_visit_order(block.false_block) > post_order)
back_edge_dominator = true;
break;
case SPIRBlock::MultiSelect:
for (auto &target : block.cases)
{
if (cfg.get_visit_order(target.block) > post_order)
back_edge_dominator = true;
}
if (block.default_block && cfg.get_visit_order(block.default_block) > post_order)
back_edge_dominator = true;
break;
default:
break;
}
if (back_edge_dominator)
dominator = cfg.get_function().entry_block;
}
} // namespace SPIRV_CROSS_NAMESPACE
+156
View File
@@ -0,0 +1,156 @@
/*
* Copyright 2016-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_CFG_HPP
#define SPIRV_CROSS_CFG_HPP
#include "spirv_common.hpp"
#include <assert.h>
namespace SPIRV_CROSS_NAMESPACE
{
class Compiler;
class CFG
{
public:
CFG(Compiler &compiler, const SPIRFunction &function);
Compiler &get_compiler()
{
return compiler;
}
const Compiler &get_compiler() const
{
return compiler;
}
const SPIRFunction &get_function() const
{
return func;
}
uint32_t get_immediate_dominator(uint32_t block) const
{
auto itr = immediate_dominators.find(block);
if (itr != std::end(immediate_dominators))
return itr->second;
else
return 0;
}
uint32_t get_visit_order(uint32_t block) const
{
auto itr = visit_order.find(block);
assert(itr != std::end(visit_order));
int v = itr->second.get();
assert(v > 0);
return uint32_t(v);
}
uint32_t find_common_dominator(uint32_t a, uint32_t b) const;
const SmallVector<uint32_t> &get_preceding_edges(uint32_t block) const
{
auto itr = preceding_edges.find(block);
if (itr != std::end(preceding_edges))
return itr->second;
else
return empty_vector;
}
const SmallVector<uint32_t> &get_succeeding_edges(uint32_t block) const
{
auto itr = succeeding_edges.find(block);
if (itr != std::end(succeeding_edges))
return itr->second;
else
return empty_vector;
}
template <typename Op>
void walk_from(std::unordered_set<uint32_t> &seen_blocks, uint32_t block, const Op &op) const
{
if (seen_blocks.count(block))
return;
seen_blocks.insert(block);
if (op(block))
{
for (auto b : get_succeeding_edges(block))
walk_from(seen_blocks, b, op);
}
}
uint32_t find_loop_dominator(uint32_t block) const;
bool node_terminates_control_flow_in_sub_graph(BlockID from, BlockID to) const;
private:
struct VisitOrder
{
int &get()
{
return v;
}
const int &get() const
{
return v;
}
int v = -1;
};
Compiler &compiler;
const SPIRFunction &func;
std::unordered_map<uint32_t, SmallVector<uint32_t>> preceding_edges;
std::unordered_map<uint32_t, SmallVector<uint32_t>> succeeding_edges;
std::unordered_map<uint32_t, uint32_t> immediate_dominators;
std::unordered_map<uint32_t, VisitOrder> visit_order;
SmallVector<uint32_t> post_order;
SmallVector<uint32_t> empty_vector;
void add_branch(uint32_t from, uint32_t to);
void build_post_order_visit_order();
void build_immediate_dominators();
bool post_order_visit(uint32_t block);
uint32_t visit_count = 0;
bool is_back_edge(uint32_t to) const;
bool has_visited_forward_edge(uint32_t to) const;
};
class DominatorBuilder
{
public:
DominatorBuilder(const CFG &cfg);
void add_block(uint32_t block);
uint32_t get_dominator() const
{
return dominator;
}
void lift_continue_block_dominator();
private:
const CFG &cfg;
uint32_t dominator = 0;
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
File diff suppressed because it is too large Load Diff
+549
View File
@@ -0,0 +1,549 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "spirv_cpp.hpp"
using namespace spv;
using namespace SPIRV_CROSS_NAMESPACE;
using namespace std;
void CompilerCPP::emit_buffer_block(const SPIRVariable &var)
{
add_resource_name(var.self);
auto &type = get<SPIRType>(var.basetype);
auto instance_name = to_name(var.self);
uint32_t descriptor_set = ir.meta[var.self].decoration.set;
uint32_t binding = ir.meta[var.self].decoration.binding;
emit_block_struct(type);
auto buffer_name = to_name(type.self);
statement("internal::Resource<", buffer_name, type_to_array_glsl(type), "> ", instance_name, "__;");
statement_no_indent("#define ", instance_name, " __res->", instance_name, "__.get()");
resource_registrations.push_back(
join("s.register_resource(", instance_name, "__", ", ", descriptor_set, ", ", binding, ");"));
statement("");
}
void CompilerCPP::emit_interface_block(const SPIRVariable &var)
{
add_resource_name(var.self);
auto &type = get<SPIRType>(var.basetype);
const char *qual = var.storage == StorageClassInput ? "StageInput" : "StageOutput";
const char *lowerqual = var.storage == StorageClassInput ? "stage_input" : "stage_output";
auto instance_name = to_name(var.self);
uint32_t location = ir.meta[var.self].decoration.location;
string buffer_name;
auto flags = ir.meta[type.self].decoration.decoration_flags;
if (flags.get(DecorationBlock))
{
emit_block_struct(type);
buffer_name = to_name(type.self);
}
else
buffer_name = type_to_glsl(type);
statement("internal::", qual, "<", buffer_name, type_to_array_glsl(type), "> ", instance_name, "__;");
statement_no_indent("#define ", instance_name, " __res->", instance_name, "__.get()");
resource_registrations.push_back(join("s.register_", lowerqual, "(", instance_name, "__", ", ", location, ");"));
statement("");
}
void CompilerCPP::emit_shared(const SPIRVariable &var)
{
add_resource_name(var.self);
auto instance_name = to_name(var.self);
statement(CompilerGLSL::variable_decl(var), ";");
statement_no_indent("#define ", instance_name, " __res->", instance_name);
}
void CompilerCPP::emit_uniform(const SPIRVariable &var)
{
add_resource_name(var.self);
auto &type = get<SPIRType>(var.basetype);
auto instance_name = to_name(var.self);
uint32_t descriptor_set = ir.meta[var.self].decoration.set;
uint32_t binding = ir.meta[var.self].decoration.binding;
uint32_t location = ir.meta[var.self].decoration.location;
string type_name = type_to_glsl(type);
remap_variable_type_name(type, instance_name, type_name);
if (type.basetype == SPIRType::Image || type.basetype == SPIRType::SampledImage ||
type.basetype == SPIRType::AtomicCounter)
{
statement("internal::Resource<", type_name, type_to_array_glsl(type), "> ", instance_name, "__;");
statement_no_indent("#define ", instance_name, " __res->", instance_name, "__.get()");
resource_registrations.push_back(
join("s.register_resource(", instance_name, "__", ", ", descriptor_set, ", ", binding, ");"));
}
else
{
statement("internal::UniformConstant<", type_name, type_to_array_glsl(type), "> ", instance_name, "__;");
statement_no_indent("#define ", instance_name, " __res->", instance_name, "__.get()");
resource_registrations.push_back(
join("s.register_uniform_constant(", instance_name, "__", ", ", location, ");"));
}
statement("");
}
void CompilerCPP::emit_push_constant_block(const SPIRVariable &var)
{
add_resource_name(var.self);
auto &type = get<SPIRType>(var.basetype);
auto &flags = ir.meta[var.self].decoration.decoration_flags;
if (flags.get(DecorationBinding) || flags.get(DecorationDescriptorSet))
SPIRV_CROSS_THROW("Push constant blocks cannot be compiled to GLSL with Binding or Set syntax. "
"Remap to location with reflection API first or disable these decorations.");
emit_block_struct(type);
auto buffer_name = to_name(type.self);
auto instance_name = to_name(var.self);
statement("internal::PushConstant<", buffer_name, type_to_array_glsl(type), "> ", instance_name, ";");
statement_no_indent("#define ", instance_name, " __res->", instance_name, ".get()");
resource_registrations.push_back(join("s.register_push_constant(", instance_name, "__", ");"));
statement("");
}
void CompilerCPP::emit_block_struct(SPIRType &type)
{
// C++ can't do interface blocks, so we fake it by emitting a separate struct.
// However, these structs are not allowed to alias anything, so remove it before
// emitting the struct.
//
// The type we have here needs to be resolved to the non-pointer type so we can remove aliases.
auto &self = get<SPIRType>(type.self);
self.type_alias = 0;
emit_struct(self);
}
void CompilerCPP::emit_resources()
{
for (auto &id : ir.ids)
{
if (id.get_type() == TypeConstant)
{
auto &c = id.get<SPIRConstant>();
bool needs_declaration = c.specialization || c.is_used_as_lut;
if (needs_declaration)
{
if (!options.vulkan_semantics && c.specialization)
{
c.specialization_constant_macro_name =
constant_value_macro_name(get_decoration(c.self, DecorationSpecId));
}
emit_constant(c);
}
}
else if (id.get_type() == TypeConstantOp)
{
emit_specialization_constant_op(id.get<SPIRConstantOp>());
}
}
// Output all basic struct types which are not Block or BufferBlock as these are declared inplace
// when such variables are instantiated.
for (auto &id : ir.ids)
{
if (id.get_type() == TypeType)
{
auto &type = id.get<SPIRType>();
if (type.basetype == SPIRType::Struct && type.array.empty() && !type.pointer &&
(!ir.meta[type.self].decoration.decoration_flags.get(DecorationBlock) &&
!ir.meta[type.self].decoration.decoration_flags.get(DecorationBufferBlock)))
{
emit_struct(type);
}
}
}
statement("struct Resources : ", resource_type);
begin_scope();
// Output UBOs and SSBOs
for (auto &id : ir.ids)
{
if (id.get_type() == TypeVariable)
{
auto &var = id.get<SPIRVariable>();
auto &type = get<SPIRType>(var.basetype);
if (var.storage != StorageClassFunction && type.pointer && type.storage == StorageClassUniform &&
!is_hidden_variable(var) &&
(ir.meta[type.self].decoration.decoration_flags.get(DecorationBlock) ||
ir.meta[type.self].decoration.decoration_flags.get(DecorationBufferBlock)))
{
emit_buffer_block(var);
}
}
}
// Output push constant blocks
for (auto &id : ir.ids)
{
if (id.get_type() == TypeVariable)
{
auto &var = id.get<SPIRVariable>();
auto &type = get<SPIRType>(var.basetype);
if (!is_hidden_variable(var) && var.storage != StorageClassFunction && type.pointer &&
type.storage == StorageClassPushConstant)
{
emit_push_constant_block(var);
}
}
}
// Output in/out interfaces.
for (auto &id : ir.ids)
{
if (id.get_type() == TypeVariable)
{
auto &var = id.get<SPIRVariable>();
auto &type = get<SPIRType>(var.basetype);
if (var.storage != StorageClassFunction && !is_hidden_variable(var) && type.pointer &&
(var.storage == StorageClassInput || var.storage == StorageClassOutput) &&
interface_variable_exists_in_entry_point(var.self))
{
emit_interface_block(var);
}
}
}
// Output Uniform Constants (values, samplers, images, etc).
for (auto &id : ir.ids)
{
if (id.get_type() == TypeVariable)
{
auto &var = id.get<SPIRVariable>();
auto &type = get<SPIRType>(var.basetype);
if (var.storage != StorageClassFunction && !is_hidden_variable(var) && type.pointer &&
(type.storage == StorageClassUniformConstant || type.storage == StorageClassAtomicCounter))
{
emit_uniform(var);
}
}
}
// Global variables.
bool emitted = false;
for (auto global : global_variables)
{
auto &var = get<SPIRVariable>(global);
if (var.storage == StorageClassWorkgroup)
{
emit_shared(var);
emitted = true;
}
}
if (emitted)
statement("");
declare_undefined_values();
statement("inline void init(spirv_cross_shader& s)");
begin_scope();
statement(resource_type, "::init(s);");
for (auto &reg : resource_registrations)
statement(reg);
end_scope();
resource_registrations.clear();
end_scope_decl();
statement("");
statement("Resources* __res;");
if (get_entry_point().model == ExecutionModelGLCompute)
statement("ComputePrivateResources __priv_res;");
statement("");
// Emit regular globals which are allocated per invocation.
emitted = false;
for (auto global : global_variables)
{
auto &var = get<SPIRVariable>(global);
if (var.storage == StorageClassPrivate)
{
if (var.storage == StorageClassWorkgroup)
emit_shared(var);
else
statement(CompilerGLSL::variable_decl(var), ";");
emitted = true;
}
}
if (emitted)
statement("");
}
string CompilerCPP::compile()
{
// Do not deal with ES-isms like precision, older extensions and such.
options.es = false;
options.version = 450;
backend.float_literal_suffix = true;
backend.double_literal_suffix = false;
backend.long_long_literal_suffix = true;
backend.uint32_t_literal_suffix = true;
backend.basic_int_type = "int32_t";
backend.basic_uint_type = "uint32_t";
backend.swizzle_is_function = true;
backend.shared_is_implied = true;
backend.unsized_array_supported = false;
backend.explicit_struct_type = true;
backend.use_initializer_list = true;
fixup_type_alias();
reorder_type_alias();
build_function_control_flow_graphs_and_analyze();
update_active_builtins();
uint32_t pass_count = 0;
do
{
if (pass_count >= 3)
SPIRV_CROSS_THROW("Over 3 compilation loops detected. Must be a bug!");
resource_registrations.clear();
reset();
// Move constructor for this type is broken on GCC 4.9 ...
buffer.reset();
emit_header();
emit_resources();
emit_function(get<SPIRFunction>(ir.default_entry_point), Bitset());
pass_count++;
} while (is_forcing_recompilation());
// Match opening scope of emit_header().
end_scope_decl();
// namespace
end_scope();
// Emit C entry points
emit_c_linkage();
// Entry point in CPP is always main() for the time being.
get_entry_point().name = "main";
return buffer.str();
}
void CompilerCPP::emit_c_linkage()
{
statement("");
statement("spirv_cross_shader_t *spirv_cross_construct(void)");
begin_scope();
statement("return new ", impl_type, "();");
end_scope();
statement("");
statement("void spirv_cross_destruct(spirv_cross_shader_t *shader)");
begin_scope();
statement("delete static_cast<", impl_type, "*>(shader);");
end_scope();
statement("");
statement("void spirv_cross_invoke(spirv_cross_shader_t *shader)");
begin_scope();
statement("static_cast<", impl_type, "*>(shader)->invoke();");
end_scope();
statement("");
statement("static const struct spirv_cross_interface vtable =");
begin_scope();
statement("spirv_cross_construct,");
statement("spirv_cross_destruct,");
statement("spirv_cross_invoke,");
end_scope_decl();
statement("");
statement("const struct spirv_cross_interface *",
interface_name.empty() ? string("spirv_cross_get_interface") : interface_name, "(void)");
begin_scope();
statement("return &vtable;");
end_scope();
}
void CompilerCPP::emit_function_prototype(SPIRFunction &func, const Bitset &)
{
if (func.self != ir.default_entry_point)
add_function_overload(func);
local_variable_names = resource_names;
string decl;
auto &type = get<SPIRType>(func.return_type);
decl += "inline ";
decl += type_to_glsl(type);
decl += " ";
if (func.self == ir.default_entry_point)
{
decl += "main";
processing_entry_point = true;
}
else
decl += to_name(func.self);
decl += "(";
for (auto &arg : func.arguments)
{
add_local_variable_name(arg.id);
decl += argument_decl(arg);
if (&arg != &func.arguments.back())
decl += ", ";
// Hold a pointer to the parameter so we can invalidate the readonly field if needed.
auto *var = maybe_get<SPIRVariable>(arg.id);
if (var)
var->parameter = &arg;
}
decl += ")";
statement(decl);
}
string CompilerCPP::argument_decl(const SPIRFunction::Parameter &arg)
{
auto &type = expression_type(arg.id);
bool constref = !type.pointer || arg.write_count == 0;
auto &var = get<SPIRVariable>(arg.id);
string base = type_to_glsl(type);
string variable_name = to_name(var.self);
remap_variable_type_name(type, variable_name, base);
for (uint32_t i = 0; i < type.array.size(); i++)
base = join("std::array<", base, ", ", to_array_size(type, i), ">");
return join(constref ? "const " : "", base, " &", variable_name);
}
string CompilerCPP::variable_decl(const SPIRType &type, const string &name, uint32_t /* id */)
{
string base = type_to_glsl(type);
remap_variable_type_name(type, name, base);
bool runtime = false;
for (uint32_t i = 0; i < type.array.size(); i++)
{
auto &array = type.array[i];
if (!array && type.array_size_literal[i])
{
// Avoid using runtime arrays with std::array since this is undefined.
// Runtime arrays cannot be passed around as values, so this is fine.
runtime = true;
}
else
base = join("std::array<", base, ", ", to_array_size(type, i), ">");
}
base += ' ';
return base + name + (runtime ? "[1]" : "");
}
void CompilerCPP::emit_header()
{
auto &execution = get_entry_point();
statement("// This C++ shader is autogenerated by spirv-cross.");
statement("#include \"spirv_cross/internal_interface.hpp\"");
statement("#include \"spirv_cross/external_interface.h\"");
// Needed to properly implement GLSL-style arrays.
statement("#include <array>");
statement("#include <stdint.h>");
statement("");
statement("using namespace spirv_cross;");
statement("using namespace glm;");
statement("");
statement("namespace Impl");
begin_scope();
switch (execution.model)
{
case ExecutionModelGeometry:
case ExecutionModelTessellationControl:
case ExecutionModelTessellationEvaluation:
case ExecutionModelGLCompute:
case ExecutionModelFragment:
case ExecutionModelVertex:
statement("struct Shader");
begin_scope();
break;
default:
SPIRV_CROSS_THROW("Unsupported execution model.");
}
switch (execution.model)
{
case ExecutionModelGeometry:
impl_type = "GeometryShader<Impl::Shader, Impl::Shader::Resources>";
resource_type = "GeometryResources";
break;
case ExecutionModelVertex:
impl_type = "VertexShader<Impl::Shader, Impl::Shader::Resources>";
resource_type = "VertexResources";
break;
case ExecutionModelFragment:
impl_type = "FragmentShader<Impl::Shader, Impl::Shader::Resources>";
resource_type = "FragmentResources";
break;
case ExecutionModelGLCompute:
impl_type = join("ComputeShader<Impl::Shader, Impl::Shader::Resources, ", execution.workgroup_size.x, ", ",
execution.workgroup_size.y, ", ", execution.workgroup_size.z, ">");
resource_type = "ComputeResources";
break;
case ExecutionModelTessellationControl:
impl_type = "TessControlShader<Impl::Shader, Impl::Shader::Resources>";
resource_type = "TessControlResources";
break;
case ExecutionModelTessellationEvaluation:
impl_type = "TessEvaluationShader<Impl::Shader, Impl::Shader::Resources>";
resource_type = "TessEvaluationResources";
break;
default:
SPIRV_CROSS_THROW("Unsupported execution model.");
}
}
+86
View File
@@ -0,0 +1,86 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_CPP_HPP
#define SPIRV_CROSS_CPP_HPP
#include "spirv_glsl.hpp"
#include <utility>
namespace SPIRV_CROSS_NAMESPACE
{
class CompilerCPP : public CompilerGLSL
{
public:
explicit CompilerCPP(std::vector<uint32_t> spirv_)
: CompilerGLSL(std::move(spirv_))
{
}
CompilerCPP(const uint32_t *ir_, size_t word_count)
: CompilerGLSL(ir_, word_count)
{
}
explicit CompilerCPP(const ParsedIR &ir_)
: CompilerGLSL(ir_)
{
}
explicit CompilerCPP(ParsedIR &&ir_)
: CompilerGLSL(std::move(ir_))
{
}
std::string compile() override;
// Sets a custom symbol name that can override
// spirv_cross_get_interface.
//
// Useful when several shader interfaces are linked
// statically into the same binary.
void set_interface_name(std::string name)
{
interface_name = std::move(name);
}
private:
void emit_header() override;
void emit_c_linkage();
void emit_function_prototype(SPIRFunction &func, const Bitset &return_flags) override;
void emit_resources();
void emit_buffer_block(const SPIRVariable &type) override;
void emit_push_constant_block(const SPIRVariable &var) override;
void emit_interface_block(const SPIRVariable &type);
void emit_block_chain(SPIRBlock &block);
void emit_uniform(const SPIRVariable &var) override;
void emit_shared(const SPIRVariable &var);
void emit_block_struct(SPIRType &type);
std::string variable_decl(const SPIRType &type, const std::string &name, uint32_t id) override;
std::string argument_decl(const SPIRFunction::Parameter &arg);
SmallVector<std::string> resource_registrations;
std::string impl_type;
std::string resource_type;
uint32_t shared_counter = 0;
std::string interface_name;
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
File diff suppressed because it is too large Load Diff
File diff suppressed because it is too large Load Diff
File diff suppressed because it is too large Load Diff
+859
View File
@@ -0,0 +1,859 @@
/*
* Copyright 2019 Hans-Kristian Arntzen
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_C_API_H
#define SPIRV_CROSS_C_API_H
#include <stddef.h>
#include "spirv.h"
/*
* C89-compatible wrapper for SPIRV-Cross' API.
* Documentation here is sparse unless the behavior does not map 1:1 with C++ API.
* It is recommended to look at the canonical C++ API for more detailed information.
*/
#ifdef __cplusplus
extern "C" {
#endif
/* Bumped if ABI or API breaks backwards compatibility. */
#define SPVC_C_API_VERSION_MAJOR 0
/* Bumped if APIs or enumerations are added in a backwards compatible way. */
#define SPVC_C_API_VERSION_MINOR 21
/* Bumped if internal implementation details change. */
#define SPVC_C_API_VERSION_PATCH 0
#if !defined(SPVC_PUBLIC_API)
#if defined(SPVC_EXPORT_SYMBOLS)
/* Exports symbols. Standard C calling convention is used. */
#if defined(__GNUC__)
#define SPVC_PUBLIC_API __attribute__((visibility("default")))
#elif defined(_MSC_VER)
#define SPVC_PUBLIC_API __declspec(dllexport)
#else
#define SPVC_PUBLIC_API
#endif
#else
#define SPVC_PUBLIC_API
#endif
#endif
/*
* Gets the SPVC_C_API_VERSION_* used to build this library.
* Can be used to check for ABI mismatch if so-versioning did not catch it.
*/
SPVC_PUBLIC_API void spvc_get_version(unsigned *major, unsigned *minor, unsigned *patch);
/* Gets a human readable version string to identify which commit a particular binary was created from. */
SPVC_PUBLIC_API const char *spvc_get_commit_revision_and_timestamp(void);
/* These types are opaque to the user. */
typedef struct spvc_context_s *spvc_context;
typedef struct spvc_parsed_ir_s *spvc_parsed_ir;
typedef struct spvc_compiler_s *spvc_compiler;
typedef struct spvc_compiler_options_s *spvc_compiler_options;
typedef struct spvc_resources_s *spvc_resources;
struct spvc_type_s;
typedef const struct spvc_type_s *spvc_type;
typedef struct spvc_constant_s *spvc_constant;
struct spvc_set_s;
typedef const struct spvc_set_s *spvc_set;
/*
* Shallow typedefs. All SPIR-V IDs are plain 32-bit numbers, but this helps communicate which data is used.
* Maps to a SPIRType.
*/
typedef SpvId spvc_type_id;
/* Maps to a SPIRVariable. */
typedef SpvId spvc_variable_id;
/* Maps to a SPIRConstant. */
typedef SpvId spvc_constant_id;
/* See C++ API. */
typedef struct spvc_reflected_resource
{
spvc_variable_id id;
spvc_type_id base_type_id;
spvc_type_id type_id;
const char *name;
} spvc_reflected_resource;
/* See C++ API. */
typedef struct spvc_entry_point
{
SpvExecutionModel execution_model;
const char *name;
} spvc_entry_point;
/* See C++ API. */
typedef struct spvc_combined_image_sampler
{
spvc_variable_id combined_id;
spvc_variable_id image_id;
spvc_variable_id sampler_id;
} spvc_combined_image_sampler;
/* See C++ API. */
typedef struct spvc_specialization_constant
{
spvc_constant_id id;
unsigned constant_id;
} spvc_specialization_constant;
/* See C++ API. */
typedef struct spvc_buffer_range
{
unsigned index;
size_t offset;
size_t range;
} spvc_buffer_range;
/* See C++ API. */
typedef struct spvc_hlsl_root_constants
{
unsigned start;
unsigned end;
unsigned binding;
unsigned space;
} spvc_hlsl_root_constants;
/* See C++ API. */
typedef struct spvc_hlsl_vertex_attribute_remap
{
unsigned location;
const char *semantic;
} spvc_hlsl_vertex_attribute_remap;
/*
* Be compatible with non-C99 compilers, which do not have stdbool.
* Only recent MSVC compilers supports this for example, and ideally SPIRV-Cross should be linkable
* from a wide range of compilers in its C wrapper.
*/
typedef unsigned char spvc_bool;
#define SPVC_TRUE ((spvc_bool)1)
#define SPVC_FALSE ((spvc_bool)0)
typedef enum spvc_result
{
/* Success. */
SPVC_SUCCESS = 0,
/* The SPIR-V is invalid. Should have been caught by validation ideally. */
SPVC_ERROR_INVALID_SPIRV = -1,
/* The SPIR-V might be valid or invalid, but SPIRV-Cross currently cannot correctly translate this to your target language. */
SPVC_ERROR_UNSUPPORTED_SPIRV = -2,
/* If for some reason we hit this, new or malloc failed. */
SPVC_ERROR_OUT_OF_MEMORY = -3,
/* Invalid API argument. */
SPVC_ERROR_INVALID_ARGUMENT = -4,
SPVC_ERROR_INT_MAX = 0x7fffffff
} spvc_result;
typedef enum spvc_capture_mode
{
/* The Parsed IR payload will be copied, and the handle can be reused to create other compiler instances. */
SPVC_CAPTURE_MODE_COPY = 0,
/*
* The payload will now be owned by the compiler.
* parsed_ir should now be considered a dead blob and must not be used further.
* This is optimal for performance and should be the go-to option.
*/
SPVC_CAPTURE_MODE_TAKE_OWNERSHIP = 1,
SPVC_CAPTURE_MODE_INT_MAX = 0x7fffffff
} spvc_capture_mode;
typedef enum spvc_backend
{
/* This backend can only perform reflection, no compiler options are supported. Maps to spirv_cross::Compiler. */
SPVC_BACKEND_NONE = 0,
SPVC_BACKEND_GLSL = 1, /* spirv_cross::CompilerGLSL */
SPVC_BACKEND_HLSL = 2, /* CompilerHLSL */
SPVC_BACKEND_MSL = 3, /* CompilerMSL */
SPVC_BACKEND_CPP = 4, /* CompilerCPP */
SPVC_BACKEND_JSON = 5, /* CompilerReflection w/ JSON backend */
SPVC_BACKEND_INT_MAX = 0x7fffffff
} spvc_backend;
/* Maps to C++ API. */
typedef enum spvc_resource_type
{
SPVC_RESOURCE_TYPE_UNKNOWN = 0,
SPVC_RESOURCE_TYPE_UNIFORM_BUFFER = 1,
SPVC_RESOURCE_TYPE_STORAGE_BUFFER = 2,
SPVC_RESOURCE_TYPE_STAGE_INPUT = 3,
SPVC_RESOURCE_TYPE_STAGE_OUTPUT = 4,
SPVC_RESOURCE_TYPE_SUBPASS_INPUT = 5,
SPVC_RESOURCE_TYPE_STORAGE_IMAGE = 6,
SPVC_RESOURCE_TYPE_SAMPLED_IMAGE = 7,
SPVC_RESOURCE_TYPE_ATOMIC_COUNTER = 8,
SPVC_RESOURCE_TYPE_PUSH_CONSTANT = 9,
SPVC_RESOURCE_TYPE_SEPARATE_IMAGE = 10,
SPVC_RESOURCE_TYPE_SEPARATE_SAMPLERS = 11,
SPVC_RESOURCE_TYPE_ACCELERATION_STRUCTURE = 12,
SPVC_RESOURCE_TYPE_INT_MAX = 0x7fffffff
} spvc_resource_type;
/* Maps to spirv_cross::SPIRType::BaseType. */
typedef enum spvc_basetype
{
SPVC_BASETYPE_UNKNOWN = 0,
SPVC_BASETYPE_VOID = 1,
SPVC_BASETYPE_BOOLEAN = 2,
SPVC_BASETYPE_INT8 = 3,
SPVC_BASETYPE_UINT8 = 4,
SPVC_BASETYPE_INT16 = 5,
SPVC_BASETYPE_UINT16 = 6,
SPVC_BASETYPE_INT32 = 7,
SPVC_BASETYPE_UINT32 = 8,
SPVC_BASETYPE_INT64 = 9,
SPVC_BASETYPE_UINT64 = 10,
SPVC_BASETYPE_ATOMIC_COUNTER = 11,
SPVC_BASETYPE_FP16 = 12,
SPVC_BASETYPE_FP32 = 13,
SPVC_BASETYPE_FP64 = 14,
SPVC_BASETYPE_STRUCT = 15,
SPVC_BASETYPE_IMAGE = 16,
SPVC_BASETYPE_SAMPLED_IMAGE = 17,
SPVC_BASETYPE_SAMPLER = 18,
SPVC_BASETYPE_ACCELERATION_STRUCTURE = 19,
SPVC_BASETYPE_INT_MAX = 0x7fffffff
} spvc_basetype;
#define SPVC_COMPILER_OPTION_COMMON_BIT 0x1000000
#define SPVC_COMPILER_OPTION_GLSL_BIT 0x2000000
#define SPVC_COMPILER_OPTION_HLSL_BIT 0x4000000
#define SPVC_COMPILER_OPTION_MSL_BIT 0x8000000
#define SPVC_COMPILER_OPTION_LANG_BITS 0x0f000000
#define SPVC_COMPILER_OPTION_ENUM_BITS 0xffffff
#define SPVC_MAKE_MSL_VERSION(major, minor, patch) ((major) * 10000 + (minor) * 100 + (patch))
/* Maps to C++ API. */
typedef enum spvc_msl_platform
{
SPVC_MSL_PLATFORM_IOS = 0,
SPVC_MSL_PLATFORM_MACOS = 1,
SPVC_MSL_PLATFORM_MAX_INT = 0x7fffffff
} spvc_msl_platform;
/* Maps to C++ API. */
typedef enum spvc_msl_vertex_format
{
SPVC_MSL_VERTEX_FORMAT_OTHER = 0,
SPVC_MSL_VERTEX_FORMAT_UINT8 = 1,
SPVC_MSL_VERTEX_FORMAT_UINT16 = 2
} spvc_msl_vertex_format;
/* Maps to C++ API. */
typedef struct spvc_msl_vertex_attribute
{
unsigned location;
unsigned msl_buffer;
unsigned msl_offset;
unsigned msl_stride;
spvc_bool per_instance;
spvc_msl_vertex_format format;
SpvBuiltIn builtin;
} spvc_msl_vertex_attribute;
/*
* Initializes the vertex attribute struct.
*/
SPVC_PUBLIC_API void spvc_msl_vertex_attribute_init(spvc_msl_vertex_attribute *attr);
/* Maps to C++ API. */
typedef struct spvc_msl_resource_binding
{
SpvExecutionModel stage;
unsigned desc_set;
unsigned binding;
unsigned msl_buffer;
unsigned msl_texture;
unsigned msl_sampler;
} spvc_msl_resource_binding;
/*
* Initializes the resource binding struct.
* The defaults are non-zero.
*/
SPVC_PUBLIC_API void spvc_msl_resource_binding_init(spvc_msl_resource_binding *binding);
#define SPVC_MSL_PUSH_CONSTANT_DESC_SET (~(0u))
#define SPVC_MSL_PUSH_CONSTANT_BINDING (0)
#define SPVC_MSL_SWIZZLE_BUFFER_BINDING (~(1u))
#define SPVC_MSL_BUFFER_SIZE_BUFFER_BINDING (~(2u))
#define SPVC_MSL_ARGUMENT_BUFFER_BINDING (~(3u))
/* Obsolete. Sticks around for backwards compatibility. */
#define SPVC_MSL_AUX_BUFFER_STRUCT_VERSION 1
/* Runtime check for incompatibility. Obsolete. */
SPVC_PUBLIC_API unsigned spvc_msl_get_aux_buffer_struct_version(void);
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_coord
{
SPVC_MSL_SAMPLER_COORD_NORMALIZED = 0,
SPVC_MSL_SAMPLER_COORD_PIXEL = 1,
SPVC_MSL_SAMPLER_INT_MAX = 0x7fffffff
} spvc_msl_sampler_coord;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_filter
{
SPVC_MSL_SAMPLER_FILTER_NEAREST = 0,
SPVC_MSL_SAMPLER_FILTER_LINEAR = 1,
SPVC_MSL_SAMPLER_FILTER_INT_MAX = 0x7fffffff
} spvc_msl_sampler_filter;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_mip_filter
{
SPVC_MSL_SAMPLER_MIP_FILTER_NONE = 0,
SPVC_MSL_SAMPLER_MIP_FILTER_NEAREST = 1,
SPVC_MSL_SAMPLER_MIP_FILTER_LINEAR = 2,
SPVC_MSL_SAMPLER_MIP_FILTER_INT_MAX = 0x7fffffff
} spvc_msl_sampler_mip_filter;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_address
{
SPVC_MSL_SAMPLER_ADDRESS_CLAMP_TO_ZERO = 0,
SPVC_MSL_SAMPLER_ADDRESS_CLAMP_TO_EDGE = 1,
SPVC_MSL_SAMPLER_ADDRESS_CLAMP_TO_BORDER = 2,
SPVC_MSL_SAMPLER_ADDRESS_REPEAT = 3,
SPVC_MSL_SAMPLER_ADDRESS_MIRRORED_REPEAT = 4,
SPVC_MSL_SAMPLER_ADDRESS_INT_MAX = 0x7fffffff
} spvc_msl_sampler_address;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_compare_func
{
SPVC_MSL_SAMPLER_COMPARE_FUNC_NEVER = 0,
SPVC_MSL_SAMPLER_COMPARE_FUNC_LESS = 1,
SPVC_MSL_SAMPLER_COMPARE_FUNC_LESS_EQUAL = 2,
SPVC_MSL_SAMPLER_COMPARE_FUNC_GREATER = 3,
SPVC_MSL_SAMPLER_COMPARE_FUNC_GREATER_EQUAL = 4,
SPVC_MSL_SAMPLER_COMPARE_FUNC_EQUAL = 5,
SPVC_MSL_SAMPLER_COMPARE_FUNC_NOT_EQUAL = 6,
SPVC_MSL_SAMPLER_COMPARE_FUNC_ALWAYS = 7,
SPVC_MSL_SAMPLER_COMPARE_FUNC_INT_MAX = 0x7fffffff
} spvc_msl_sampler_compare_func;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_border_color
{
SPVC_MSL_SAMPLER_BORDER_COLOR_TRANSPARENT_BLACK = 0,
SPVC_MSL_SAMPLER_BORDER_COLOR_OPAQUE_BLACK = 1,
SPVC_MSL_SAMPLER_BORDER_COLOR_OPAQUE_WHITE = 2,
SPVC_MSL_SAMPLER_BORDER_COLOR_INT_MAX = 0x7fffffff
} spvc_msl_sampler_border_color;
/* Maps to C++ API. */
typedef enum spvc_msl_format_resolution
{
SPVC_MSL_FORMAT_RESOLUTION_444 = 0,
SPVC_MSL_FORMAT_RESOLUTION_422,
SPVC_MSL_FORMAT_RESOLUTION_420,
SPVC_MSL_FORMAT_RESOLUTION_INT_MAX = 0x7fffffff
} spvc_msl_format_resolution;
/* Maps to C++ API. */
typedef enum spvc_msl_chroma_location
{
SPVC_MSL_CHROMA_LOCATION_COSITED_EVEN = 0,
SPVC_MSL_CHROMA_LOCATION_MIDPOINT,
SPVC_MSL_CHROMA_LOCATION_INT_MAX = 0x7fffffff
} spvc_msl_chroma_location;
/* Maps to C++ API. */
typedef enum spvc_msl_component_swizzle
{
SPVC_MSL_COMPONENT_SWIZZLE_IDENTITY = 0,
SPVC_MSL_COMPONENT_SWIZZLE_ZERO,
SPVC_MSL_COMPONENT_SWIZZLE_ONE,
SPVC_MSL_COMPONENT_SWIZZLE_R,
SPVC_MSL_COMPONENT_SWIZZLE_G,
SPVC_MSL_COMPONENT_SWIZZLE_B,
SPVC_MSL_COMPONENT_SWIZZLE_A,
SPVC_MSL_COMPONENT_SWIZZLE_INT_MAX = 0x7fffffff
} spvc_msl_component_swizzle;
/* Maps to C++ API. */
typedef enum spvc_msl_sampler_ycbcr_model_conversion
{
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_RGB_IDENTITY = 0,
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_IDENTITY,
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_709,
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_601,
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_2020,
SPVC_MSL_SAMPLER_YCBCR_MODEL_CONVERSION_INT_MAX = 0x7fffffff
} spvc_msl_sampler_ycbcr_model_conversion;
/* Maps to C+ API. */
typedef enum spvc_msl_sampler_ycbcr_range
{
SPVC_MSL_SAMPLER_YCBCR_RANGE_ITU_FULL = 0,
SPVC_MSL_SAMPLER_YCBCR_RANGE_ITU_NARROW,
SPVC_MSL_SAMPLER_YCBCR_RANGE_INT_MAX = 0x7fffffff
} spvc_msl_sampler_ycbcr_range;
/* Maps to C++ API. */
typedef struct spvc_msl_constexpr_sampler
{
spvc_msl_sampler_coord coord;
spvc_msl_sampler_filter min_filter;
spvc_msl_sampler_filter mag_filter;
spvc_msl_sampler_mip_filter mip_filter;
spvc_msl_sampler_address s_address;
spvc_msl_sampler_address t_address;
spvc_msl_sampler_address r_address;
spvc_msl_sampler_compare_func compare_func;
spvc_msl_sampler_border_color border_color;
float lod_clamp_min;
float lod_clamp_max;
int max_anisotropy;
spvc_bool compare_enable;
spvc_bool lod_clamp_enable;
spvc_bool anisotropy_enable;
} spvc_msl_constexpr_sampler;
/*
* Initializes the constexpr sampler struct.
* The defaults are non-zero.
*/
SPVC_PUBLIC_API void spvc_msl_constexpr_sampler_init(spvc_msl_constexpr_sampler *sampler);
/* Maps to the sampler Y'CbCr conversion-related portions of MSLConstexprSampler. See C++ API for defaults and details. */
typedef struct spvc_msl_sampler_ycbcr_conversion
{
unsigned planes;
spvc_msl_format_resolution resolution;
spvc_msl_sampler_filter chroma_filter;
spvc_msl_chroma_location x_chroma_offset;
spvc_msl_chroma_location y_chroma_offset;
spvc_msl_component_swizzle swizzle[4];
spvc_msl_sampler_ycbcr_model_conversion ycbcr_model;
spvc_msl_sampler_ycbcr_range ycbcr_range;
unsigned bpc;
} spvc_msl_sampler_ycbcr_conversion;
/*
* Initializes the constexpr sampler struct.
* The defaults are non-zero.
*/
SPVC_PUBLIC_API void spvc_msl_sampler_ycbcr_conversion_init(spvc_msl_sampler_ycbcr_conversion *conv);
/* Maps to C++ API. */
typedef enum spvc_hlsl_binding_flag_bits
{
SPVC_HLSL_BINDING_AUTO_PUSH_CONSTANT_BIT = 1 << 0,
SPVC_HLSL_BINDING_AUTO_CBV_BIT = 1 << 1,
SPVC_HLSL_BINDING_AUTO_SRV_BIT = 1 << 2,
SPVC_HLSL_BINDING_AUTO_UAV_BIT = 1 << 3,
SPVC_HLSL_BINDING_AUTO_SAMPLER_BIT = 1 << 4,
SPVC_HLSL_BINDING_AUTO_ALL = 0x7fffffff
} spvc_hlsl_binding_flag_bits;
typedef unsigned spvc_hlsl_binding_flags;
/* Maps to the various spirv_cross::Compiler*::Option structures. See C++ API for defaults and details. */
typedef enum spvc_compiler_option
{
SPVC_COMPILER_OPTION_UNKNOWN = 0,
SPVC_COMPILER_OPTION_FORCE_TEMPORARY = 1 | SPVC_COMPILER_OPTION_COMMON_BIT,
SPVC_COMPILER_OPTION_FLATTEN_MULTIDIMENSIONAL_ARRAYS = 2 | SPVC_COMPILER_OPTION_COMMON_BIT,
SPVC_COMPILER_OPTION_FIXUP_DEPTH_CONVENTION = 3 | SPVC_COMPILER_OPTION_COMMON_BIT,
SPVC_COMPILER_OPTION_FLIP_VERTEX_Y = 4 | SPVC_COMPILER_OPTION_COMMON_BIT,
SPVC_COMPILER_OPTION_GLSL_SUPPORT_NONZERO_BASE_INSTANCE = 5 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_SEPARATE_SHADER_OBJECTS = 6 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_ENABLE_420PACK_EXTENSION = 7 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_VERSION = 8 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_ES = 9 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_VULKAN_SEMANTICS = 10 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_ES_DEFAULT_FLOAT_PRECISION_HIGHP = 11 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_GLSL_ES_DEFAULT_INT_PRECISION_HIGHP = 12 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_HLSL_SHADER_MODEL = 13 | SPVC_COMPILER_OPTION_HLSL_BIT,
SPVC_COMPILER_OPTION_HLSL_POINT_SIZE_COMPAT = 14 | SPVC_COMPILER_OPTION_HLSL_BIT,
SPVC_COMPILER_OPTION_HLSL_POINT_COORD_COMPAT = 15 | SPVC_COMPILER_OPTION_HLSL_BIT,
SPVC_COMPILER_OPTION_HLSL_SUPPORT_NONZERO_BASE_VERTEX_BASE_INSTANCE = 16 | SPVC_COMPILER_OPTION_HLSL_BIT,
SPVC_COMPILER_OPTION_MSL_VERSION = 17 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_TEXEL_BUFFER_TEXTURE_WIDTH = 18 | SPVC_COMPILER_OPTION_MSL_BIT,
/* Obsolete, use SWIZZLE_BUFFER_INDEX instead. */
SPVC_COMPILER_OPTION_MSL_AUX_BUFFER_INDEX = 19 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SWIZZLE_BUFFER_INDEX = 19 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_INDIRECT_PARAMS_BUFFER_INDEX = 20 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SHADER_OUTPUT_BUFFER_INDEX = 21 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SHADER_PATCH_OUTPUT_BUFFER_INDEX = 22 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SHADER_TESS_FACTOR_OUTPUT_BUFFER_INDEX = 23 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SHADER_INPUT_WORKGROUP_INDEX = 24 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_ENABLE_POINT_SIZE_BUILTIN = 25 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_DISABLE_RASTERIZATION = 26 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_CAPTURE_OUTPUT_TO_BUFFER = 27 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_SWIZZLE_TEXTURE_SAMPLES = 28 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_PAD_FRAGMENT_OUTPUT_COMPONENTS = 29 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_TESS_DOMAIN_ORIGIN_LOWER_LEFT = 30 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_PLATFORM = 31 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_ARGUMENT_BUFFERS = 32 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_GLSL_EMIT_PUSH_CONSTANT_AS_UNIFORM_BUFFER = 33 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_MSL_TEXTURE_BUFFER_NATIVE = 34 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_GLSL_EMIT_UNIFORM_BUFFER_AS_PLAIN_UNIFORMS = 35 | SPVC_COMPILER_OPTION_GLSL_BIT,
SPVC_COMPILER_OPTION_MSL_BUFFER_SIZE_BUFFER_INDEX = 36 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_EMIT_LINE_DIRECTIVES = 37 | SPVC_COMPILER_OPTION_COMMON_BIT,
SPVC_COMPILER_OPTION_MSL_MULTIVIEW = 38 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_VIEW_MASK_BUFFER_INDEX = 39 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_DEVICE_INDEX = 40 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_VIEW_INDEX_FROM_DEVICE_INDEX = 41 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_DISPATCH_BASE = 42 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_DYNAMIC_OFFSETS_BUFFER_INDEX = 43 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_TEXTURE_1D_AS_2D = 44 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_ENABLE_BASE_INDEX_ZERO = 45 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_IOS_FRAMEBUFFER_FETCH_SUBPASS = 46 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_INVARIANT_FP_MATH = 47 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_EMULATE_CUBEMAP_ARRAY = 48 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_MSL_ENABLE_DECORATION_BINDING = 49 | SPVC_COMPILER_OPTION_MSL_BIT,
SPVC_COMPILER_OPTION_INT_MAX = 0x7fffffff
} spvc_compiler_option;
/*
* Context is the highest-level API construct.
* The context owns all memory allocations made by its child object hierarchy, including various non-opaque structs and strings.
* This means that the API user only has to care about one "destroy" call ever when using the C API.
* All pointers handed out by the APIs are only valid as long as the context
* is alive and spvc_context_release_allocations has not been called.
*/
SPVC_PUBLIC_API spvc_result spvc_context_create(spvc_context *context);
/* Frees all memory allocations and objects associated with the context and its child objects. */
SPVC_PUBLIC_API void spvc_context_destroy(spvc_context context);
/* Frees all memory allocations and objects associated with the context and its child objects, but keeps the context alive. */
SPVC_PUBLIC_API void spvc_context_release_allocations(spvc_context context);
/* Get the string for the last error which was logged. */
SPVC_PUBLIC_API const char *spvc_context_get_last_error_string(spvc_context context);
/* Get notified in a callback when an error triggers. Useful for debugging. */
typedef void (*spvc_error_callback)(void *userdata, const char *error);
SPVC_PUBLIC_API void spvc_context_set_error_callback(spvc_context context, spvc_error_callback cb, void *userdata);
/* SPIR-V parsing interface. Maps to Parser which then creates a ParsedIR, and that IR is extracted into the handle. */
SPVC_PUBLIC_API spvc_result spvc_context_parse_spirv(spvc_context context, const SpvId *spirv, size_t word_count,
spvc_parsed_ir *parsed_ir);
/*
* Create a compiler backend. Capture mode controls if we construct by copy or move semantics.
* It is always recommended to use SPVC_CAPTURE_MODE_TAKE_OWNERSHIP if you only intend to cross-compile the IR once.
*/
SPVC_PUBLIC_API spvc_result spvc_context_create_compiler(spvc_context context, spvc_backend backend,
spvc_parsed_ir parsed_ir, spvc_capture_mode mode,
spvc_compiler *compiler);
/* Maps directly to C++ API. */
SPVC_PUBLIC_API unsigned spvc_compiler_get_current_id_bound(spvc_compiler compiler);
/* Create compiler options, which will initialize defaults. */
SPVC_PUBLIC_API spvc_result spvc_compiler_create_compiler_options(spvc_compiler compiler,
spvc_compiler_options *options);
/* Override options. Will return error if e.g. MSL options are used for the HLSL backend, etc. */
SPVC_PUBLIC_API spvc_result spvc_compiler_options_set_bool(spvc_compiler_options options,
spvc_compiler_option option, spvc_bool value);
SPVC_PUBLIC_API spvc_result spvc_compiler_options_set_uint(spvc_compiler_options options,
spvc_compiler_option option, unsigned value);
/* Set compiler options. */
SPVC_PUBLIC_API spvc_result spvc_compiler_install_compiler_options(spvc_compiler compiler,
spvc_compiler_options options);
/* Compile IR into a string. *source is owned by the context, and caller must not free it themselves. */
SPVC_PUBLIC_API spvc_result spvc_compiler_compile(spvc_compiler compiler, const char **source);
/* Maps to C++ API. */
SPVC_PUBLIC_API spvc_result spvc_compiler_add_header_line(spvc_compiler compiler, const char *line);
SPVC_PUBLIC_API spvc_result spvc_compiler_require_extension(spvc_compiler compiler, const char *ext);
SPVC_PUBLIC_API spvc_result spvc_compiler_flatten_buffer_block(spvc_compiler compiler, spvc_variable_id id);
/*
* HLSL specifics.
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_hlsl_set_root_constants_layout(spvc_compiler compiler,
const spvc_hlsl_root_constants *constant_info,
size_t count);
SPVC_PUBLIC_API spvc_result spvc_compiler_hlsl_add_vertex_attribute_remap(spvc_compiler compiler,
const spvc_hlsl_vertex_attribute_remap *remap,
size_t remaps);
SPVC_PUBLIC_API spvc_variable_id spvc_compiler_hlsl_remap_num_workgroups_builtin(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_result spvc_compiler_hlsl_set_resource_binding_flags(spvc_compiler compiler,
spvc_hlsl_binding_flags flags);
/*
* MSL specifics.
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_is_rasterization_disabled(spvc_compiler compiler);
/* Obsolete. Renamed to needs_swizzle_buffer. */
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_aux_buffer(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_swizzle_buffer(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_buffer_size_buffer(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_output_buffer(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_patch_output_buffer(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_needs_input_threadgroup_mem(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_vertex_attribute(spvc_compiler compiler,
const spvc_msl_vertex_attribute *attrs);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_resource_binding(spvc_compiler compiler,
const spvc_msl_resource_binding *binding);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_discrete_descriptor_set(spvc_compiler compiler, unsigned desc_set);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_set_argument_buffer_device_address_space(spvc_compiler compiler, unsigned desc_set, spvc_bool device_address);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_is_vertex_attribute_used(spvc_compiler compiler, unsigned location);
SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_is_resource_used(spvc_compiler compiler,
SpvExecutionModel model,
unsigned set,
unsigned binding);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_remap_constexpr_sampler(spvc_compiler compiler, spvc_variable_id id, const spvc_msl_constexpr_sampler *sampler);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_remap_constexpr_sampler_by_binding(spvc_compiler compiler, unsigned desc_set, unsigned binding, const spvc_msl_constexpr_sampler *sampler);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_remap_constexpr_sampler_ycbcr(spvc_compiler compiler, spvc_variable_id id, const spvc_msl_constexpr_sampler *sampler, const spvc_msl_sampler_ycbcr_conversion *conv);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_remap_constexpr_sampler_by_binding_ycbcr(spvc_compiler compiler, unsigned desc_set, unsigned binding, const spvc_msl_constexpr_sampler *sampler, const spvc_msl_sampler_ycbcr_conversion *conv);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_set_fragment_output_components(spvc_compiler compiler, unsigned location, unsigned components);
SPVC_PUBLIC_API unsigned spvc_compiler_msl_get_automatic_resource_binding(spvc_compiler compiler, spvc_variable_id id);
SPVC_PUBLIC_API unsigned spvc_compiler_msl_get_automatic_resource_binding_secondary(spvc_compiler compiler, spvc_variable_id id);
SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_dynamic_buffer(spvc_compiler compiler, unsigned desc_set, unsigned binding, unsigned index);
/*
* Reflect resources.
* Maps almost 1:1 to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_get_active_interface_variables(spvc_compiler compiler, spvc_set *set);
SPVC_PUBLIC_API spvc_result spvc_compiler_set_enabled_interface_variables(spvc_compiler compiler, spvc_set set);
SPVC_PUBLIC_API spvc_result spvc_compiler_create_shader_resources(spvc_compiler compiler, spvc_resources *resources);
SPVC_PUBLIC_API spvc_result spvc_compiler_create_shader_resources_for_active_variables(spvc_compiler compiler,
spvc_resources *resources,
spvc_set active);
SPVC_PUBLIC_API spvc_result spvc_resources_get_resource_list_for_type(spvc_resources resources, spvc_resource_type type,
const spvc_reflected_resource **resource_list,
size_t *resource_size);
/*
* Decorations.
* Maps to C++ API.
*/
SPVC_PUBLIC_API void spvc_compiler_set_decoration(spvc_compiler compiler, SpvId id, SpvDecoration decoration,
unsigned argument);
SPVC_PUBLIC_API void spvc_compiler_set_decoration_string(spvc_compiler compiler, SpvId id, SpvDecoration decoration,
const char *argument);
SPVC_PUBLIC_API void spvc_compiler_set_name(spvc_compiler compiler, SpvId id, const char *argument);
SPVC_PUBLIC_API void spvc_compiler_set_member_decoration(spvc_compiler compiler, spvc_type_id id, unsigned member_index,
SpvDecoration decoration, unsigned argument);
SPVC_PUBLIC_API void spvc_compiler_set_member_decoration_string(spvc_compiler compiler, spvc_type_id id,
unsigned member_index, SpvDecoration decoration,
const char *argument);
SPVC_PUBLIC_API void spvc_compiler_set_member_name(spvc_compiler compiler, spvc_type_id id, unsigned member_index,
const char *argument);
SPVC_PUBLIC_API void spvc_compiler_unset_decoration(spvc_compiler compiler, SpvId id, SpvDecoration decoration);
SPVC_PUBLIC_API void spvc_compiler_unset_member_decoration(spvc_compiler compiler, spvc_type_id id,
unsigned member_index, SpvDecoration decoration);
SPVC_PUBLIC_API spvc_bool spvc_compiler_has_decoration(spvc_compiler compiler, SpvId id, SpvDecoration decoration);
SPVC_PUBLIC_API spvc_bool spvc_compiler_has_member_decoration(spvc_compiler compiler, spvc_type_id id,
unsigned member_index, SpvDecoration decoration);
SPVC_PUBLIC_API const char *spvc_compiler_get_name(spvc_compiler compiler, SpvId id);
SPVC_PUBLIC_API unsigned spvc_compiler_get_decoration(spvc_compiler compiler, SpvId id, SpvDecoration decoration);
SPVC_PUBLIC_API const char *spvc_compiler_get_decoration_string(spvc_compiler compiler, SpvId id,
SpvDecoration decoration);
SPVC_PUBLIC_API unsigned spvc_compiler_get_member_decoration(spvc_compiler compiler, spvc_type_id id,
unsigned member_index, SpvDecoration decoration);
SPVC_PUBLIC_API const char *spvc_compiler_get_member_decoration_string(spvc_compiler compiler, spvc_type_id id,
unsigned member_index, SpvDecoration decoration);
SPVC_PUBLIC_API const char *spvc_compiler_get_member_name(spvc_compiler compiler, spvc_type_id id, unsigned member_index);
/*
* Entry points.
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_get_entry_points(spvc_compiler compiler,
const spvc_entry_point **entry_points,
size_t *num_entry_points);
SPVC_PUBLIC_API spvc_result spvc_compiler_set_entry_point(spvc_compiler compiler, const char *name,
SpvExecutionModel model);
SPVC_PUBLIC_API spvc_result spvc_compiler_rename_entry_point(spvc_compiler compiler, const char *old_name,
const char *new_name, SpvExecutionModel model);
SPVC_PUBLIC_API const char *spvc_compiler_get_cleansed_entry_point_name(spvc_compiler compiler, const char *name,
SpvExecutionModel model);
SPVC_PUBLIC_API void spvc_compiler_set_execution_mode(spvc_compiler compiler, SpvExecutionMode mode);
SPVC_PUBLIC_API void spvc_compiler_unset_execution_mode(spvc_compiler compiler, SpvExecutionMode mode);
SPVC_PUBLIC_API void spvc_compiler_set_execution_mode_with_arguments(spvc_compiler compiler, SpvExecutionMode mode,
unsigned arg0, unsigned arg1, unsigned arg2);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_execution_modes(spvc_compiler compiler, const SpvExecutionMode **modes,
size_t *num_modes);
SPVC_PUBLIC_API unsigned spvc_compiler_get_execution_mode_argument(spvc_compiler compiler, SpvExecutionMode mode);
SPVC_PUBLIC_API unsigned spvc_compiler_get_execution_mode_argument_by_index(spvc_compiler compiler,
SpvExecutionMode mode, unsigned index);
SPVC_PUBLIC_API SpvExecutionModel spvc_compiler_get_execution_model(spvc_compiler compiler);
/*
* Type query interface.
* Maps to C++ API, except it's read-only.
*/
SPVC_PUBLIC_API spvc_type spvc_compiler_get_type_handle(spvc_compiler compiler, spvc_type_id id);
/* Pulls out SPIRType::self. This effectively gives the type ID without array or pointer qualifiers.
* This is necessary when reflecting decoration/name information on members of a struct,
* which are placed in the base type, not the qualified type.
* This is similar to spvc_reflected_resource::base_type_id. */
SPVC_PUBLIC_API spvc_type_id spvc_type_get_base_type_id(spvc_type type);
SPVC_PUBLIC_API spvc_basetype spvc_type_get_basetype(spvc_type type);
SPVC_PUBLIC_API unsigned spvc_type_get_bit_width(spvc_type type);
SPVC_PUBLIC_API unsigned spvc_type_get_vector_size(spvc_type type);
SPVC_PUBLIC_API unsigned spvc_type_get_columns(spvc_type type);
SPVC_PUBLIC_API unsigned spvc_type_get_num_array_dimensions(spvc_type type);
SPVC_PUBLIC_API spvc_bool spvc_type_array_dimension_is_literal(spvc_type type, unsigned dimension);
SPVC_PUBLIC_API SpvId spvc_type_get_array_dimension(spvc_type type, unsigned dimension);
SPVC_PUBLIC_API unsigned spvc_type_get_num_member_types(spvc_type type);
SPVC_PUBLIC_API spvc_type_id spvc_type_get_member_type(spvc_type type, unsigned index);
SPVC_PUBLIC_API SpvStorageClass spvc_type_get_storage_class(spvc_type type);
/* Image type query. */
SPVC_PUBLIC_API spvc_type_id spvc_type_get_image_sampled_type(spvc_type type);
SPVC_PUBLIC_API SpvDim spvc_type_get_image_dimension(spvc_type type);
SPVC_PUBLIC_API spvc_bool spvc_type_get_image_is_depth(spvc_type type);
SPVC_PUBLIC_API spvc_bool spvc_type_get_image_arrayed(spvc_type type);
SPVC_PUBLIC_API spvc_bool spvc_type_get_image_multisampled(spvc_type type);
SPVC_PUBLIC_API spvc_bool spvc_type_get_image_is_storage(spvc_type type);
SPVC_PUBLIC_API SpvImageFormat spvc_type_get_image_storage_format(spvc_type type);
SPVC_PUBLIC_API SpvAccessQualifier spvc_type_get_image_access_qualifier(spvc_type type);
/*
* Buffer layout query.
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_get_declared_struct_size(spvc_compiler compiler, spvc_type struct_type, size_t *size);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_declared_struct_size_runtime_array(spvc_compiler compiler,
spvc_type struct_type, size_t array_size, size_t *size);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_declared_struct_member_size(spvc_compiler compiler, spvc_type type, unsigned index, size_t *size);
SPVC_PUBLIC_API spvc_result spvc_compiler_type_struct_member_offset(spvc_compiler compiler,
spvc_type type, unsigned index, unsigned *offset);
SPVC_PUBLIC_API spvc_result spvc_compiler_type_struct_member_array_stride(spvc_compiler compiler,
spvc_type type, unsigned index, unsigned *stride);
SPVC_PUBLIC_API spvc_result spvc_compiler_type_struct_member_matrix_stride(spvc_compiler compiler,
spvc_type type, unsigned index, unsigned *stride);
/*
* Workaround helper functions.
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_build_dummy_sampler_for_combined_images(spvc_compiler compiler, spvc_variable_id *id);
SPVC_PUBLIC_API spvc_result spvc_compiler_build_combined_image_samplers(spvc_compiler compiler);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_combined_image_samplers(spvc_compiler compiler,
const spvc_combined_image_sampler **samplers,
size_t *num_samplers);
/*
* Constants
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_get_specialization_constants(spvc_compiler compiler,
const spvc_specialization_constant **constants,
size_t *num_constants);
SPVC_PUBLIC_API spvc_constant spvc_compiler_get_constant_handle(spvc_compiler compiler,
spvc_constant_id id);
SPVC_PUBLIC_API spvc_constant_id spvc_compiler_get_work_group_size_specialization_constants(spvc_compiler compiler,
spvc_specialization_constant *x,
spvc_specialization_constant *y,
spvc_specialization_constant *z);
/*
* Buffer ranges
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_result spvc_compiler_get_active_buffer_ranges(spvc_compiler compiler,
spvc_variable_id id,
const spvc_buffer_range **ranges,
size_t *num_ranges);
/*
* No stdint.h until C99, sigh :(
* For smaller types, the result is sign or zero-extended as appropriate.
* Maps to C++ API.
* TODO: The SPIRConstant query interface and modification interface is not quite complete.
*/
SPVC_PUBLIC_API float spvc_constant_get_scalar_fp16(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API float spvc_constant_get_scalar_fp32(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API double spvc_constant_get_scalar_fp64(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API unsigned spvc_constant_get_scalar_u32(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API int spvc_constant_get_scalar_i32(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API unsigned spvc_constant_get_scalar_u16(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API int spvc_constant_get_scalar_i16(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API unsigned spvc_constant_get_scalar_u8(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API int spvc_constant_get_scalar_i8(spvc_constant constant, unsigned column, unsigned row);
SPVC_PUBLIC_API void spvc_constant_get_subconstants(spvc_constant constant, const spvc_constant_id **constituents, size_t *count);
SPVC_PUBLIC_API spvc_type_id spvc_constant_get_type(spvc_constant constant);
/*
* Misc reflection
* Maps to C++ API.
*/
SPVC_PUBLIC_API spvc_bool spvc_compiler_get_binary_offset_for_decoration(spvc_compiler compiler,
spvc_variable_id id,
SpvDecoration decoration,
unsigned *word_offset);
SPVC_PUBLIC_API spvc_bool spvc_compiler_buffer_is_hlsl_counter_buffer(spvc_compiler compiler, spvc_variable_id id);
SPVC_PUBLIC_API spvc_bool spvc_compiler_buffer_get_hlsl_counter_buffer(spvc_compiler compiler, spvc_variable_id id,
spvc_variable_id *counter_id);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_declared_capabilities(spvc_compiler compiler,
const SpvCapability **capabilities,
size_t *num_capabilities);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_declared_extensions(spvc_compiler compiler, const char ***extensions,
size_t *num_extensions);
SPVC_PUBLIC_API const char *spvc_compiler_get_remapped_declared_block_name(spvc_compiler compiler, spvc_variable_id id);
SPVC_PUBLIC_API spvc_result spvc_compiler_get_buffer_block_decorations(spvc_compiler compiler, spvc_variable_id id,
const SpvDecoration **decorations,
size_t *num_decorations);
#ifdef __cplusplus
}
#endif
#endif
@@ -0,0 +1,720 @@
/*
* Copyright 2019 Hans-Kristian Arntzen
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_CONTAINERS_HPP
#define SPIRV_CROSS_CONTAINERS_HPP
#include "spirv_cross_error_handling.hpp"
#include <algorithm>
#include <functional>
#include <iterator>
#include <memory>
#include <stack>
#include <stdint.h>
#include <stdlib.h>
#include <string.h>
#include <type_traits>
#include <unordered_map>
#include <unordered_set>
#include <utility>
#include <vector>
#ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
#define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
#else
#define SPIRV_CROSS_NAMESPACE spirv_cross
#endif
namespace SPIRV_CROSS_NAMESPACE
{
#ifndef SPIRV_CROSS_FORCE_STL_TYPES
// std::aligned_storage does not support size == 0, so roll our own.
template <typename T, size_t N>
class AlignedBuffer
{
public:
T *data()
{
#if defined(_MSC_VER) && _MSC_VER < 1900
// MSVC 2013 workarounds, sigh ...
// Only use this workaround on MSVC 2013 due to some confusion around default initialized unions.
// Spec seems to suggest the memory will be zero-initialized, which is *not* what we want.
return reinterpret_cast<T *>(u.aligned_char);
#else
return reinterpret_cast<T *>(aligned_char);
#endif
}
private:
#if defined(_MSC_VER) && _MSC_VER < 1900
// MSVC 2013 workarounds, sigh ...
union {
char aligned_char[sizeof(T) * N];
double dummy_aligner;
} u;
#else
alignas(T) char aligned_char[sizeof(T) * N];
#endif
};
template <typename T>
class AlignedBuffer<T, 0>
{
public:
T *data()
{
return nullptr;
}
};
// An immutable version of SmallVector which erases type information about storage.
template <typename T>
class VectorView
{
public:
T &operator[](size_t i)
{
return ptr[i];
}
const T &operator[](size_t i) const
{
return ptr[i];
}
bool empty() const
{
return buffer_size == 0;
}
size_t size() const
{
return buffer_size;
}
T *data()
{
return ptr;
}
const T *data() const
{
return ptr;
}
T *begin()
{
return ptr;
}
T *end()
{
return ptr + buffer_size;
}
const T *begin() const
{
return ptr;
}
const T *end() const
{
return ptr + buffer_size;
}
T &front()
{
return ptr[0];
}
const T &front() const
{
return ptr[0];
}
T &back()
{
return ptr[buffer_size - 1];
}
const T &back() const
{
return ptr[buffer_size - 1];
}
// Makes it easier to consume SmallVector.
#if defined(_MSC_VER) && _MSC_VER < 1900
explicit operator std::vector<T>() const
{
// Another MSVC 2013 workaround. It does not understand lvalue/rvalue qualified operations.
return std::vector<T>(ptr, ptr + buffer_size);
}
#else
// Makes it easier to consume SmallVector.
explicit operator std::vector<T>() const &
{
return std::vector<T>(ptr, ptr + buffer_size);
}
// If we are converting as an r-value, we can pilfer our elements.
explicit operator std::vector<T>() &&
{
return std::vector<T>(std::make_move_iterator(ptr), std::make_move_iterator(ptr + buffer_size));
}
#endif
// Avoid sliced copies. Base class should only be read as a reference.
VectorView(const VectorView &) = delete;
void operator=(const VectorView &) = delete;
protected:
VectorView() = default;
T *ptr = nullptr;
size_t buffer_size = 0;
};
// Simple vector which supports up to N elements inline, without malloc/free.
// We use a lot of throwaway vectors all over the place which triggers allocations.
// This class only implements the subset of std::vector we need in SPIRV-Cross.
// It is *NOT* a drop-in replacement in general projects.
template <typename T, size_t N = 8>
class SmallVector : public VectorView<T>
{
public:
SmallVector()
{
this->ptr = stack_storage.data();
buffer_capacity = N;
}
SmallVector(const T *arg_list_begin, const T *arg_list_end)
: SmallVector()
{
auto count = size_t(arg_list_end - arg_list_begin);
reserve(count);
for (size_t i = 0; i < count; i++, arg_list_begin++)
new (&this->ptr[i]) T(*arg_list_begin);
this->buffer_size = count;
}
SmallVector(SmallVector &&other) SPIRV_CROSS_NOEXCEPT : SmallVector()
{
*this = std::move(other);
}
SmallVector &operator=(SmallVector &&other) SPIRV_CROSS_NOEXCEPT
{
clear();
if (other.ptr != other.stack_storage.data())
{
// Pilfer allocated pointer.
if (this->ptr != stack_storage.data())
free(this->ptr);
this->ptr = other.ptr;
this->buffer_size = other.buffer_size;
buffer_capacity = other.buffer_capacity;
other.ptr = nullptr;
other.buffer_size = 0;
other.buffer_capacity = 0;
}
else
{
// Need to move the stack contents individually.
reserve(other.buffer_size);
for (size_t i = 0; i < other.buffer_size; i++)
{
new (&this->ptr[i]) T(std::move(other.ptr[i]));
other.ptr[i].~T();
}
this->buffer_size = other.buffer_size;
other.buffer_size = 0;
}
return *this;
}
SmallVector(const SmallVector &other)
: SmallVector()
{
*this = other;
}
SmallVector &operator=(const SmallVector &other)
{
clear();
reserve(other.buffer_size);
for (size_t i = 0; i < other.buffer_size; i++)
new (&this->ptr[i]) T(other.ptr[i]);
this->buffer_size = other.buffer_size;
return *this;
}
explicit SmallVector(size_t count)
: SmallVector()
{
resize(count);
}
~SmallVector()
{
clear();
if (this->ptr != stack_storage.data())
free(this->ptr);
}
void clear()
{
for (size_t i = 0; i < this->buffer_size; i++)
this->ptr[i].~T();
this->buffer_size = 0;
}
void push_back(const T &t)
{
reserve(this->buffer_size + 1);
new (&this->ptr[this->buffer_size]) T(t);
this->buffer_size++;
}
void push_back(T &&t)
{
reserve(this->buffer_size + 1);
new (&this->ptr[this->buffer_size]) T(std::move(t));
this->buffer_size++;
}
void pop_back()
{
// Work around false positive warning on GCC 8.3.
// Calling pop_back on empty vector is undefined.
if (!this->empty())
resize(this->buffer_size - 1);
}
template <typename... Ts>
void emplace_back(Ts &&... ts)
{
reserve(this->buffer_size + 1);
new (&this->ptr[this->buffer_size]) T(std::forward<Ts>(ts)...);
this->buffer_size++;
}
void reserve(size_t count)
{
if (count > buffer_capacity)
{
size_t target_capacity = buffer_capacity;
if (target_capacity == 0)
target_capacity = 1;
if (target_capacity < N)
target_capacity = N;
while (target_capacity < count)
target_capacity <<= 1u;
T *new_buffer =
target_capacity > N ? static_cast<T *>(malloc(target_capacity * sizeof(T))) : stack_storage.data();
if (!new_buffer)
SPIRV_CROSS_THROW("Out of memory.");
// In case for some reason two allocations both come from same stack.
if (new_buffer != this->ptr)
{
// We don't deal with types which can throw in move constructor.
for (size_t i = 0; i < this->buffer_size; i++)
{
new (&new_buffer[i]) T(std::move(this->ptr[i]));
this->ptr[i].~T();
}
}
if (this->ptr != stack_storage.data())
free(this->ptr);
this->ptr = new_buffer;
buffer_capacity = target_capacity;
}
}
void insert(T *itr, const T *insert_begin, const T *insert_end)
{
auto count = size_t(insert_end - insert_begin);
if (itr == this->end())
{
reserve(this->buffer_size + count);
for (size_t i = 0; i < count; i++, insert_begin++)
new (&this->ptr[this->buffer_size + i]) T(*insert_begin);
this->buffer_size += count;
}
else
{
if (this->buffer_size + count > buffer_capacity)
{
auto target_capacity = this->buffer_size + count;
if (target_capacity == 0)
target_capacity = 1;
if (target_capacity < N)
target_capacity = N;
while (target_capacity < count)
target_capacity <<= 1u;
// Need to allocate new buffer. Move everything to a new buffer.
T *new_buffer =
target_capacity > N ? static_cast<T *>(malloc(target_capacity * sizeof(T))) : stack_storage.data();
if (!new_buffer)
SPIRV_CROSS_THROW("Out of memory.");
// First, move elements from source buffer to new buffer.
// We don't deal with types which can throw in move constructor.
auto *target_itr = new_buffer;
auto *original_source_itr = this->begin();
if (new_buffer != this->ptr)
{
while (original_source_itr != itr)
{
new (target_itr) T(std::move(*original_source_itr));
original_source_itr->~T();
++original_source_itr;
++target_itr;
}
}
// Copy-construct new elements.
for (auto *source_itr = insert_begin; source_itr != insert_end; ++source_itr, ++target_itr)
new (target_itr) T(*source_itr);
// Move over the other half.
if (new_buffer != this->ptr || insert_begin != insert_end)
{
while (original_source_itr != this->end())
{
new (target_itr) T(std::move(*original_source_itr));
original_source_itr->~T();
++original_source_itr;
++target_itr;
}
}
if (this->ptr != stack_storage.data())
free(this->ptr);
this->ptr = new_buffer;
buffer_capacity = target_capacity;
}
else
{
// Move in place, need to be a bit careful about which elements are constructed and which are not.
// Move the end and construct the new elements.
auto *target_itr = this->end() + count;
auto *source_itr = this->end();
while (target_itr != this->end() && source_itr != itr)
{
--target_itr;
--source_itr;
new (target_itr) T(std::move(*source_itr));
}
// For already constructed elements we can move-assign.
std::move_backward(itr, source_itr, target_itr);
// For the inserts which go to already constructed elements, we can do a plain copy.
while (itr != this->end() && insert_begin != insert_end)
*itr++ = *insert_begin++;
// For inserts into newly allocated memory, we must copy-construct instead.
while (insert_begin != insert_end)
{
new (itr) T(*insert_begin);
++itr;
++insert_begin;
}
}
this->buffer_size += count;
}
}
void insert(T *itr, const T &value)
{
insert(itr, &value, &value + 1);
}
T *erase(T *itr)
{
std::move(itr + 1, this->end(), itr);
this->ptr[--this->buffer_size].~T();
return itr;
}
void erase(T *start_erase, T *end_erase)
{
if (end_erase == this->end())
{
resize(size_t(start_erase - this->begin()));
}
else
{
auto new_size = this->buffer_size - (end_erase - start_erase);
std::move(end_erase, this->end(), start_erase);
resize(new_size);
}
}
void resize(size_t new_size)
{
if (new_size < this->buffer_size)
{
for (size_t i = new_size; i < this->buffer_size; i++)
this->ptr[i].~T();
}
else if (new_size > this->buffer_size)
{
reserve(new_size);
for (size_t i = this->buffer_size; i < new_size; i++)
new (&this->ptr[i]) T();
}
this->buffer_size = new_size;
}
private:
size_t buffer_capacity = 0;
AlignedBuffer<T, N> stack_storage;
};
// A vector without stack storage.
// Could also be a typedef-ed to std::vector,
// but might as well use the one we have.
template <typename T>
using Vector = SmallVector<T, 0>;
#else // SPIRV_CROSS_FORCE_STL_TYPES
template <typename T, size_t N = 8>
using SmallVector = std::vector<T>;
template <typename T>
using Vector = std::vector<T>;
template <typename T>
using VectorView = std::vector<T>;
#endif // SPIRV_CROSS_FORCE_STL_TYPES
// An object pool which we use for allocating IVariant-derived objects.
// We know we are going to allocate a bunch of objects of each type,
// so amortize the mallocs.
class ObjectPoolBase
{
public:
virtual ~ObjectPoolBase() = default;
virtual void free_opaque(void *ptr) = 0;
};
template <typename T>
class ObjectPool : public ObjectPoolBase
{
public:
explicit ObjectPool(unsigned start_object_count_ = 16)
: start_object_count(start_object_count_)
{
}
template <typename... P>
T *allocate(P &&... p)
{
if (vacants.empty())
{
unsigned num_objects = start_object_count << memory.size();
T *ptr = static_cast<T *>(malloc(num_objects * sizeof(T)));
if (!ptr)
return nullptr;
for (unsigned i = 0; i < num_objects; i++)
vacants.push_back(&ptr[i]);
memory.emplace_back(ptr);
}
T *ptr = vacants.back();
vacants.pop_back();
new (ptr) T(std::forward<P>(p)...);
return ptr;
}
void free(T *ptr)
{
ptr->~T();
vacants.push_back(ptr);
}
void free_opaque(void *ptr) override
{
free(static_cast<T *>(ptr));
}
void clear()
{
vacants.clear();
memory.clear();
}
protected:
Vector<T *> vacants;
struct MallocDeleter
{
void operator()(T *ptr)
{
::free(ptr);
}
};
SmallVector<std::unique_ptr<T, MallocDeleter>> memory;
unsigned start_object_count;
};
template <size_t StackSize = 4096, size_t BlockSize = 4096>
class StringStream
{
public:
StringStream()
{
reset();
}
~StringStream()
{
reset();
}
// Disable copies and moves. Makes it easier to implement, and we don't need it.
StringStream(const StringStream &) = delete;
void operator=(const StringStream &) = delete;
template <typename T, typename std::enable_if<!std::is_floating_point<T>::value, int>::type = 0>
StringStream &operator<<(const T &t)
{
auto s = std::to_string(t);
append(s.data(), s.size());
return *this;
}
// Only overload this to make float/double conversions ambiguous.
StringStream &operator<<(uint32_t v)
{
auto s = std::to_string(v);
append(s.data(), s.size());
return *this;
}
StringStream &operator<<(char c)
{
append(&c, 1);
return *this;
}
StringStream &operator<<(const std::string &s)
{
append(s.data(), s.size());
return *this;
}
StringStream &operator<<(const char *s)
{
append(s, strlen(s));
return *this;
}
template <size_t N>
StringStream &operator<<(const char (&s)[N])
{
append(s, strlen(s));
return *this;
}
std::string str() const
{
std::string ret;
size_t target_size = 0;
for (auto &saved : saved_buffers)
target_size += saved.offset;
target_size += current_buffer.offset;
ret.reserve(target_size);
for (auto &saved : saved_buffers)
ret.insert(ret.end(), saved.buffer, saved.buffer + saved.offset);
ret.insert(ret.end(), current_buffer.buffer, current_buffer.buffer + current_buffer.offset);
return ret;
}
void reset()
{
for (auto &saved : saved_buffers)
if (saved.buffer != stack_buffer)
free(saved.buffer);
if (current_buffer.buffer != stack_buffer)
free(current_buffer.buffer);
saved_buffers.clear();
current_buffer.buffer = stack_buffer;
current_buffer.offset = 0;
current_buffer.size = sizeof(stack_buffer);
}
private:
struct Buffer
{
char *buffer = nullptr;
size_t offset = 0;
size_t size = 0;
};
Buffer current_buffer;
char stack_buffer[StackSize];
SmallVector<Buffer> saved_buffers;
void append(const char *s, size_t len)
{
size_t avail = current_buffer.size - current_buffer.offset;
if (avail < len)
{
if (avail > 0)
{
memcpy(current_buffer.buffer + current_buffer.offset, s, avail);
s += avail;
len -= avail;
current_buffer.offset += avail;
}
saved_buffers.push_back(current_buffer);
size_t target_size = len > BlockSize ? len : BlockSize;
current_buffer.buffer = static_cast<char *>(malloc(target_size));
if (!current_buffer.buffer)
SPIRV_CROSS_THROW("Out of memory.");
memcpy(current_buffer.buffer, s, len);
current_buffer.offset = len;
current_buffer.size = target_size;
}
else
{
memcpy(current_buffer.buffer + current_buffer.offset, s, len);
current_buffer.offset += len;
}
}
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
@@ -0,0 +1,87 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_ERROR_HANDLING
#define SPIRV_CROSS_ERROR_HANDLING
#include <stdio.h>
#include <stdlib.h>
#include <string>
#ifndef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS
#include <stdexcept>
#endif
#ifdef SPIRV_CROSS_NAMESPACE_OVERRIDE
#define SPIRV_CROSS_NAMESPACE SPIRV_CROSS_NAMESPACE_OVERRIDE
#else
#define SPIRV_CROSS_NAMESPACE spirv_cross
#endif
namespace SPIRV_CROSS_NAMESPACE
{
#ifdef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS
#if !defined(_MSC_VER) || defined(__clang__)
[[noreturn]]
#elif defined(_MSC_VER)
__declspec(noreturn)
#endif
inline void
report_and_abort(const std::string &msg)
{
#ifdef NDEBUG
(void)msg;
#else
fprintf(stderr, "There was a compiler error: %s\n", msg.c_str());
#endif
fflush(stderr);
abort();
}
#define SPIRV_CROSS_THROW(x) report_and_abort(x)
#else
class CompilerError : public std::runtime_error
{
public:
explicit CompilerError(const std::string &str)
: std::runtime_error(str)
{
}
};
#define SPIRV_CROSS_THROW(x) throw CompilerError(x)
#endif
// MSVC 2013 does not have noexcept. We need this for Variant to get move constructor to work correctly
// instead of copy constructor.
// MSVC 2013 ignores that move constructors cannot throw in std::vector, so just don't define it.
#if defined(_MSC_VER) && _MSC_VER < 1900
#define SPIRV_CROSS_NOEXCEPT
#else
#define SPIRV_CROSS_NOEXCEPT noexcept
#endif
#if __cplusplus >= 201402l
#define SPIRV_CROSS_DEPRECATED(reason) [[deprecated(reason)]]
#elif defined(__GNUC__)
#define SPIRV_CROSS_DEPRECATED(reason) __attribute__((deprecated))
#elif defined(_MSC_VER)
#define SPIRV_CROSS_DEPRECATED(reason) __declspec(deprecated(reason))
#else
#define SPIRV_CROSS_DEPRECATED(reason)
#endif
} // namespace SPIRV_CROSS_NAMESPACE
#endif
@@ -0,0 +1,817 @@
/*
* Copyright 2018-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "spirv_cross_parsed_ir.hpp"
#include <algorithm>
#include <assert.h>
using namespace std;
using namespace spv;
namespace SPIRV_CROSS_NAMESPACE
{
ParsedIR::ParsedIR()
{
// If we move ParsedIR, we need to make sure the pointer stays fixed since the child Variant objects consume a pointer to this group,
// so need an extra pointer here.
pool_group.reset(new ObjectPoolGroup);
pool_group->pools[TypeType].reset(new ObjectPool<SPIRType>);
pool_group->pools[TypeVariable].reset(new ObjectPool<SPIRVariable>);
pool_group->pools[TypeConstant].reset(new ObjectPool<SPIRConstant>);
pool_group->pools[TypeFunction].reset(new ObjectPool<SPIRFunction>);
pool_group->pools[TypeFunctionPrototype].reset(new ObjectPool<SPIRFunctionPrototype>);
pool_group->pools[TypeBlock].reset(new ObjectPool<SPIRBlock>);
pool_group->pools[TypeExtension].reset(new ObjectPool<SPIRExtension>);
pool_group->pools[TypeExpression].reset(new ObjectPool<SPIRExpression>);
pool_group->pools[TypeConstantOp].reset(new ObjectPool<SPIRConstantOp>);
pool_group->pools[TypeCombinedImageSampler].reset(new ObjectPool<SPIRCombinedImageSampler>);
pool_group->pools[TypeAccessChain].reset(new ObjectPool<SPIRAccessChain>);
pool_group->pools[TypeUndef].reset(new ObjectPool<SPIRUndef>);
pool_group->pools[TypeString].reset(new ObjectPool<SPIRString>);
}
// Should have been default-implemented, but need this on MSVC 2013.
ParsedIR::ParsedIR(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT
{
*this = move(other);
}
ParsedIR &ParsedIR::operator=(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT
{
if (this != &other)
{
pool_group = move(other.pool_group);
spirv = move(other.spirv);
meta = move(other.meta);
for (int i = 0; i < TypeCount; i++)
ids_for_type[i] = move(other.ids_for_type[i]);
ids_for_constant_or_type = move(other.ids_for_constant_or_type);
ids_for_constant_or_variable = move(other.ids_for_constant_or_variable);
declared_capabilities = move(other.declared_capabilities);
declared_extensions = move(other.declared_extensions);
block_meta = move(other.block_meta);
continue_block_to_loop_header = move(other.continue_block_to_loop_header);
entry_points = move(other.entry_points);
ids = move(other.ids);
addressing_model = other.addressing_model;
memory_model = other.memory_model;
default_entry_point = other.default_entry_point;
source = other.source;
loop_iteration_depth_hard = other.loop_iteration_depth_hard;
loop_iteration_depth_soft = other.loop_iteration_depth_soft;
}
return *this;
}
ParsedIR::ParsedIR(const ParsedIR &other)
: ParsedIR()
{
*this = other;
}
ParsedIR &ParsedIR::operator=(const ParsedIR &other)
{
if (this != &other)
{
spirv = other.spirv;
meta = other.meta;
for (int i = 0; i < TypeCount; i++)
ids_for_type[i] = other.ids_for_type[i];
ids_for_constant_or_type = other.ids_for_constant_or_type;
ids_for_constant_or_variable = other.ids_for_constant_or_variable;
declared_capabilities = other.declared_capabilities;
declared_extensions = other.declared_extensions;
block_meta = other.block_meta;
continue_block_to_loop_header = other.continue_block_to_loop_header;
entry_points = other.entry_points;
default_entry_point = other.default_entry_point;
source = other.source;
loop_iteration_depth_hard = other.loop_iteration_depth_hard;
loop_iteration_depth_soft = other.loop_iteration_depth_soft;
addressing_model = other.addressing_model;
memory_model = other.memory_model;
// Very deliberate copying of IDs. There is no default copy constructor, nor a simple default constructor.
// Construct object first so we have the correct allocator set-up, then we can copy object into our new pool group.
ids.clear();
ids.reserve(other.ids.size());
for (size_t i = 0; i < other.ids.size(); i++)
{
ids.emplace_back(pool_group.get());
ids.back() = other.ids[i];
}
}
return *this;
}
void ParsedIR::set_id_bounds(uint32_t bounds)
{
ids.reserve(bounds);
while (ids.size() < bounds)
ids.emplace_back(pool_group.get());
block_meta.resize(bounds);
}
// Roll our own versions of these functions to avoid potential locale shenanigans.
static bool is_alpha(char c)
{
return (c >= 'a' && c <= 'z') || (c >= 'A' && c <= 'Z');
}
static bool is_alphanumeric(char c)
{
return is_alpha(c) || (c >= '0' && c <= '9');
}
static string ensure_valid_identifier(const string &name, bool member)
{
// Functions in glslangValidator are mangled with name(<mangled> stuff.
// Normally, we would never see '(' in any legal identifiers, so just strip them out.
auto str = name.substr(0, name.find('('));
for (uint32_t i = 0; i < str.size(); i++)
{
auto &c = str[i];
if (member)
{
// _m<num> variables are reserved by the internal implementation,
// otherwise, make sure the name is a valid identifier.
if (i == 0)
c = is_alpha(c) ? c : '_';
else if (i == 2 && str[0] == '_' && str[1] == 'm')
c = is_alpha(c) ? c : '_';
else
c = is_alphanumeric(c) ? c : '_';
}
else
{
// _<num> variables are reserved by the internal implementation,
// otherwise, make sure the name is a valid identifier.
if (i == 0 || (str[0] == '_' && i == 1))
c = is_alpha(c) ? c : '_';
else
c = is_alphanumeric(c) ? c : '_';
}
}
return str;
}
const string &ParsedIR::get_name(ID id) const
{
auto *m = find_meta(id);
if (m)
return m->decoration.alias;
else
return empty_string;
}
const string &ParsedIR::get_member_name(TypeID id, uint32_t index) const
{
auto *m = find_meta(id);
if (m)
{
if (index >= m->members.size())
return empty_string;
return m->members[index].alias;
}
else
return empty_string;
}
void ParsedIR::set_name(ID id, const string &name)
{
auto &str = meta[id].decoration.alias;
str.clear();
if (name.empty())
return;
// Reserved for temporaries.
if (name[0] == '_' && name.size() >= 2 && isdigit(name[1]))
return;
str = ensure_valid_identifier(name, false);
}
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));
auto &str = meta[id].members[index].alias;
str.clear();
if (name.empty())
return;
// Reserved for unnamed members.
if (name[0] == '_' && name.size() >= 3 && name[1] == 'm' && isdigit(name[2]))
return;
str = ensure_valid_identifier(name, true);
}
void ParsedIR::set_decoration_string(ID id, Decoration decoration, const string &argument)
{
auto &dec = meta[id].decoration;
dec.decoration_flags.set(decoration);
switch (decoration)
{
case DecorationHlslSemanticGOOGLE:
dec.hlsl_semantic = argument;
break;
default:
break;
}
}
void ParsedIR::set_decoration(ID id, Decoration decoration, uint32_t argument)
{
auto &dec = meta[id].decoration;
dec.decoration_flags.set(decoration);
switch (decoration)
{
case DecorationBuiltIn:
dec.builtin = true;
dec.builtin_type = static_cast<BuiltIn>(argument);
break;
case DecorationLocation:
dec.location = argument;
break;
case DecorationComponent:
dec.component = argument;
break;
case DecorationOffset:
dec.offset = argument;
break;
case DecorationArrayStride:
dec.array_stride = argument;
break;
case DecorationMatrixStride:
dec.matrix_stride = argument;
break;
case DecorationBinding:
dec.binding = argument;
break;
case DecorationDescriptorSet:
dec.set = argument;
break;
case DecorationInputAttachmentIndex:
dec.input_attachment = argument;
break;
case DecorationSpecId:
dec.spec_id = argument;
break;
case DecorationIndex:
dec.index = argument;
break;
case DecorationHlslCounterBufferGOOGLE:
meta[id].hlsl_magic_counter_buffer = argument;
meta[argument].hlsl_is_magic_counter_buffer = true;
break;
case DecorationFPRoundingMode:
dec.fp_rounding_mode = static_cast<FPRoundingMode>(argument);
break;
default:
break;
}
}
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];
dec.decoration_flags.set(decoration);
switch (decoration)
{
case DecorationBuiltIn:
dec.builtin = true;
dec.builtin_type = static_cast<BuiltIn>(argument);
break;
case DecorationLocation:
dec.location = argument;
break;
case DecorationComponent:
dec.component = argument;
break;
case DecorationBinding:
dec.binding = argument;
break;
case DecorationOffset:
dec.offset = argument;
break;
case DecorationSpecId:
dec.spec_id = argument;
break;
case DecorationMatrixStride:
dec.matrix_stride = argument;
break;
case DecorationIndex:
dec.index = argument;
break;
default:
break;
}
}
// 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(ID id)
{
switch (ids[id].get_type())
{
case TypeConstant:
get<SPIRConstant>(id).is_used_as_array_length = true;
break;
case TypeConstantOp:
{
auto &cop = get<SPIRConstantOp>(id);
if (cop.opcode == OpCompositeExtract)
mark_used_as_array_length(cop.arguments[0]);
else if (cop.opcode == OpCompositeInsert)
{
mark_used_as_array_length(cop.arguments[0]);
mark_used_as_array_length(cop.arguments[1]);
}
else
for (uint32_t arg_id : cop.arguments)
mark_used_as_array_length(arg_id);
break;
}
case TypeUndef:
break;
default:
assert(0);
}
}
Bitset ParsedIR::get_buffer_block_flags(const SPIRVariable &var) const
{
auto &type = get<SPIRType>(var.basetype);
assert(type.basetype == SPIRType::Struct);
// Some flags like non-writable, non-readable are actually found
// as member decorations. If all members have a decoration set, propagate
// the decoration up as a regular variable decoration.
Bitset base_flags;
auto *m = find_meta(var.self);
if (m)
base_flags = m->decoration.decoration_flags;
if (type.member_types.empty())
return base_flags;
Bitset all_members_flags = get_member_decoration_bitset(type.self, 0);
for (uint32_t i = 1; i < uint32_t(type.member_types.size()); i++)
all_members_flags.merge_and(get_member_decoration_bitset(type.self, i));
base_flags.merge_or(all_members_flags);
return base_flags;
}
const Bitset &ParsedIR::get_member_decoration_bitset(TypeID id, uint32_t index) const
{
auto *m = find_meta(id);
if (m)
{
if (index >= m->members.size())
return cleared_bitset;
return m->members[index].decoration_flags;
}
else
return cleared_bitset;
}
bool ParsedIR::has_decoration(ID id, Decoration decoration) const
{
return get_decoration_bitset(id).get(decoration);
}
uint32_t ParsedIR::get_decoration(ID id, Decoration decoration) const
{
auto *m = find_meta(id);
if (!m)
return 0;
auto &dec = m->decoration;
if (!dec.decoration_flags.get(decoration))
return 0;
switch (decoration)
{
case DecorationBuiltIn:
return dec.builtin_type;
case DecorationLocation:
return dec.location;
case DecorationComponent:
return dec.component;
case DecorationOffset:
return dec.offset;
case DecorationBinding:
return dec.binding;
case DecorationDescriptorSet:
return dec.set;
case DecorationInputAttachmentIndex:
return dec.input_attachment;
case DecorationSpecId:
return dec.spec_id;
case DecorationArrayStride:
return dec.array_stride;
case DecorationMatrixStride:
return dec.matrix_stride;
case DecorationIndex:
return dec.index;
case DecorationFPRoundingMode:
return dec.fp_rounding_mode;
default:
return 1;
}
}
const string &ParsedIR::get_decoration_string(ID id, Decoration decoration) const
{
auto *m = find_meta(id);
if (!m)
return empty_string;
auto &dec = m->decoration;
if (!dec.decoration_flags.get(decoration))
return empty_string;
switch (decoration)
{
case DecorationHlslSemanticGOOGLE:
return dec.hlsl_semantic;
default:
return empty_string;
}
}
void ParsedIR::unset_decoration(ID id, Decoration decoration)
{
auto &dec = meta[id].decoration;
dec.decoration_flags.clear(decoration);
switch (decoration)
{
case DecorationBuiltIn:
dec.builtin = false;
break;
case DecorationLocation:
dec.location = 0;
break;
case DecorationComponent:
dec.component = 0;
break;
case DecorationOffset:
dec.offset = 0;
break;
case DecorationBinding:
dec.binding = 0;
break;
case DecorationDescriptorSet:
dec.set = 0;
break;
case DecorationInputAttachmentIndex:
dec.input_attachment = 0;
break;
case DecorationSpecId:
dec.spec_id = 0;
break;
case DecorationHlslSemanticGOOGLE:
dec.hlsl_semantic.clear();
break;
case DecorationFPRoundingMode:
dec.fp_rounding_mode = FPRoundingModeMax;
break;
case DecorationHlslCounterBufferGOOGLE:
{
auto &counter = meta[id].hlsl_magic_counter_buffer;
if (counter)
{
meta[counter].hlsl_is_magic_counter_buffer = false;
counter = 0;
}
break;
}
default:
break;
}
}
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(TypeID id, uint32_t index, Decoration decoration) const
{
auto *m = find_meta(id);
if (!m)
return 0;
if (index >= m->members.size())
return 0;
auto &dec = m->members[index];
if (!dec.decoration_flags.get(decoration))
return 0;
switch (decoration)
{
case DecorationBuiltIn:
return dec.builtin_type;
case DecorationLocation:
return dec.location;
case DecorationComponent:
return dec.component;
case DecorationBinding:
return dec.binding;
case DecorationOffset:
return dec.offset;
case DecorationSpecId:
return dec.spec_id;
case DecorationIndex:
return dec.index;
default:
return 1;
}
}
const Bitset &ParsedIR::get_decoration_bitset(ID id) const
{
auto *m = find_meta(id);
if (m)
{
auto &dec = m->decoration;
return dec.decoration_flags;
}
else
return cleared_bitset;
}
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];
dec.decoration_flags.set(decoration);
switch (decoration)
{
case DecorationHlslSemanticGOOGLE:
dec.hlsl_semantic = argument;
break;
default:
break;
}
}
const string &ParsedIR::get_member_decoration_string(TypeID id, uint32_t index, Decoration decoration) const
{
auto *m = find_meta(id);
if (m)
{
if (!has_member_decoration(id, index, decoration))
return empty_string;
auto &dec = m->members[index];
switch (decoration)
{
case DecorationHlslSemanticGOOGLE:
return dec.hlsl_semantic;
default:
return empty_string;
}
}
else
return empty_string;
}
void ParsedIR::unset_member_decoration(TypeID id, uint32_t index, Decoration decoration)
{
auto &m = meta[id];
if (index >= m.members.size())
return;
auto &dec = m.members[index];
dec.decoration_flags.clear(decoration);
switch (decoration)
{
case DecorationBuiltIn:
dec.builtin = false;
break;
case DecorationLocation:
dec.location = 0;
break;
case DecorationComponent:
dec.component = 0;
break;
case DecorationOffset:
dec.offset = 0;
break;
case DecorationSpecId:
dec.spec_id = 0;
break;
case DecorationHlslSemanticGOOGLE:
dec.hlsl_semantic.clear();
break;
default:
break;
}
}
uint32_t ParsedIR::increase_bound_by(uint32_t incr_amount)
{
auto curr_bound = ids.size();
auto new_bound = curr_bound + incr_amount;
ids.reserve(ids.size() + incr_amount);
for (uint32_t i = 0; i < incr_amount; i++)
ids.emplace_back(pool_group.get());
block_meta.resize(new_bound);
return uint32_t(curr_bound);
}
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));
}
void ParsedIR::reset_all_of_type(Types type)
{
for (auto &id : ids_for_type[type])
if (ids[id].get_type() == type)
ids[id].reset();
ids_for_type[type].clear();
}
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.");
if (loop_iteration_depth_soft != 0)
{
if (!ids[id].empty())
SPIRV_CROSS_THROW("Cannot override IDs when loop is soft locked.");
return;
}
if (ids[id].empty() || ids[id].get_type() != type)
{
switch (type)
{
case TypeConstant:
ids_for_constant_or_variable.push_back(id);
ids_for_constant_or_type.push_back(id);
break;
case TypeVariable:
ids_for_constant_or_variable.push_back(id);
break;
case TypeType:
case TypeConstantOp:
ids_for_constant_or_type.push_back(id);
break;
default:
break;
}
}
if (ids[id].empty())
{
ids_for_type[type].push_back(id);
}
else if (ids[id].get_type() != type)
{
remove_typed_id(ids[id].get_type(), id);
ids_for_type[type].push_back(id);
}
}
const Meta *ParsedIR::find_meta(ID id) const
{
auto itr = meta.find(id);
if (itr != end(meta))
return &itr->second;
else
return nullptr;
}
Meta *ParsedIR::find_meta(ID id)
{
auto itr = meta.find(id);
if (itr != end(meta))
return &itr->second;
else
return nullptr;
}
ParsedIR::LoopLock ParsedIR::create_loop_hard_lock() const
{
return ParsedIR::LoopLock(&loop_iteration_depth_hard);
}
ParsedIR::LoopLock ParsedIR::create_loop_soft_lock() const
{
return ParsedIR::LoopLock(&loop_iteration_depth_soft);
}
ParsedIR::LoopLock::~LoopLock()
{
if (lock)
(*lock)--;
}
ParsedIR::LoopLock::LoopLock(uint32_t *lock_)
: lock(lock_)
{
if (lock)
(*lock)++;
}
ParsedIR::LoopLock::LoopLock(LoopLock &&other) SPIRV_CROSS_NOEXCEPT
{
*this = move(other);
}
ParsedIR::LoopLock &ParsedIR::LoopLock::operator=(LoopLock &&other) SPIRV_CROSS_NOEXCEPT
{
if (lock)
(*lock)--;
lock = other.lock;
other.lock = nullptr;
return *this;
}
} // namespace SPIRV_CROSS_NAMESPACE
@@ -0,0 +1,229 @@
/*
* Copyright 2018-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_PARSED_IR_HPP
#define SPIRV_CROSS_PARSED_IR_HPP
#include "spirv_common.hpp"
#include <stdint.h>
#include <unordered_map>
namespace SPIRV_CROSS_NAMESPACE
{
// This data structure holds all information needed to perform cross-compilation and reflection.
// It is the output of the Parser, but any implementation could create this structure.
// It is intentionally very "open" and struct-like with some helper functions to deal with decorations.
// Parser is the reference implementation of how this data structure should be filled in.
class ParsedIR
{
private:
// This must be destroyed after the "ids" vector.
std::unique_ptr<ObjectPoolGroup> pool_group;
public:
ParsedIR();
// Due to custom allocations from object pools, we cannot use a default copy constructor.
ParsedIR(const ParsedIR &other);
ParsedIR &operator=(const ParsedIR &other);
// Moves are unproblematic, but we need to implement it anyways, since MSVC 2013 does not understand
// how to default-implement these.
ParsedIR(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT;
ParsedIR &operator=(ParsedIR &&other) SPIRV_CROSS_NOEXCEPT;
// Resizes ids, meta and block_meta.
void set_id_bounds(uint32_t bounds);
// The raw SPIR-V, instructions and opcodes refer to this by offset + count.
std::vector<uint32_t> spirv;
// Holds various data structures which inherit from IVariant.
SmallVector<Variant> ids;
// Various meta data for IDs, decorations, names, etc.
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<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<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.
SmallVector<spv::Capability> declared_capabilities;
SmallVector<std::string> declared_extensions;
// Meta data about blocks. The cross-compiler needs to query if a block is either of these types.
// It is a bitset as there can be more than one tag per block.
enum BlockMetaFlagBits
{
BLOCK_META_LOOP_HEADER_BIT = 1 << 0,
BLOCK_META_CONTINUE_BIT = 1 << 1,
BLOCK_META_LOOP_MERGE_BIT = 1 << 2,
BLOCK_META_SELECTION_MERGE_BIT = 1 << 3,
BLOCK_META_MULTISELECT_MERGE_BIT = 1 << 4
};
using BlockMetaFlags = uint8_t;
SmallVector<BlockMetaFlags> block_meta;
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<FunctionID, SPIREntryPoint> entry_points;
FunctionID default_entry_point = 0;
struct Source
{
uint32_t version = 0;
bool es = false;
bool known = false;
bool hlsl = false;
Source() = default;
};
Source source;
spv::AddressingModel addressing_model = spv::AddressingModelMax;
spv::MemoryModel memory_model = spv::MemoryModelMax;
// Decoration handling methods.
// 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(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(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(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(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, ID id);
void remove_typed_id(Types type, ID id);
class LoopLock
{
public:
explicit LoopLock(uint32_t *counter);
LoopLock(const LoopLock &) = delete;
void operator=(const LoopLock &) = delete;
LoopLock(LoopLock &&other) SPIRV_CROSS_NOEXCEPT;
LoopLock &operator=(LoopLock &&other) SPIRV_CROSS_NOEXCEPT;
~LoopLock();
private:
uint32_t *lock;
};
// This must be held while iterating over a type ID array.
// It is undefined if someone calls set<>() while we're iterating over a data structure, so we must
// make sure that this case is avoided.
// If we have a hard lock, it is an error to call set<>(), and an exception is thrown.
// If we have a soft lock, we silently ignore any additions to the typed arrays.
// This should only be used for physical ID remapping where we need to create an ID, but we will never
// care about iterating over them.
LoopLock create_loop_hard_lock() const;
LoopLock create_loop_soft_lock() const;
template <typename T, typename Op>
void for_each_typed_id(const Op &op)
{
auto loop_lock = create_loop_hard_lock();
for (auto &id : ids_for_type[T::type])
{
if (ids[id].get_type() == static_cast<Types>(T::type))
op(id, get<T>(id));
}
}
template <typename T, typename Op>
void for_each_typed_id(const Op &op) const
{
auto loop_lock = create_loop_hard_lock();
for (auto &id : ids_for_type[T::type])
{
if (ids[id].get_type() == static_cast<Types>(T::type))
op(id, get<T>(id));
}
}
template <typename T>
void reset_all_of_type()
{
reset_all_of_type(static_cast<Types>(T::type));
}
void reset_all_of_type(Types type);
Meta *find_meta(ID id);
const Meta *find_meta(ID id) const;
const std::string &get_empty_string() const
{
return empty_string;
}
private:
template <typename T>
T &get(uint32_t id)
{
return variant_get<T>(ids[id]);
}
template <typename T>
const T &get(uint32_t id) const
{
return variant_get<T>(ids[id]);
}
mutable uint32_t loop_iteration_depth_hard = 0;
mutable uint32_t loop_iteration_depth_soft = 0;
std::string empty_string;
Bitset cleared_bitset;
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
@@ -0,0 +1,70 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "spirv_cross_util.hpp"
#include "spirv_common.hpp"
using namespace spv;
using namespace SPIRV_CROSS_NAMESPACE;
namespace spirv_cross_util
{
void rename_interface_variable(Compiler &compiler, const SmallVector<Resource> &resources, uint32_t location,
const std::string &name)
{
for (auto &v : resources)
{
if (!compiler.has_decoration(v.id, spv::DecorationLocation))
continue;
auto loc = compiler.get_decoration(v.id, spv::DecorationLocation);
if (loc != location)
continue;
auto &type = compiler.get_type(v.base_type_id);
// This is more of a friendly variant. If we need to rename interface variables, we might have to rename
// structs as well and make sure all the names match up.
if (type.basetype == SPIRType::Struct)
{
compiler.set_name(v.base_type_id, join("SPIRV_Cross_Interface_Location", location));
for (uint32_t i = 0; i < uint32_t(type.member_types.size()); i++)
compiler.set_member_name(v.base_type_id, i, join("InterfaceMember", i));
}
compiler.set_name(v.id, name);
}
}
void inherit_combined_sampler_bindings(Compiler &compiler)
{
auto &samplers = compiler.get_combined_image_samplers();
for (auto &s : samplers)
{
if (compiler.has_decoration(s.image_id, spv::DecorationDescriptorSet))
{
uint32_t set = compiler.get_decoration(s.image_id, spv::DecorationDescriptorSet);
compiler.set_decoration(s.combined_id, spv::DecorationDescriptorSet, set);
}
if (compiler.has_decoration(s.image_id, spv::DecorationBinding))
{
uint32_t binding = compiler.get_decoration(s.image_id, spv::DecorationBinding);
compiler.set_decoration(s.combined_id, spv::DecorationBinding, binding);
}
}
}
} // namespace spirv_cross_util
@@ -0,0 +1,30 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_UTIL_HPP
#define SPIRV_CROSS_UTIL_HPP
#include "spirv_cross.hpp"
namespace spirv_cross_util
{
void rename_interface_variable(SPIRV_CROSS_NAMESPACE::Compiler &compiler,
const SPIRV_CROSS_NAMESPACE::SmallVector<SPIRV_CROSS_NAMESPACE::Resource> &resources,
uint32_t location, const std::string &name);
void inherit_combined_sampler_bindings(SPIRV_CROSS_NAMESPACE::Compiler &compiler);
} // namespace spirv_cross_util
#endif
File diff suppressed because it is too large Load Diff
+720
View File
@@ -0,0 +1,720 @@
/*
* Copyright 2015-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_GLSL_HPP
#define SPIRV_CROSS_GLSL_HPP
#include "GLSL.std.450.h"
#include "spirv_cross.hpp"
#include <unordered_map>
#include <unordered_set>
#include <utility>
namespace SPIRV_CROSS_NAMESPACE
{
enum PlsFormat
{
PlsNone = 0,
PlsR11FG11FB10F,
PlsR32F,
PlsRG16F,
PlsRGB10A2,
PlsRGBA8,
PlsRG16,
PlsRGBA8I,
PlsRG16I,
PlsRGB10A2UI,
PlsRGBA8UI,
PlsRG16UI,
PlsR32UI
};
struct PlsRemap
{
uint32_t id;
PlsFormat format;
};
enum AccessChainFlagBits
{
ACCESS_CHAIN_INDEX_IS_LITERAL_BIT = 1 << 0,
ACCESS_CHAIN_CHAIN_ONLY_BIT = 1 << 1,
ACCESS_CHAIN_PTR_CHAIN_BIT = 1 << 2,
ACCESS_CHAIN_SKIP_REGISTER_EXPRESSION_READ_BIT = 1 << 3,
ACCESS_CHAIN_LITERAL_MSB_FORCE_ID = 1 << 4
};
typedef uint32_t AccessChainFlags;
class CompilerGLSL : public Compiler
{
public:
struct Options
{
// The shading language version. Corresponds to #version $VALUE.
uint32_t version = 450;
// Emit the OpenGL ES shading language instead of desktop OpenGL.
bool es = false;
// Debug option to always emit temporary variables for all expressions.
bool force_temporary = false;
// If true, Vulkan GLSL features are used instead of GL-compatible features.
// Mostly useful for debugging SPIR-V files.
bool vulkan_semantics = false;
// If true, gl_PerVertex is explicitly redeclared in vertex, geometry and tessellation shaders.
// The members of gl_PerVertex is determined by which built-ins are declared by the shader.
// This option is ignored in ES versions, as redeclaration in ES is not required, and it depends on a different extension
// (EXT_shader_io_blocks) which makes things a bit more fuzzy.
bool separate_shader_objects = false;
// Flattens multidimensional arrays, e.g. float foo[a][b][c] into single-dimensional arrays,
// e.g. float foo[a * b * c].
// This function does not change the actual SPIRType of any object.
// Only the generated code, including declarations of interface variables are changed to be single array dimension.
bool flatten_multidimensional_arrays = false;
// For older desktop GLSL targets than version 420, the
// GL_ARB_shading_language_420pack extensions is used to be able to support
// layout(binding) on UBOs and samplers.
// If disabled on older targets, binding decorations will be stripped.
bool enable_420pack_extension = true;
// In non-Vulkan GLSL, emit push constant blocks as UBOs rather than plain uniforms.
bool emit_push_constant_as_uniform_buffer = false;
// Always emit uniform blocks as plain uniforms, regardless of the GLSL version, even when UBOs are supported.
// Does not apply to shader storage or push constant blocks.
bool emit_uniform_buffer_as_plain_uniforms = false;
// Emit OpLine directives if present in the module.
// May not correspond exactly to original source, but should be a good approximation.
bool emit_line_directives = false;
enum Precision
{
DontCare,
Lowp,
Mediump,
Highp
};
struct
{
// GLSL: In vertex shaders, rewrite [0, w] depth (Vulkan/D3D style) to [-w, w] depth (GL style).
// MSL: In vertex shaders, rewrite [-w, w] depth (GL style) to [0, w] depth.
// HLSL: In vertex shaders, rewrite [-w, w] depth (GL style) to [0, w] depth.
bool fixup_clipspace = false;
// Inverts gl_Position.y or equivalent.
bool flip_vert_y = false;
// GLSL only, for HLSL version of this option, see CompilerHLSL.
// If true, the backend will assume that InstanceIndex will need to apply
// a base instance offset. Set to false if you know you will never use base instance
// functionality as it might remove some internal uniforms.
bool support_nonzero_base_instance = true;
} vertex;
struct
{
// Add precision mediump float in ES targets when emitting GLES source.
// Add precision highp int in ES targets when emitting GLES source.
Precision default_float_precision = Mediump;
Precision default_int_precision = Highp;
} fragment;
};
void remap_pixel_local_storage(std::vector<PlsRemap> inputs, std::vector<PlsRemap> outputs)
{
pls_inputs = std::move(inputs);
pls_outputs = std::move(outputs);
remap_pls_variables();
}
explicit CompilerGLSL(std::vector<uint32_t> spirv_)
: Compiler(std::move(spirv_))
{
init();
}
CompilerGLSL(const uint32_t *ir_, size_t word_count)
: Compiler(ir_, word_count)
{
init();
}
explicit CompilerGLSL(const ParsedIR &ir_)
: Compiler(ir_)
{
init();
}
explicit CompilerGLSL(ParsedIR &&ir_)
: Compiler(std::move(ir_))
{
init();
}
const Options &get_common_options() const
{
return options;
}
void set_common_options(const Options &opts)
{
options = opts;
}
std::string compile() override;
// Returns the current string held in the conversion buffer. Useful for
// capturing what has been converted so far when compile() throws an error.
std::string get_partial_source();
// Adds a line to be added right after #version in GLSL backend.
// This is useful for enabling custom extensions which are outside the scope of SPIRV-Cross.
// This can be combined with variable remapping.
// A new-line will be added.
//
// While add_header_line() is a more generic way of adding arbitrary text to the header
// of a GLSL file, require_extension() should be used when adding extensions since it will
// avoid creating collisions with SPIRV-Cross generated extensions.
//
// Code added via add_header_line() is typically backend-specific.
void add_header_line(const std::string &str);
// Adds an extension which is required to run this shader, e.g.
// require_extension("GL_KHR_my_extension");
void require_extension(const std::string &ext);
// Legacy GLSL compatibility method.
// Takes a uniform or push constant variable and flattens it into a (i|u)vec4 array[N]; array instead.
// 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(VariableID id);
protected:
void reset();
void emit_function(SPIRFunction &func, const Bitset &return_flags);
bool has_extension(const std::string &ext) const;
void require_extension_internal(const std::string &ext);
// Virtualize methods which need to be overridden by subclass targets like C++ and such.
virtual void emit_function_prototype(SPIRFunction &func, const Bitset &return_flags);
SPIRBlock *current_emitting_block = nullptr;
SPIRBlock *current_emitting_switch = nullptr;
bool current_emitting_switch_fallthrough = false;
virtual void emit_instruction(const Instruction &instr);
void emit_block_instructions(SPIRBlock &block);
virtual void emit_glsl_op(uint32_t result_type, uint32_t result_id, uint32_t op, const uint32_t *args,
uint32_t count);
virtual void emit_spv_amd_shader_ballot_op(uint32_t result_type, uint32_t result_id, uint32_t op,
const uint32_t *args, uint32_t count);
virtual void emit_spv_amd_shader_explicit_vertex_parameter_op(uint32_t result_type, uint32_t result_id, uint32_t op,
const uint32_t *args, uint32_t count);
virtual void emit_spv_amd_shader_trinary_minmax_op(uint32_t result_type, uint32_t result_id, uint32_t op,
const uint32_t *args, uint32_t count);
virtual void emit_spv_amd_gcn_shader_op(uint32_t result_type, uint32_t result_id, uint32_t op, const uint32_t *args,
uint32_t count);
virtual void emit_header();
void emit_line_directive(uint32_t file_id, uint32_t line_literal);
void build_workgroup_size(SmallVector<std::string> &arguments, const SpecializationConstant &x,
const SpecializationConstant &y, const SpecializationConstant &z);
virtual void emit_sampled_image_op(uint32_t result_type, uint32_t result_id, uint32_t image_id, uint32_t samp_id);
virtual void emit_texture_op(const Instruction &i);
virtual std::string to_texture_op(const Instruction &i, bool *forward,
SmallVector<uint32_t> &inherited_expressions);
virtual void emit_subgroup_op(const Instruction &i);
virtual std::string type_to_glsl(const SPIRType &type, uint32_t id = 0);
virtual std::string builtin_to_glsl(spv::BuiltIn builtin, spv::StorageClass storage);
virtual void emit_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index,
const std::string &qualifier = "", uint32_t base_offset = 0);
virtual void emit_struct_padding_target(const SPIRType &type);
virtual std::string image_type_glsl(const SPIRType &type, uint32_t id = 0);
std::string constant_expression(const SPIRConstant &c);
std::string constant_op_expression(const SPIRConstantOp &cop);
virtual std::string constant_expression_vector(const SPIRConstant &c, uint32_t vector);
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(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(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);
virtual void emit_buffer_block(const SPIRVariable &type);
virtual void emit_push_constant_block(const SPIRVariable &var);
virtual void emit_uniform(const SPIRVariable &var);
virtual std::string unpack_expression_type(std::string expr_str, const SPIRType &type, uint32_t physical_type_id,
bool packed_type, bool row_major);
virtual bool builtin_translates_to_nonarray(spv::BuiltIn builtin) const;
void emit_copy_logical_type(uint32_t lhs_id, uint32_t lhs_type_id, uint32_t rhs_id, uint32_t rhs_type_id,
SmallVector<uint32_t> chain);
StringStream<> buffer;
template <typename T>
inline void statement_inner(T &&t)
{
buffer << std::forward<T>(t);
statement_count++;
}
template <typename T, typename... Ts>
inline void statement_inner(T &&t, Ts &&... ts)
{
buffer << std::forward<T>(t);
statement_count++;
statement_inner(std::forward<Ts>(ts)...);
}
template <typename... Ts>
inline void statement(Ts &&... ts)
{
if (is_forcing_recompilation())
{
// Do not bother emitting code while force_recompile is active.
// We will compile again.
statement_count++;
return;
}
if (redirect_statement)
{
redirect_statement->push_back(join(std::forward<Ts>(ts)...));
statement_count++;
}
else
{
for (uint32_t i = 0; i < indent; i++)
buffer << " ";
statement_inner(std::forward<Ts>(ts)...);
buffer << '\n';
}
}
template <typename... Ts>
inline void statement_no_indent(Ts &&... ts)
{
auto old_indent = indent;
indent = 0;
statement(std::forward<Ts>(ts)...);
indent = old_indent;
}
// Used for implementing continue blocks where
// we want to obtain a list of statements we can merge
// on a single line separated by comma.
SmallVector<std::string> *redirect_statement = nullptr;
const SPIRBlock *current_continue_block = nullptr;
void begin_scope();
void end_scope();
void end_scope(const std::string &trailer);
void end_scope_decl();
void end_scope_decl(const std::string &decl);
Options options;
virtual std::string type_to_array_glsl(
const SPIRType &type); // Allow Metal to use the array<T> template to make arrays a value type
std::string to_array_size(const SPIRType &type, uint32_t index);
uint32_t to_array_size_literal(const SPIRType &type, uint32_t index) const;
uint32_t to_array_size_literal(const SPIRType &type) const;
virtual std::string variable_decl(const SPIRVariable &variable); // Threadgroup arrays can't have a wrapper type
std::string variable_decl_function_local(SPIRVariable &variable);
void add_local_variable_name(uint32_t id);
void add_resource_name(uint32_t id);
void add_member_name(SPIRType &type, uint32_t name);
void add_function_overload(const SPIRFunction &func);
virtual bool is_non_native_row_major_matrix(uint32_t id);
virtual bool member_is_non_native_row_major_matrix(const SPIRType &type, uint32_t index);
bool member_is_remapped_physical_type(const SPIRType &type, uint32_t index) const;
bool member_is_packed_physical_type(const SPIRType &type, uint32_t index) const;
virtual std::string convert_row_major_matrix(std::string exp_str, const SPIRType &exp_type,
uint32_t physical_type_id, bool is_packed);
std::unordered_set<std::string> local_variable_names;
std::unordered_set<std::string> resource_names;
std::unordered_set<std::string> block_input_names;
std::unordered_set<std::string> block_output_names;
std::unordered_set<std::string> block_ubo_names;
std::unordered_set<std::string> block_ssbo_names;
std::unordered_set<std::string> block_names; // A union of all block_*_names.
std::unordered_map<std::string, std::unordered_set<uint64_t>> function_overloads;
std::unordered_map<uint32_t, std::string> preserved_aliases;
void preserve_alias_on_reset(uint32_t id);
void reset_name_caches();
bool processing_entry_point = false;
// Can be overriden by subclass backends for trivial things which
// shouldn't need polymorphism.
struct BackendVariations
{
std::string discard_literal = "discard";
std::string demote_literal = "demote";
std::string null_pointer_literal = "";
bool float_literal_suffix = false;
bool double_literal_suffix = true;
bool uint32_t_literal_suffix = true;
bool long_long_literal_suffix = false;
const char *basic_int_type = "int";
const char *basic_uint_type = "uint";
const char *basic_int8_type = "int8_t";
const char *basic_uint8_type = "uint8_t";
const char *basic_int16_type = "int16_t";
const char *basic_uint16_type = "uint16_t";
const char *int16_t_literal_suffix = "s";
const char *uint16_t_literal_suffix = "us";
const char *nonuniform_qualifier = "nonuniformEXT";
const char *boolean_mix_function = "mix";
bool swizzle_is_function = false;
bool shared_is_implied = false;
bool unsized_array_supported = true;
bool explicit_struct_type = false;
bool use_initializer_list = false;
bool use_typed_initializer_list = false;
bool can_declare_struct_inline = true;
bool can_declare_arrays_inline = true;
bool native_row_major_matrix = true;
bool use_constructor_splatting = true;
bool allow_precision_qualifiers = false;
bool can_swizzle_scalar = false;
bool force_gl_in_out_block = false;
bool can_return_array = true;
bool allow_truncated_access_chain = false;
bool supports_extensions = false;
bool supports_empty_struct = false;
bool array_is_value_type = true;
bool comparison_image_samples_scalar = false;
bool native_pointers = false;
bool support_small_type_sampling_result = false;
bool support_case_fallthrough = true;
bool use_array_constructor = false;
} backend;
void emit_struct(SPIRType &type);
void emit_resources();
void emit_buffer_block_native(const SPIRVariable &var);
void emit_buffer_reference_block(SPIRType &type, bool forward_declaration);
void emit_buffer_block_legacy(const SPIRVariable &var);
void emit_buffer_block_flattened(const SPIRVariable &type);
void emit_declared_builtin_block(spv::StorageClass storage, spv::ExecutionModel model);
void emit_push_constant_block_vulkan(const SPIRVariable &var);
void emit_push_constant_block_glsl(const SPIRVariable &var);
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<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(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);
bool should_dereference(uint32_t id);
bool should_forward(uint32_t id) const;
bool should_suppress_usage_tracking(uint32_t id) const;
void emit_mix_op(uint32_t result_type, uint32_t id, uint32_t left, uint32_t right, uint32_t lerp);
void emit_nminmax_op(uint32_t result_type, uint32_t id, uint32_t op0, uint32_t op1, GLSLstd450 op);
bool to_trivial_mix_op(const SPIRType &type, std::string &op, uint32_t left, uint32_t right, uint32_t lerp);
void emit_quaternary_func_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2,
uint32_t op3, const char *op);
void emit_trinary_func_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2,
const char *op);
void emit_binary_func_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op);
void emit_unary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, const char *op,
SPIRType::BaseType input_type, SPIRType::BaseType expected_result_type);
void emit_binary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op,
SPIRType::BaseType input_type, bool skip_cast_if_equal_type);
void emit_trinary_func_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2,
const char *op, SPIRType::BaseType input_type);
void emit_trinary_func_op_bitextract(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1,
uint32_t op2, const char *op, SPIRType::BaseType expected_result_type,
SPIRType::BaseType input_type0, SPIRType::BaseType input_type1,
SPIRType::BaseType input_type2);
void emit_bitfield_insert_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, uint32_t op2,
uint32_t op3, const char *op, SPIRType::BaseType offset_count_type);
void emit_unary_func_op(uint32_t result_type, uint32_t result_id, uint32_t op0, const char *op);
void emit_unrolled_unary_op(uint32_t result_type, uint32_t result_id, uint32_t operand, const char *op);
void emit_binary_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op);
void emit_unrolled_binary_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op,
bool negate, SPIRType::BaseType expected_type);
void emit_binary_op_cast(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op,
SPIRType::BaseType input_type, bool skip_cast_if_equal_type);
SPIRType binary_op_bitcast_helper(std::string &cast_op0, std::string &cast_op1, SPIRType::BaseType &input_type,
uint32_t op0, uint32_t op1, bool skip_cast_if_equal_type);
std::string to_ternary_expression(const SPIRType &result_type, uint32_t select, uint32_t true_value,
uint32_t false_value);
void emit_unary_op(uint32_t result_type, uint32_t result_id, uint32_t op0, const char *op);
bool expression_is_forwarded(uint32_t id) const;
bool expression_suppresses_usage_tracking(uint32_t id) const;
SPIRExpression &emit_op(uint32_t result_type, uint32_t result_id, const std::string &rhs, bool forward_rhs,
bool suppress_usage_tracking = false);
void access_chain_internal_append_index(std::string &expr, uint32_t base, const SPIRType *type,
AccessChainFlags flags, bool &access_chain_is_arrayed, uint32_t index);
std::string access_chain_internal(uint32_t base, const uint32_t *indices, uint32_t count, AccessChainFlags flags,
AccessChainMeta *meta);
std::string access_chain(uint32_t base, const uint32_t *indices, uint32_t count, const SPIRType &target_type,
AccessChainMeta *meta = nullptr, bool ptr_chain = false);
std::string flattened_access_chain(uint32_t base, const uint32_t *indices, uint32_t count,
const SPIRType &target_type, uint32_t offset, uint32_t matrix_stride,
uint32_t array_stride, bool need_transpose);
std::string flattened_access_chain_struct(uint32_t base, const uint32_t *indices, uint32_t count,
const SPIRType &target_type, uint32_t offset);
std::string flattened_access_chain_matrix(uint32_t base, const uint32_t *indices, uint32_t count,
const SPIRType &target_type, uint32_t offset, uint32_t matrix_stride,
bool need_transpose);
std::string flattened_access_chain_vector(uint32_t base, const uint32_t *indices, uint32_t count,
const SPIRType &target_type, uint32_t offset, uint32_t matrix_stride,
bool need_transpose);
std::pair<std::string, uint32_t> flattened_access_chain_offset(const SPIRType &basetype, const uint32_t *indices,
uint32_t count, uint32_t offset,
uint32_t word_stride, bool *need_transpose = nullptr,
uint32_t *matrix_stride = nullptr,
uint32_t *array_stride = nullptr,
bool ptr_chain = false);
const char *index_to_swizzle(uint32_t index);
std::string remap_swizzle(const SPIRType &result_type, uint32_t input_components, const std::string &expr);
std::string declare_temporary(uint32_t type, uint32_t id);
void emit_uninitialized_temporary(uint32_t type, uint32_t id);
SPIRExpression &emit_uninitialized_temporary_expression(uint32_t type, uint32_t id);
void append_global_func_args(const SPIRFunction &func, uint32_t index, SmallVector<std::string> &arglist);
std::string to_expression(uint32_t id, bool register_expression_read = true);
std::string to_composite_constructor_expression(uint32_t id);
std::string to_rerolled_array_expression(const std::string &expr, const SPIRType &type);
std::string to_enclosed_expression(uint32_t id, bool register_expression_read = true);
std::string to_unpacked_expression(uint32_t id, bool register_expression_read = true);
std::string to_unpacked_row_major_matrix_expression(uint32_t id);
std::string to_enclosed_unpacked_expression(uint32_t id, bool register_expression_read = true);
std::string to_dereferenced_expression(uint32_t id, bool register_expression_read = true);
std::string to_pointer_expression(uint32_t id, bool register_expression_read = true);
std::string to_enclosed_pointer_expression(uint32_t id, bool register_expression_read = true);
std::string to_extract_component_expression(uint32_t id, uint32_t index);
std::string enclose_expression(const std::string &expr);
std::string dereference_expression(const SPIRType &expression_type, const std::string &expr);
std::string address_of_expression(const std::string &expr);
void strip_enclosed_expression(std::string &expr);
std::string to_member_name(const SPIRType &type, uint32_t index);
virtual std::string to_member_reference(uint32_t base, const SPIRType &type, uint32_t index, bool ptr_chain);
std::string type_to_glsl_constructor(const SPIRType &type);
std::string argument_decl(const SPIRFunction::Parameter &arg);
virtual std::string to_qualifiers_glsl(uint32_t id);
const char *to_precision_qualifiers_glsl(uint32_t id);
virtual const char *to_storage_qualifiers_glsl(const SPIRVariable &var);
const char *flags_to_qualifiers_glsl(const SPIRType &type, const Bitset &flags);
const char *format_to_glsl(spv::ImageFormat format);
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(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);
virtual void emit_block_hints(const SPIRBlock &block);
virtual std::string to_initializer_expression(const SPIRVariable &var);
bool buffer_is_packing_standard(const SPIRType &type, BufferPackingStandard packing,
uint32_t *failed_index = nullptr, uint32_t start_offset = 0,
uint32_t end_offset = ~(0u));
std::string buffer_to_packing_standard(const SPIRType &type, bool support_std430_without_scalar_layout);
uint32_t type_to_packed_base_size(const SPIRType &type, BufferPackingStandard packing);
uint32_t type_to_packed_alignment(const SPIRType &type, const Bitset &flags, BufferPackingStandard packing);
uint32_t type_to_packed_array_stride(const SPIRType &type, const Bitset &flags, BufferPackingStandard packing);
uint32_t type_to_packed_size(const SPIRType &type, const Bitset &flags, BufferPackingStandard packing);
std::string bitcast_glsl(const SPIRType &result_type, uint32_t arg);
virtual std::string bitcast_glsl_op(const SPIRType &result_type, const SPIRType &argument_type);
std::string bitcast_expression(SPIRType::BaseType target_type, uint32_t arg);
std::string bitcast_expression(const SPIRType &target_type, SPIRType::BaseType expr_type, const std::string &expr);
std::string build_composite_combiner(uint32_t result_type, const uint32_t *elems, uint32_t length);
bool remove_duplicate_swizzle(std::string &op);
bool remove_unity_swizzle(uint32_t base, std::string &op);
// Can modify flags to remote readonly/writeonly if image type
// and force recompile.
bool check_atomic_image(uint32_t id);
virtual void replace_illegal_names();
virtual void emit_entry_point_declarations();
void replace_fragment_output(SPIRVariable &var);
void replace_fragment_outputs();
bool check_explicit_lod_allowed(uint32_t lod);
std::string legacy_tex_op(const std::string &op, const SPIRType &imgtype, uint32_t lod, uint32_t id);
uint32_t indent = 0;
std::unordered_set<uint32_t> emitted_functions;
// Ensure that we declare phi-variable copies even if the original declaration isn't deferred
std::unordered_set<uint32_t> flushed_phi_variables;
std::unordered_set<uint32_t> flattened_buffer_blocks;
std::unordered_set<uint32_t> flattened_structs;
std::string load_flattened_struct(SPIRVariable &var);
std::string to_flattened_struct_member(const SPIRVariable &var, uint32_t index);
void store_flattened_struct(SPIRVariable &var, uint32_t value);
// Usage tracking. If a temporary is used more than once, use the temporary instead to
// avoid AST explosion when SPIRV is generated with pure SSA and doesn't write stuff to variables.
std::unordered_map<uint32_t, uint32_t> expression_usage_counts;
void track_expression_read(uint32_t id);
SmallVector<std::string> forced_extensions;
SmallVector<std::string> header_lines;
// Used when expressions emit extra opcodes with their own unique IDs,
// and we need to reuse the IDs across recompilation loops.
// Currently used by NMin/Max/Clamp implementations.
std::unordered_map<uint32_t, uint32_t> extra_sub_expressions;
uint32_t statement_count = 0;
inline bool is_legacy() const
{
return (options.es && options.version < 300) || (!options.es && options.version < 130);
}
inline bool is_legacy_es() const
{
return options.es && options.version < 300;
}
inline bool is_legacy_desktop() const
{
return !options.es && options.version < 130;
}
bool args_will_forward(uint32_t id, const uint32_t *args, uint32_t num_args, bool pure);
void register_call_out_argument(uint32_t id);
void register_impure_function_call();
void register_control_dependent_expression(uint32_t expr);
// GL_EXT_shader_pixel_local_storage support.
std::vector<PlsRemap> pls_inputs;
std::vector<PlsRemap> pls_outputs;
std::string pls_decl(const PlsRemap &variable);
const char *to_pls_qualifiers_glsl(const SPIRVariable &variable);
void emit_pls();
void remap_pls_variables();
// A variant which takes two sets of name. The secondary is only used to verify there are no collisions,
// but the set is not updated when we have found a new name.
// Used primarily when adding block interface names.
void add_variable(std::unordered_set<std::string> &variables_primary,
const std::unordered_set<std::string> &variables_secondary, std::string &name);
void check_function_call_constraints(const uint32_t *args, uint32_t length);
void handle_invalid_expression(uint32_t id);
void find_static_extensions();
std::string emit_for_loop_initializers(const SPIRBlock &block);
void emit_while_loop_initializers(const SPIRBlock &block);
bool for_loop_initializers_are_same_type(const SPIRBlock &block);
bool optimize_read_modify_write(const SPIRType &type, const std::string &lhs, const std::string &rhs);
void fixup_image_load_store_access();
bool type_is_empty(const SPIRType &type);
virtual void declare_undefined_values();
static std::string sanitize_underscores(const std::string &str);
bool can_use_io_location(spv::StorageClass storage, bool block);
const Instruction *get_next_instruction_in_block(const Instruction &instr);
static uint32_t mask_relevant_memory_semantics(uint32_t semantics);
std::string convert_half_to_string(const SPIRConstant &value, uint32_t col, uint32_t row);
std::string convert_float_to_string(const SPIRConstant &value, uint32_t col, uint32_t row);
std::string convert_double_to_string(const SPIRConstant &value, uint32_t col, uint32_t row);
std::string convert_separate_image_to_expression(uint32_t id);
// Builtins in GLSL are always specific signedness, but the SPIR-V can declare them
// as either unsigned or signed.
// Sometimes we will need to automatically perform bitcasts on load and store to make this work.
virtual void bitcast_to_builtin_store(uint32_t target_id, std::string &expr, const SPIRType &expr_type);
virtual void bitcast_from_builtin_load(uint32_t source_id, std::string &expr, const SPIRType &expr_type);
void unroll_array_from_complex_load(uint32_t target_id, uint32_t source_id, std::string &expr);
void convert_non_uniform_expression(const SPIRType &type, std::string &expr);
void handle_store_to_invariant_variable(uint32_t store_id, uint32_t value_id);
void disallow_forwarding_in_expression_chain(const SPIRExpression &expr);
bool expression_is_constant_null(uint32_t id) const;
virtual void emit_store_statement(uint32_t lhs_expression, uint32_t rhs_expression);
uint32_t get_integer_width_for_instruction(const Instruction &instr) const;
uint32_t get_integer_width_for_glsl_instruction(GLSLstd450 op, const uint32_t *arguments, uint32_t length) const;
bool variable_is_lut(const SPIRVariable &var) const;
char current_locale_radix_character = '.';
void fixup_type_alias();
void reorder_type_alias();
void propagate_nonuniform_qualifier(uint32_t id);
static const char *vector_swizzle(int vecsize, int index);
private:
void init();
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
File diff suppressed because it is too large Load Diff
+273
View File
@@ -0,0 +1,273 @@
/*
* Copyright 2016-2019 Robert Konrad
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_HLSL_HPP
#define SPIRV_HLSL_HPP
#include "spirv_glsl.hpp"
#include <utility>
namespace SPIRV_CROSS_NAMESPACE
{
// Interface which remaps vertex inputs to a fixed semantic name to make linking easier.
struct HLSLVertexAttributeRemap
{
uint32_t location;
std::string semantic;
};
// Specifying a root constant (d3d12) or push constant range (vulkan).
//
// `start` and `end` denotes the range of the root constant in bytes.
// Both values need to be multiple of 4.
struct RootConstants
{
uint32_t start;
uint32_t end;
uint32_t binding;
uint32_t space;
};
// For finer control, decorations may be removed from specific resources instead with unset_decoration().
enum HLSLBindingFlagBits
{
// Push constant (root constant) resources will be declared as CBVs (b-space) without a register() declaration.
// A register will be automatically assigned by the D3D compiler, but must therefore be reflected in D3D-land.
// Push constants do not normally have a DecorationBinding set, but if they do, this can be used to ignore it.
HLSL_BINDING_AUTO_PUSH_CONSTANT_BIT = 1 << 0,
// cbuffer resources will be declared as CBVs (b-space) without a register() declaration.
// A register will be automatically assigned, but must be reflected in D3D-land.
HLSL_BINDING_AUTO_CBV_BIT = 1 << 1,
// All SRVs (t-space) will be declared without a register() declaration.
HLSL_BINDING_AUTO_SRV_BIT = 1 << 2,
// All UAVs (u-space) will be declared without a register() declaration.
HLSL_BINDING_AUTO_UAV_BIT = 1 << 3,
// All samplers (s-space) will be declared without a register() declaration.
HLSL_BINDING_AUTO_SAMPLER_BIT = 1 << 4,
// No resources will be declared with register().
HLSL_BINDING_AUTO_ALL = 0x7fffffff
};
using HLSLBindingFlags = uint32_t;
class CompilerHLSL : public CompilerGLSL
{
public:
struct Options
{
uint32_t shader_model = 30; // TODO: map ps_4_0_level_9_0,... somehow
// Allows the PointSize builtin, and ignores it, as PointSize is not supported in HLSL.
bool point_size_compat = false;
// Allows the PointCoord builtin, returns float2(0.5, 0.5), as PointCoord is not supported in HLSL.
bool point_coord_compat = false;
// If true, the backend will assume that VertexIndex and InstanceIndex will need to apply
// a base offset, and you will need to fill in a cbuffer with offsets.
// Set to false if you know you will never use base instance or base vertex
// functionality as it might remove an internal cbuffer.
bool support_nonzero_base_vertex_base_instance = false;
};
explicit CompilerHLSL(std::vector<uint32_t> spirv_)
: CompilerGLSL(std::move(spirv_))
{
}
CompilerHLSL(const uint32_t *ir_, size_t size)
: CompilerGLSL(ir_, size)
{
}
explicit CompilerHLSL(const ParsedIR &ir_)
: CompilerGLSL(ir_)
{
}
explicit CompilerHLSL(ParsedIR &&ir_)
: CompilerGLSL(std::move(ir_))
{
}
const Options &get_hlsl_options() const
{
return hlsl_options;
}
void set_hlsl_options(const Options &opts)
{
hlsl_options = opts;
}
// Optionally specify a custom root constant layout.
//
// Push constants ranges will be split up according to the
// layout specified.
void set_root_constant_layouts(std::vector<RootConstants> layout);
// Compiles and remaps vertex attributes at specific locations to a fixed semantic.
// The default is TEXCOORD# where # denotes location.
// Matrices are unrolled to vectors with notation ${SEMANTIC}_#, where # denotes row.
// $SEMANTIC is either TEXCOORD# or a semantic name specified here.
void add_vertex_attribute_remap(const HLSLVertexAttributeRemap &vertex_attributes);
std::string compile() override;
// This is a special HLSL workaround for the NumWorkGroups builtin.
// This does not exist in HLSL, so the calling application must create a dummy cbuffer in
// which the application will store this builtin.
// The cbuffer layout will be:
// cbuffer SPIRV_Cross_NumWorkgroups : register(b#, space#) { uint3 SPIRV_Cross_NumWorkgroups_count; };
// This must be called before compile().
// The function returns 0 if NumWorkGroups builtin is not statically used in the shader from the current entry point.
// 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().
VariableID remap_num_workgroups_builtin();
// Controls how resource bindings are declared in the output HLSL.
void set_resource_binding_flags(HLSLBindingFlags flags);
private:
std::string type_to_glsl(const SPIRType &type, uint32_t id = 0) override;
std::string image_type_hlsl(const SPIRType &type, uint32_t id);
std::string image_type_hlsl_modern(const SPIRType &type, uint32_t id);
std::string image_type_hlsl_legacy(const SPIRType &type, uint32_t id);
void emit_function_prototype(SPIRFunction &func, const Bitset &return_flags) override;
void emit_hlsl_entry_point();
void emit_header() override;
void emit_resources();
void emit_interface_block_globally(const SPIRVariable &type);
void emit_interface_block_in_struct(const SPIRVariable &type, std::unordered_set<uint32_t> &active_locations);
void emit_builtin_inputs_in_struct();
void emit_builtin_outputs_in_struct();
void emit_texture_op(const Instruction &i) override;
void emit_instruction(const Instruction &instruction) override;
void emit_glsl_op(uint32_t result_type, uint32_t result_id, uint32_t op, const uint32_t *args,
uint32_t count) override;
void emit_buffer_block(const SPIRVariable &type) override;
void emit_push_constant_block(const SPIRVariable &var) override;
void emit_uniform(const SPIRVariable &var) override;
void emit_modern_uniform(const SPIRVariable &var);
void emit_legacy_uniform(const SPIRVariable &var);
void emit_specialization_constants_and_structs();
void emit_composite_constants();
void emit_fixup() override;
std::string builtin_to_glsl(spv::BuiltIn builtin, spv::StorageClass storage) override;
std::string layout_for_member(const SPIRType &type, uint32_t index) override;
std::string to_interpolation_qualifiers(const Bitset &flags) override;
std::string bitcast_glsl_op(const SPIRType &result_type, const SPIRType &argument_type) override;
std::string to_func_call_arg(const SPIRFunction::Parameter &arg, uint32_t id) override;
std::string to_sampler_expression(uint32_t id);
std::string to_resource_binding(const SPIRVariable &var);
std::string to_resource_binding_sampler(const SPIRVariable &var);
std::string to_resource_register(HLSLBindingFlags flags, char space, uint32_t binding, uint32_t set);
void emit_sampled_image_op(uint32_t result_type, uint32_t result_id, uint32_t image_id, uint32_t samp_id) override;
void emit_access_chain(const Instruction &instruction);
void emit_load(const Instruction &instruction);
void read_access_chain(std::string *expr, const std::string &lhs, const SPIRAccessChain &chain);
void read_access_chain_struct(const std::string &lhs, const SPIRAccessChain &chain);
void read_access_chain_array(const std::string &lhs, const SPIRAccessChain &chain);
void write_access_chain(const SPIRAccessChain &chain, uint32_t value, const SmallVector<uint32_t> &composite_chain);
void write_access_chain_struct(const SPIRAccessChain &chain, uint32_t value,
const SmallVector<uint32_t> &composite_chain);
void write_access_chain_array(const SPIRAccessChain &chain, uint32_t value,
const SmallVector<uint32_t> &composite_chain);
std::string write_access_chain_value(uint32_t value, const SmallVector<uint32_t> &composite_chain, bool enclose);
void emit_store(const Instruction &instruction);
void emit_atomic(const uint32_t *ops, uint32_t length, spv::Op op);
void emit_subgroup_op(const Instruction &i) override;
void emit_block_hints(const SPIRBlock &block) override;
void emit_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index, const std::string &qualifier,
uint32_t base_offset = 0) override;
const char *to_storage_qualifiers_glsl(const SPIRVariable &var) override;
void replace_illegal_names() override;
Options hlsl_options;
// TODO: Refactor this to be more similar to MSL, maybe have some common system in place?
bool requires_op_fmod = false;
bool requires_fp16_packing = false;
bool requires_explicit_fp16_packing = false;
bool requires_unorm8_packing = false;
bool requires_snorm8_packing = false;
bool requires_unorm16_packing = false;
bool requires_snorm16_packing = false;
bool requires_bitfield_insert = false;
bool requires_bitfield_extract = false;
bool requires_inverse_2x2 = false;
bool requires_inverse_3x3 = false;
bool requires_inverse_4x4 = false;
bool requires_scalar_reflect = false;
bool requires_scalar_refract = false;
bool requires_scalar_faceforward = false;
uint64_t required_textureSizeVariants = 0;
void require_texture_query_variant(const SPIRType &type);
enum TextureQueryVariantDim
{
Query1D = 0,
Query1DArray,
Query2D,
Query2DArray,
Query3D,
QueryBuffer,
QueryCube,
QueryCubeArray,
Query2DMS,
Query2DMSArray,
QueryDimCount
};
enum TextureQueryVariantType
{
QueryTypeFloat = 0,
QueryTypeInt = 16,
QueryTypeUInt = 32,
QueryTypeCount = 3
};
void emit_builtin_variables();
bool require_output = false;
bool require_input = false;
SmallVector<HLSLVertexAttributeRemap> remap_vertex_attributes;
uint32_t type_to_consumed_locations(const SPIRType &type) const;
void emit_io_block(const SPIRVariable &var);
std::string to_semantic(uint32_t location, spv::ExecutionModel em, spv::StorageClass sc);
uint32_t num_workgroups_builtin = 0;
HLSLBindingFlags resource_binding_flags = 0;
// Custom root constant layout, which should be emitted
// when translating push constant ranges.
std::vector<RootConstants> root_constants_layout;
void validate_shader_model();
std::string get_unique_identifier();
uint32_t unique_identifier_count = 0;
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
File diff suppressed because it is too large Load Diff
+945
View File
@@ -0,0 +1,945 @@
/*
* Copyright 2016-2019 The Brenwill Workshop Ltd.
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_MSL_HPP
#define SPIRV_CROSS_MSL_HPP
#include "spirv_glsl.hpp"
#include <map>
#include <set>
#include <stddef.h>
#include <unordered_map>
#include <unordered_set>
namespace SPIRV_CROSS_NAMESPACE
{
// Indicates the format of the vertex attribute. Currently limited to specifying
// if the attribute is an 8-bit unsigned integer, 16-bit unsigned integer, or
// some other format.
enum MSLVertexFormat
{
MSL_VERTEX_FORMAT_OTHER = 0,
MSL_VERTEX_FORMAT_UINT8 = 1,
MSL_VERTEX_FORMAT_UINT16 = 2,
MSL_VERTEX_FORMAT_INT_MAX = 0x7fffffff
};
// Defines MSL characteristics of a vertex attribute at a particular location.
// After compilation, it is possible to query whether or not this location was used.
struct MSLVertexAttr
{
uint32_t location = 0;
uint32_t msl_buffer = 0;
uint32_t msl_offset = 0;
uint32_t msl_stride = 0;
bool per_instance = false;
MSLVertexFormat format = MSL_VERTEX_FORMAT_OTHER;
spv::BuiltIn builtin = spv::BuiltInMax;
};
// Matches the binding index of a MSL resource for a binding within a descriptor set.
// Taken together, the stage, desc_set and binding combine to form a reference to a resource
// descriptor used in a particular shading stage.
// If using MSL 2.0 argument buffers, the descriptor set is not marked as a discrete descriptor set,
// and (for iOS only) the resource is not a storage image (sampled != 2), the binding reference we
// remap to will become an [[id(N)]] attribute within the "descriptor set" argument buffer structure.
// For resources which are bound in the "classic" MSL 1.0 way or discrete descriptors, the remap will become a
// [[buffer(N)]], [[texture(N)]] or [[sampler(N)]] depending on the resource types used.
struct MSLResourceBinding
{
spv::ExecutionModel stage = spv::ExecutionModelMax;
uint32_t desc_set = 0;
uint32_t binding = 0;
uint32_t msl_buffer = 0;
uint32_t msl_texture = 0;
uint32_t msl_sampler = 0;
};
enum MSLSamplerCoord
{
MSL_SAMPLER_COORD_NORMALIZED = 0,
MSL_SAMPLER_COORD_PIXEL = 1,
MSL_SAMPLER_INT_MAX = 0x7fffffff
};
enum MSLSamplerFilter
{
MSL_SAMPLER_FILTER_NEAREST = 0,
MSL_SAMPLER_FILTER_LINEAR = 1,
MSL_SAMPLER_FILTER_INT_MAX = 0x7fffffff
};
enum MSLSamplerMipFilter
{
MSL_SAMPLER_MIP_FILTER_NONE = 0,
MSL_SAMPLER_MIP_FILTER_NEAREST = 1,
MSL_SAMPLER_MIP_FILTER_LINEAR = 2,
MSL_SAMPLER_MIP_FILTER_INT_MAX = 0x7fffffff
};
enum MSLSamplerAddress
{
MSL_SAMPLER_ADDRESS_CLAMP_TO_ZERO = 0,
MSL_SAMPLER_ADDRESS_CLAMP_TO_EDGE = 1,
MSL_SAMPLER_ADDRESS_CLAMP_TO_BORDER = 2,
MSL_SAMPLER_ADDRESS_REPEAT = 3,
MSL_SAMPLER_ADDRESS_MIRRORED_REPEAT = 4,
MSL_SAMPLER_ADDRESS_INT_MAX = 0x7fffffff
};
enum MSLSamplerCompareFunc
{
MSL_SAMPLER_COMPARE_FUNC_NEVER = 0,
MSL_SAMPLER_COMPARE_FUNC_LESS = 1,
MSL_SAMPLER_COMPARE_FUNC_LESS_EQUAL = 2,
MSL_SAMPLER_COMPARE_FUNC_GREATER = 3,
MSL_SAMPLER_COMPARE_FUNC_GREATER_EQUAL = 4,
MSL_SAMPLER_COMPARE_FUNC_EQUAL = 5,
MSL_SAMPLER_COMPARE_FUNC_NOT_EQUAL = 6,
MSL_SAMPLER_COMPARE_FUNC_ALWAYS = 7,
MSL_SAMPLER_COMPARE_FUNC_INT_MAX = 0x7fffffff
};
enum MSLSamplerBorderColor
{
MSL_SAMPLER_BORDER_COLOR_TRANSPARENT_BLACK = 0,
MSL_SAMPLER_BORDER_COLOR_OPAQUE_BLACK = 1,
MSL_SAMPLER_BORDER_COLOR_OPAQUE_WHITE = 2,
MSL_SAMPLER_BORDER_COLOR_INT_MAX = 0x7fffffff
};
enum MSLFormatResolution
{
MSL_FORMAT_RESOLUTION_444 = 0,
MSL_FORMAT_RESOLUTION_422,
MSL_FORMAT_RESOLUTION_420,
MSL_FORMAT_RESOLUTION_INT_MAX = 0x7fffffff
};
enum MSLChromaLocation
{
MSL_CHROMA_LOCATION_COSITED_EVEN = 0,
MSL_CHROMA_LOCATION_MIDPOINT,
MSL_CHROMA_LOCATION_INT_MAX = 0x7fffffff
};
enum MSLComponentSwizzle
{
MSL_COMPONENT_SWIZZLE_IDENTITY = 0,
MSL_COMPONENT_SWIZZLE_ZERO,
MSL_COMPONENT_SWIZZLE_ONE,
MSL_COMPONENT_SWIZZLE_R,
MSL_COMPONENT_SWIZZLE_G,
MSL_COMPONENT_SWIZZLE_B,
MSL_COMPONENT_SWIZZLE_A,
MSL_COMPONENT_SWIZZLE_INT_MAX = 0x7fffffff
};
enum MSLSamplerYCbCrModelConversion
{
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_RGB_IDENTITY = 0,
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_IDENTITY,
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_709,
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_601,
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_YCBCR_BT_2020,
MSL_SAMPLER_YCBCR_MODEL_CONVERSION_INT_MAX = 0x7fffffff
};
enum MSLSamplerYCbCrRange
{
MSL_SAMPLER_YCBCR_RANGE_ITU_FULL = 0,
MSL_SAMPLER_YCBCR_RANGE_ITU_NARROW,
MSL_SAMPLER_YCBCR_RANGE_INT_MAX = 0x7fffffff
};
struct MSLConstexprSampler
{
MSLSamplerCoord coord = MSL_SAMPLER_COORD_NORMALIZED;
MSLSamplerFilter min_filter = MSL_SAMPLER_FILTER_NEAREST;
MSLSamplerFilter mag_filter = MSL_SAMPLER_FILTER_NEAREST;
MSLSamplerMipFilter mip_filter = MSL_SAMPLER_MIP_FILTER_NONE;
MSLSamplerAddress s_address = MSL_SAMPLER_ADDRESS_CLAMP_TO_EDGE;
MSLSamplerAddress t_address = MSL_SAMPLER_ADDRESS_CLAMP_TO_EDGE;
MSLSamplerAddress r_address = MSL_SAMPLER_ADDRESS_CLAMP_TO_EDGE;
MSLSamplerCompareFunc compare_func = MSL_SAMPLER_COMPARE_FUNC_NEVER;
MSLSamplerBorderColor border_color = MSL_SAMPLER_BORDER_COLOR_TRANSPARENT_BLACK;
float lod_clamp_min = 0.0f;
float lod_clamp_max = 1000.0f;
int max_anisotropy = 1;
// Sampler Y'CbCr conversion parameters
uint32_t planes = 0;
MSLFormatResolution resolution = MSL_FORMAT_RESOLUTION_444;
MSLSamplerFilter chroma_filter = MSL_SAMPLER_FILTER_NEAREST;
MSLChromaLocation x_chroma_offset = MSL_CHROMA_LOCATION_COSITED_EVEN;
MSLChromaLocation y_chroma_offset = MSL_CHROMA_LOCATION_COSITED_EVEN;
MSLComponentSwizzle swizzle[4]; // IDENTITY, IDENTITY, IDENTITY, IDENTITY
MSLSamplerYCbCrModelConversion ycbcr_model = MSL_SAMPLER_YCBCR_MODEL_CONVERSION_RGB_IDENTITY;
MSLSamplerYCbCrRange ycbcr_range = MSL_SAMPLER_YCBCR_RANGE_ITU_FULL;
uint32_t bpc = 8;
bool compare_enable = false;
bool lod_clamp_enable = false;
bool anisotropy_enable = false;
bool ycbcr_conversion_enable = false;
MSLConstexprSampler()
{
for (uint32_t i = 0; i < 4; i++)
swizzle[i] = MSL_COMPONENT_SWIZZLE_IDENTITY;
}
bool swizzle_is_identity() const
{
return (swizzle[0] == MSL_COMPONENT_SWIZZLE_IDENTITY && swizzle[1] == MSL_COMPONENT_SWIZZLE_IDENTITY &&
swizzle[2] == MSL_COMPONENT_SWIZZLE_IDENTITY && swizzle[3] == MSL_COMPONENT_SWIZZLE_IDENTITY);
}
bool swizzle_has_one_or_zero() const
{
return (swizzle[0] == MSL_COMPONENT_SWIZZLE_ZERO || swizzle[0] == MSL_COMPONENT_SWIZZLE_ONE ||
swizzle[1] == MSL_COMPONENT_SWIZZLE_ZERO || swizzle[1] == MSL_COMPONENT_SWIZZLE_ONE ||
swizzle[2] == MSL_COMPONENT_SWIZZLE_ZERO || swizzle[2] == MSL_COMPONENT_SWIZZLE_ONE ||
swizzle[3] == MSL_COMPONENT_SWIZZLE_ZERO || swizzle[3] == MSL_COMPONENT_SWIZZLE_ONE);
}
};
// Special constant used in a MSLResourceBinding desc_set
// element to indicate the bindings for the push constants.
static const uint32_t kPushConstDescSet = ~(0u);
// Special constant used in a MSLResourceBinding binding
// element to indicate the bindings for the push constants.
static const uint32_t kPushConstBinding = 0;
// Special constant used in a MSLResourceBinding binding
// element to indicate the buffer binding for swizzle buffers.
static const uint32_t kSwizzleBufferBinding = ~(1u);
// Special constant used in a MSLResourceBinding binding
// element to indicate the buffer binding for buffer size buffers to support OpArrayLength.
static const uint32_t kBufferSizeBufferBinding = ~(2u);
// Special constant used in a MSLResourceBinding binding
// element to indicate the buffer binding used for the argument buffer itself.
// This buffer binding should be kept as small as possible as all automatic bindings for buffers
// will start at max(kArgumentBufferBinding) + 1.
static const uint32_t kArgumentBufferBinding = ~(3u);
static const uint32_t kMaxArgumentBuffers = 8;
// Decompiles SPIR-V to Metal Shading Language
class CompilerMSL : public CompilerGLSL
{
public:
// Options for compiling to Metal Shading Language
struct Options
{
typedef enum
{
iOS = 0,
macOS = 1
} Platform;
Platform platform = macOS;
uint32_t msl_version = make_msl_version(1, 2);
uint32_t texel_buffer_texture_width = 4096; // Width of 2D Metal textures used as 1D texel buffers
uint32_t swizzle_buffer_index = 30;
uint32_t indirect_params_buffer_index = 29;
uint32_t shader_output_buffer_index = 28;
uint32_t shader_patch_output_buffer_index = 27;
uint32_t shader_tess_factor_buffer_index = 26;
uint32_t buffer_size_buffer_index = 25;
uint32_t view_mask_buffer_index = 24;
uint32_t dynamic_offsets_buffer_index = 23;
uint32_t shader_input_wg_index = 0;
uint32_t device_index = 0;
bool enable_point_size_builtin = true;
bool disable_rasterization = false;
bool capture_output_to_buffer = false;
bool swizzle_texture_samples = false;
bool tess_domain_origin_lower_left = false;
bool multiview = false;
bool view_index_from_device_index = false;
bool dispatch_base = false;
bool texture_1D_as_2D = false;
// Enable use of MSL 2.0 indirect argument buffers.
// MSL 2.0 must also be enabled.
bool argument_buffers = false;
// Ensures vertex and instance indices start at zero. This reflects the behavior of HLSL with SV_VertexID and SV_InstanceID.
bool enable_base_index_zero = false;
// Fragment output in MSL must have at least as many components as the render pass.
// Add support to explicit pad out components.
bool pad_fragment_output_components = false;
// Specifies whether the iOS target version supports the [[base_vertex]] and [[base_instance]] attributes.
bool ios_support_base_vertex_instance = false;
// Use Metal's native frame-buffer fetch API for subpass inputs.
bool ios_use_framebuffer_fetch_subpasses = false;
// Enables use of "fma" intrinsic for invariant float math
bool invariant_float_math = false;
// Emulate texturecube_array with texture2d_array for iOS where this type is not available
bool emulate_cube_array = false;
// Allow user to enable decoration binding
bool enable_decoration_binding = false;
// Requires MSL 2.1, use the native support for texel buffers.
bool texture_buffer_native = false;
bool is_ios()
{
return platform == iOS;
}
bool is_macos()
{
return platform == macOS;
}
void set_msl_version(uint32_t major, uint32_t minor = 0, uint32_t patch = 0)
{
msl_version = make_msl_version(major, minor, patch);
}
bool supports_msl_version(uint32_t major, uint32_t minor = 0, uint32_t patch = 0) const
{
return msl_version >= make_msl_version(major, minor, patch);
}
static uint32_t make_msl_version(uint32_t major, uint32_t minor = 0, uint32_t patch = 0)
{
return (major * 10000) + (minor * 100) + patch;
}
};
const Options &get_msl_options() const
{
return msl_options;
}
void set_msl_options(const Options &opts)
{
msl_options = opts;
}
// Provide feedback to calling API to allow runtime to disable pipeline
// rasterization if vertex shader requires rasterization to be disabled.
bool get_is_rasterization_disabled() const
{
return is_rasterization_disabled && (get_entry_point().model == spv::ExecutionModelVertex ||
get_entry_point().model == spv::ExecutionModelTessellationControl ||
get_entry_point().model == spv::ExecutionModelTessellationEvaluation);
}
// Provide feedback to calling API to allow it to pass an auxiliary
// swizzle buffer if the shader needs it.
bool needs_swizzle_buffer() const
{
return used_swizzle_buffer;
}
// Provide feedback to calling API to allow it to pass a buffer
// containing STORAGE_BUFFER buffer sizes to support OpArrayLength.
bool needs_buffer_size_buffer() const
{
return !buffers_requiring_array_length.empty();
}
// Provide feedback to calling API to allow it to pass a buffer
// containing the view mask for the current multiview subpass.
bool needs_view_mask_buffer() const
{
return msl_options.multiview && !msl_options.view_index_from_device_index;
}
// Provide feedback to calling API to allow it to pass a buffer
// containing the dispatch base workgroup ID.
bool needs_dispatch_base_buffer() const
{
return msl_options.dispatch_base && !msl_options.supports_msl_version(1, 2);
}
// Provide feedback to calling API to allow it to pass an output
// buffer if the shader needs it.
bool needs_output_buffer() const
{
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 != 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 != ID(0);
}
explicit CompilerMSL(std::vector<uint32_t> spirv);
CompilerMSL(const uint32_t *ir, size_t word_count);
explicit CompilerMSL(const ParsedIR &ir);
explicit CompilerMSL(ParsedIR &&ir);
// attr is a vertex attribute binding used to match
// vertex content locations to MSL attributes. If vertex attributes are provided,
// is_msl_vertex_attribute_used() will return true after calling ::compile() if
// the location was used by the MSL code.
void add_msl_vertex_attribute(const MSLVertexAttr &attr);
// resource is a resource binding to indicate the MSL buffer,
// texture or sampler index to use for a particular SPIR-V description set
// and binding. If resource bindings are provided,
// is_msl_resource_binding_used() will return true after calling ::compile() if
// the set/binding combination was used by the MSL code.
void add_msl_resource_binding(const MSLResourceBinding &resource);
// desc_set and binding are the SPIR-V descriptor set and binding of a buffer resource
// in this shader. index is the index within the dynamic offset buffer to use. This
// function marks that resource as using a dynamic offset (VK_DESCRIPTOR_TYPE_UNIFORM_BUFFER_DYNAMIC
// or VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC). This function only has any effect if argument buffers
// are enabled. If so, the buffer will have its address adjusted at the beginning of the shader with
// an offset taken from the dynamic offset buffer.
void add_dynamic_buffer(uint32_t desc_set, uint32_t binding, uint32_t index);
// When using MSL argument buffers, we can force "classic" MSL 1.0 binding schemes for certain descriptor sets.
// This corresponds to VK_KHR_push_descriptor in Vulkan.
void add_discrete_descriptor_set(uint32_t desc_set);
// If an argument buffer is large enough, it may need to be in the device storage space rather than
// constant. Opt-in to this behavior here on a per set basis.
void set_argument_buffer_device_address_space(uint32_t desc_set, bool device_storage);
// Query after compilation is done. This allows you to check if a location or set/binding combination was used by the shader.
bool is_msl_vertex_attribute_used(uint32_t location);
// NOTE: Only resources which are remapped using add_msl_resource_binding will be reported here.
// Constexpr samplers are always assumed to be emitted.
// No specific MSLResourceBinding remapping is required for constexpr samplers as long as they are remapped
// by remap_constexpr_sampler(_by_binding).
bool is_msl_resource_binding_used(spv::ExecutionModel model, uint32_t set, uint32_t binding);
// This must only be called after a successful call to CompilerMSL::compile().
// For a variable resource ID obtained through reflection API, report the automatically assigned resource index.
// If the descriptor set was part of an argument buffer, report the [[id(N)]],
// or [[buffer/texture/sampler]] binding for other resources.
// If the resource was a combined image sampler, report the image binding here,
// use the _secondary version of this call to query the sampler half of the resource.
// If no binding exists, uint32_t(-1) is returned.
uint32_t get_automatic_msl_resource_binding(uint32_t id) const;
// Same as get_automatic_msl_resource_binding, but should only be used for combined image samplers, in which case the
// sampler's binding is returned instead. For any other resource type, -1 is returned.
uint32_t get_automatic_msl_resource_binding_secondary(uint32_t id) const;
// Same as get_automatic_msl_resource_binding, but should only be used for combined image samplers for multiplanar images,
// in which case the second plane's binding is returned instead. For any other resource type, -1 is returned.
uint32_t get_automatic_msl_resource_binding_tertiary(uint32_t id) const;
// Same as get_automatic_msl_resource_binding, but should only be used for combined image samplers for triplanar images,
// in which case the third plane's binding is returned instead. For any other resource type, -1 is returned.
uint32_t get_automatic_msl_resource_binding_quaternary(uint32_t id) const;
// Compiles the SPIR-V code into Metal Shading Language.
std::string compile() override;
// Remap a sampler with ID to a constexpr sampler.
// Older iOS targets must use constexpr samplers in certain cases (PCF),
// so a static sampler must be used.
// The sampler will not consume a binding, but be declared in the entry point as a constexpr sampler.
// 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(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.
void remap_constexpr_sampler_by_binding(uint32_t desc_set, uint32_t binding, const MSLConstexprSampler &sampler);
// If using CompilerMSL::Options::pad_fragment_output_components, override the number of components we expect
// to use for a particular location. The default is 4 if number of components is not overridden.
void set_fragment_output_components(uint32_t location, uint32_t components);
protected:
// An enum of SPIR-V functions that are implemented in additional
// source code that is added to the shader if necessary.
enum SPVFuncImpl
{
SPVFuncImplNone,
SPVFuncImplMod,
SPVFuncImplRadians,
SPVFuncImplDegrees,
SPVFuncImplFindILsb,
SPVFuncImplFindSMsb,
SPVFuncImplFindUMsb,
SPVFuncImplSSign,
SPVFuncImplArrayCopyMultidimBase,
// Unfortunately, we cannot use recursive templates in the MSL compiler properly,
// so stamp out variants up to some arbitrary maximum.
SPVFuncImplArrayCopy = SPVFuncImplArrayCopyMultidimBase + 1,
SPVFuncImplArrayOfArrayCopy2Dim = SPVFuncImplArrayCopyMultidimBase + 2,
SPVFuncImplArrayOfArrayCopy3Dim = SPVFuncImplArrayCopyMultidimBase + 3,
SPVFuncImplArrayOfArrayCopy4Dim = SPVFuncImplArrayCopyMultidimBase + 4,
SPVFuncImplArrayOfArrayCopy5Dim = SPVFuncImplArrayCopyMultidimBase + 5,
SPVFuncImplArrayOfArrayCopy6Dim = SPVFuncImplArrayCopyMultidimBase + 6,
SPVFuncImplTexelBufferCoords,
SPVFuncImplImage2DAtomicCoords, // Emulate texture2D atomic operations
SPVFuncImplFMul,
SPVFuncImplFAdd,
SPVFuncImplCubemapTo2DArrayFace,
SPVFuncImplUnsafeArray, // Allow Metal to use the array<T> template to make arrays a value type
SPVFuncImplInverse4x4,
SPVFuncImplInverse3x3,
SPVFuncImplInverse2x2,
// It is very important that this come before *Swizzle and ChromaReconstruct*, to ensure it's
// emitted before them.
SPVFuncImplForwardArgs,
// Likewise, this must come before *Swizzle.
SPVFuncImplGetSwizzle,
SPVFuncImplTextureSwizzle,
SPVFuncImplGatherSwizzle,
SPVFuncImplGatherCompareSwizzle,
SPVFuncImplSubgroupBallot,
SPVFuncImplSubgroupBallotBitExtract,
SPVFuncImplSubgroupBallotFindLSB,
SPVFuncImplSubgroupBallotFindMSB,
SPVFuncImplSubgroupBallotBitCount,
SPVFuncImplSubgroupAllEqual,
SPVFuncImplReflectScalar,
SPVFuncImplRefractScalar,
SPVFuncImplFaceForwardScalar,
SPVFuncImplChromaReconstructNearest2Plane,
SPVFuncImplChromaReconstructNearest3Plane,
SPVFuncImplChromaReconstructLinear422CositedEven2Plane,
SPVFuncImplChromaReconstructLinear422CositedEven3Plane,
SPVFuncImplChromaReconstructLinear422Midpoint2Plane,
SPVFuncImplChromaReconstructLinear422Midpoint3Plane,
SPVFuncImplChromaReconstructLinear420XCositedEvenYCositedEven2Plane,
SPVFuncImplChromaReconstructLinear420XCositedEvenYCositedEven3Plane,
SPVFuncImplChromaReconstructLinear420XMidpointYCositedEven2Plane,
SPVFuncImplChromaReconstructLinear420XMidpointYCositedEven3Plane,
SPVFuncImplChromaReconstructLinear420XCositedEvenYMidpoint2Plane,
SPVFuncImplChromaReconstructLinear420XCositedEvenYMidpoint3Plane,
SPVFuncImplChromaReconstructLinear420XMidpointYMidpoint2Plane,
SPVFuncImplChromaReconstructLinear420XMidpointYMidpoint3Plane,
SPVFuncImplExpandITUFullRange,
SPVFuncImplExpandITUNarrowRange,
SPVFuncImplConvertYCbCrBT709,
SPVFuncImplConvertYCbCrBT601,
SPVFuncImplConvertYCbCrBT2020,
SPVFuncImplDynamicImageSampler,
SPVFuncImplArrayCopyMultidimMax = 6
};
// If the underlying resource has been used for comparison then duplicate loads of that resource must be too
// Use Metal's native frame-buffer fetch API for subpass inputs.
void emit_texture_op(const Instruction &i) override;
void emit_binary_unord_op(uint32_t result_type, uint32_t result_id, uint32_t op0, uint32_t op1, const char *op);
void emit_instruction(const Instruction &instr) override;
void emit_glsl_op(uint32_t result_type, uint32_t result_id, uint32_t op, const uint32_t *args,
uint32_t count) override;
void emit_spv_amd_shader_trinary_minmax_op(uint32_t result_type, uint32_t result_id, uint32_t op,
const uint32_t *args, uint32_t count) override;
void emit_header() override;
void emit_function_prototype(SPIRFunction &func, const Bitset &return_flags) override;
void emit_sampled_image_op(uint32_t result_type, uint32_t result_id, uint32_t image_id, uint32_t samp_id) override;
void emit_subgroup_op(const Instruction &i) override;
std::string to_texture_op(const Instruction &i, bool *forward,
SmallVector<uint32_t> &inherited_expressions) override;
void emit_fixup() override;
std::string to_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index,
const std::string &qualifier = "");
void emit_struct_member(const SPIRType &type, uint32_t member_type_id, uint32_t index,
const std::string &qualifier = "", uint32_t base_offset = 0) override;
void emit_struct_padding_target(const SPIRType &type) override;
std::string type_to_glsl(const SPIRType &type, uint32_t id = 0) override;
// Allow Metal to use the array<T> template to make arrays a value type
std::string type_to_array_glsl(const SPIRType &type) override;
// Threadgroup arrays can't have a wrapper type
std::string variable_decl(const SPIRVariable &variable) override;
// GCC workaround of lambdas calling protected functions (for older GCC versions)
std::string variable_decl(const SPIRType &type, const std::string &name, uint32_t id = 0) override;
std::string image_type_glsl(const SPIRType &type, uint32_t id = 0) override;
std::string sampler_type(const SPIRType &type);
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(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(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;
std::string to_initializer_expression(const SPIRVariable &var) override;
std::string unpack_expression_type(std::string expr_str, const SPIRType &type, uint32_t physical_type_id,
bool is_packed, bool row_major) override;
// Returns true for BuiltInSampleMask because gl_SampleMask[] is an array in SPIR-V, but [[sample_mask]] is a scalar in Metal.
bool builtin_translates_to_nonarray(spv::BuiltIn builtin) const override;
std::string bitcast_glsl_op(const SPIRType &result_type, const SPIRType &argument_type) override;
bool skip_argument(uint32_t id) const override;
std::string to_member_reference(uint32_t base, const SPIRType &type, uint32_t index, bool ptr_chain) override;
std::string to_qualifiers_glsl(uint32_t id) override;
void replace_illegal_names() override;
void declare_undefined_values() override;
void declare_constant_arrays();
// Constant arrays of non-primitive types (i.e. matrices) won't link properly into Metal libraries
void declare_complex_constant_arrays();
bool is_patch_block(const SPIRType &type);
bool is_non_native_row_major_matrix(uint32_t id) override;
bool member_is_non_native_row_major_matrix(const SPIRType &type, uint32_t index) override;
std::string convert_row_major_matrix(std::string exp_str, const SPIRType &exp_type, uint32_t physical_type_id,
bool is_packed) override;
void preprocess_op_codes();
void localize_global_variables();
void extract_global_variables_from_functions();
void mark_packable_structs();
void mark_as_packable(SPIRType &type);
std::unordered_map<uint32_t, std::set<uint32_t>> function_global_vars;
void extract_global_variables_from_function(uint32_t func_id, std::set<uint32_t> &added_arg_ids,
std::unordered_set<uint32_t> &global_var_ids,
std::unordered_set<uint32_t> &processed_func_ids);
uint32_t add_interface_block(spv::StorageClass storage, bool patch = false);
uint32_t add_interface_block_pointer(uint32_t ib_var_id, spv::StorageClass storage);
struct InterfaceBlockMeta
{
struct LocationMeta
{
uint32_t num_components = 0;
uint32_t ib_index = ~0u;
};
std::unordered_map<uint32_t, LocationMeta> location_meta;
bool strip_array = false;
};
void add_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref, SPIRType &ib_type,
SPIRVariable &var, InterfaceBlockMeta &meta);
void add_composite_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
SPIRType &ib_type, SPIRVariable &var, InterfaceBlockMeta &meta);
void add_plain_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
SPIRType &ib_type, SPIRVariable &var, InterfaceBlockMeta &meta);
void add_plain_member_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
SPIRType &ib_type, SPIRVariable &var, uint32_t index,
InterfaceBlockMeta &meta);
void add_composite_member_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
SPIRType &ib_type, SPIRVariable &var, uint32_t index,
InterfaceBlockMeta &meta);
uint32_t get_accumulated_member_location(const SPIRVariable &var, uint32_t mbr_idx, bool strip_array);
void add_tess_level_input_to_interface_block(const std::string &ib_var_ref, SPIRType &ib_type, SPIRVariable &var);
void fix_up_interface_member_indices(spv::StorageClass storage, uint32_t ib_type_id);
void mark_location_as_used_by_shader(uint32_t location, spv::StorageClass storage);
uint32_t ensure_correct_builtin_type(uint32_t type_id, spv::BuiltIn builtin);
uint32_t ensure_correct_attribute_type(uint32_t type_id, uint32_t location, uint32_t num_components = 0);
void emit_custom_templates();
void emit_custom_functions();
void emit_resources();
void emit_specialization_constants_and_structs();
void emit_interface_block(uint32_t ib_var_id);
bool maybe_emit_array_assignment(uint32_t id_lhs, uint32_t id_rhs);
void fix_up_shader_inputs_outputs();
std::string func_type_decl(SPIRType &type);
std::string entry_point_args_classic(bool append_comma);
std::string entry_point_args_argument_buffer(bool append_comma);
std::string entry_point_arg_stage_in();
void entry_point_args_builtin(std::string &args);
void entry_point_args_discrete_descriptors(std::string &args);
std::string to_qualified_member_name(const SPIRType &type, uint32_t index);
std::string ensure_valid_name(std::string name, std::string pfx);
std::string to_sampler_expression(uint32_t id);
std::string to_swizzle_expression(uint32_t id);
std::string to_buffer_size_expression(uint32_t id);
std::string builtin_qualifier(spv::BuiltIn builtin);
std::string builtin_type_decl(spv::BuiltIn builtin, uint32_t id = 0);
std::string built_in_func_arg(spv::BuiltIn builtin, bool prefix_comma);
std::string member_attribute_qualifier(const SPIRType &type, uint32_t index);
std::string argument_decl(const SPIRFunction::Parameter &arg);
std::string round_fp_tex_coords(std::string tex_coords, bool coord_is_fp);
uint32_t get_metal_resource_index(SPIRVariable &var, SPIRType::BaseType basetype, uint32_t plane = 0);
uint32_t get_ordered_member_location(uint32_t type_id, uint32_t index, uint32_t *comp = nullptr);
// MSL packing rules. These compute the effective packing rules as observed by the MSL compiler in the MSL output.
// These values can change depending on various extended decorations which control packing rules.
// We need to make these rules match up with SPIR-V declared rules.
uint32_t get_declared_type_size_msl(const SPIRType &type, bool packed, bool row_major) const;
uint32_t get_declared_type_array_stride_msl(const SPIRType &type, bool packed, bool row_major) const;
uint32_t get_declared_type_matrix_stride_msl(const SPIRType &type, bool packed, bool row_major) const;
uint32_t get_declared_type_alignment_msl(const SPIRType &type, bool packed, bool row_major) const;
uint32_t get_declared_struct_member_size_msl(const SPIRType &struct_type, uint32_t index) const;
uint32_t get_declared_struct_member_array_stride_msl(const SPIRType &struct_type, uint32_t index) const;
uint32_t get_declared_struct_member_matrix_stride_msl(const SPIRType &struct_type, uint32_t index) const;
uint32_t get_declared_struct_member_alignment_msl(const SPIRType &struct_type, uint32_t index) const;
const SPIRType &get_physical_member_type(const SPIRType &struct_type, uint32_t index) const;
uint32_t get_declared_struct_size_msl(const SPIRType &struct_type, bool ignore_alignment = false,
bool ignore_padding = false) const;
std::string to_component_argument(uint32_t id);
void align_struct(SPIRType &ib_type, std::unordered_set<uint32_t> &aligned_structs);
void mark_scalar_layout_structs(const SPIRType &ib_type);
void mark_struct_members_packed(const SPIRType &type);
void ensure_member_packing_rules_msl(SPIRType &ib_type, uint32_t index);
bool validate_member_packing_rules_msl(const SPIRType &type, uint32_t index) const;
std::string get_argument_address_space(const SPIRVariable &argument);
std::string get_type_address_space(const SPIRType &type, uint32_t id, bool argument = false);
const char *to_restrict(uint32_t id, bool space = true);
SPIRType &get_stage_in_struct_type();
SPIRType &get_stage_out_struct_type();
SPIRType &get_patch_stage_in_struct_type();
SPIRType &get_patch_stage_out_struct_type();
std::string get_tess_factor_struct_name();
void emit_atomic_func_op(uint32_t result_type, uint32_t result_id, const char *op, uint32_t mem_order_1,
uint32_t mem_order_2, bool has_mem_order_2, uint32_t op0, uint32_t op1 = 0,
bool op1_is_pointer = false, bool op1_is_literal = false, uint32_t op2 = 0);
const char *get_memory_order(uint32_t spv_mem_sem);
void add_pragma_line(const std::string &line);
void add_typedef_line(const std::string &line);
void emit_barrier(uint32_t id_exe_scope, uint32_t id_mem_scope, uint32_t id_mem_sem);
void emit_array_copy(const std::string &lhs, uint32_t rhs_id, spv::StorageClass lhs_storage,
spv::StorageClass rhs_storage) override;
void build_implicit_builtins();
uint32_t build_constant_uint_array_pointer();
void emit_entry_point_declarations() override;
uint32_t builtin_frag_coord_id = 0;
uint32_t builtin_sample_id_id = 0;
uint32_t builtin_vertex_idx_id = 0;
uint32_t builtin_base_vertex_id = 0;
uint32_t builtin_instance_idx_id = 0;
uint32_t builtin_base_instance_id = 0;
uint32_t builtin_view_idx_id = 0;
uint32_t builtin_layer_id = 0;
uint32_t builtin_invocation_id_id = 0;
uint32_t builtin_primitive_id_id = 0;
uint32_t builtin_subgroup_invocation_id_id = 0;
uint32_t builtin_subgroup_size_id = 0;
uint32_t builtin_dispatch_base_id = 0;
uint32_t swizzle_buffer_id = 0;
uint32_t buffer_size_buffer_id = 0;
uint32_t view_mask_buffer_id = 0;
uint32_t dynamic_offsets_buffer_id = 0;
void bitcast_to_builtin_store(uint32_t target_id, std::string &expr, const SPIRType &expr_type) override;
void bitcast_from_builtin_load(uint32_t source_id, std::string &expr, const SPIRType &expr_type) override;
void emit_store_statement(uint32_t lhs_expression, uint32_t rhs_expression) override;
void analyze_sampled_image_usage();
bool emit_tessellation_access_chain(const uint32_t *ops, uint32_t length);
bool emit_tessellation_io_load(uint32_t result_type, uint32_t id, uint32_t ptr);
bool is_out_of_bounds_tessellation_level(uint32_t id_lhs);
void ensure_builtin(spv::StorageClass storage, spv::BuiltIn builtin);
void mark_implicit_builtin(spv::StorageClass storage, spv::BuiltIn builtin, uint32_t id);
std::string convert_to_f32(const std::string &expr, uint32_t components);
Options msl_options;
std::set<SPVFuncImpl> spv_function_implementations;
std::unordered_map<uint32_t, MSLVertexAttr> vtx_attrs_by_location;
std::unordered_map<uint32_t, MSLVertexAttr> vtx_attrs_by_builtin;
std::unordered_set<uint32_t> vtx_attrs_in_use;
std::unordered_map<uint32_t, uint32_t> fragment_output_components;
std::set<std::string> pragma_lines;
std::set<std::string> typedef_lines;
SmallVector<uint32_t> vars_needing_early_declaration;
struct SetBindingPair
{
uint32_t desc_set;
uint32_t binding;
bool operator==(const SetBindingPair &other) const;
bool operator<(const SetBindingPair &other) const;
};
struct StageSetBinding
{
spv::ExecutionModel model;
uint32_t desc_set;
uint32_t binding;
bool operator==(const StageSetBinding &other) const;
};
struct InternalHasher
{
size_t operator()(const SetBindingPair &value) const;
size_t operator()(const StageSetBinding &value) const;
};
std::unordered_map<StageSetBinding, std::pair<MSLResourceBinding, bool>, InternalHasher> resource_bindings;
uint32_t next_metal_resource_index_buffer = 0;
uint32_t next_metal_resource_index_texture = 0;
uint32_t next_metal_resource_index_sampler = 0;
// Intentionally uninitialized, works around MSVC 2013 bug.
uint32_t next_metal_resource_ids[kMaxArgumentBuffers];
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;
// Handle HLSL-style 0-based vertex/instance index.
enum class TriState
{
Neutral,
No,
Yes
};
TriState needs_base_vertex_arg = TriState::Neutral;
TriState needs_base_instance_arg = TriState::Neutral;
bool has_sampled_images = false;
bool builtin_declaration = false; // Handle HLSL-style 0-based vertex/instance index.
bool use_builtin_array = false; // Force the use of C style array declaration.
bool is_rasterization_disabled = false;
bool capture_output_to_buffer = false;
bool needs_swizzle_buffer_def = false;
bool used_swizzle_buffer = false;
bool added_builtin_tess_level = false;
bool needs_subgroup_invocation_id = false;
std::string qual_pos_var_name;
std::string stage_in_var_name = "in";
std::string stage_out_var_name = "out";
std::string patch_stage_in_var_name = "patchIn";
std::string patch_stage_out_var_name = "patchOut";
std::string sampler_name_suffix = "Smplr";
std::string swizzle_name_suffix = "Swzl";
std::string buffer_size_name_suffix = "BufferSize";
std::string plane_name_suffix = "Plane";
std::string input_wg_var_name = "gl_in";
std::string output_buffer_var_name = "spvOut";
std::string patch_output_buffer_var_name = "spvPatchOut";
std::string tess_factor_buffer_var_name = "spvTessLevel";
spv::Op previous_instruction_opcode = spv::OpNop;
// Must be ordered since declaration is in a specific order.
std::map<uint32_t, MSLConstexprSampler> constexpr_samplers_by_id;
std::unordered_map<SetBindingPair, MSLConstexprSampler, InternalHasher> constexpr_samplers_by_binding;
const MSLConstexprSampler *find_constexpr_sampler(uint32_t id) const;
std::unordered_set<uint32_t> buffers_requiring_array_length;
SmallVector<uint32_t> buffer_arrays;
std::unordered_set<uint32_t> atomic_image_vars; // Emulate texture2D atomic operations
// Must be ordered since array is in a specific order.
std::map<SetBindingPair, std::pair<uint32_t, uint32_t>> buffers_requiring_dynamic_offset;
uint32_t argument_buffer_ids[kMaxArgumentBuffers];
uint32_t argument_buffer_discrete_mask = 0;
uint32_t argument_buffer_device_storage_mask = 0;
void analyze_argument_buffers();
bool descriptor_set_is_argument_buffer(uint32_t desc_set) const;
uint32_t get_target_components_for_fragment_location(uint32_t location) const;
uint32_t build_extended_vector_type(uint32_t type_id, uint32_t components);
bool suppress_missing_prototypes = false;
void add_spv_func_and_recompile(SPVFuncImpl spv_func);
// OpcodeHandler that handles several MSL preprocessing operations.
struct OpCodePreprocessor : OpcodeHandler
{
OpCodePreprocessor(CompilerMSL &compiler_)
: compiler(compiler_)
{
}
bool handle(spv::Op opcode, const uint32_t *args, uint32_t length) override;
CompilerMSL::SPVFuncImpl get_spv_func_impl(spv::Op opcode, const uint32_t *args);
void check_resource_write(uint32_t var_id);
CompilerMSL &compiler;
std::unordered_map<uint32_t, uint32_t> result_types;
std::unordered_map<uint32_t, uint32_t> image_pointers; // Emulate texture2D atomic operations
bool suppress_missing_prototypes = false;
bool uses_atomics = false;
bool uses_resource_write = false;
bool needs_subgroup_invocation_id = false;
};
// OpcodeHandler that scans for uses of sampled images
struct SampledImageScanner : OpcodeHandler
{
SampledImageScanner(CompilerMSL &compiler_)
: compiler(compiler_)
{
}
bool handle(spv::Op opcode, const uint32_t *args, uint32_t) override;
CompilerMSL &compiler;
};
// Sorts the members of a SPIRType and associated Meta info based on a settable sorting
// aspect, which defines which aspect of the struct members will be used to sort them.
// Regardless of the sorting aspect, built-in members always appear at the end of the struct.
struct MemberSorter
{
enum SortAspect
{
Location,
LocationReverse,
Offset,
OffsetThenLocationReverse,
Alphabetical
};
void sort();
bool operator()(uint32_t mbr_idx1, uint32_t mbr_idx2);
MemberSorter(SPIRType &t, Meta &m, SortAspect sa);
SPIRType &type;
Meta &meta;
SortAspect sort_aspect;
};
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
File diff suppressed because it is too large Load Diff
@@ -0,0 +1,94 @@
/*
* Copyright 2018-2019 Arm Limited
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_PARSER_HPP
#define SPIRV_CROSS_PARSER_HPP
#include "spirv_cross_parsed_ir.hpp"
#include <stdint.h>
namespace SPIRV_CROSS_NAMESPACE
{
class Parser
{
public:
Parser(const uint32_t *spirv_data, size_t word_count);
Parser(std::vector<uint32_t> spirv);
void parse();
ParsedIR &get_parsed_ir()
{
return ir;
}
private:
ParsedIR ir;
SPIRFunction *current_function = nullptr;
SPIRBlock *current_block = nullptr;
void parse(const Instruction &instr);
const uint32_t *stream(const Instruction &instr) const;
template <typename T, typename... P>
T &set(uint32_t id, P &&... args)
{
ir.add_typed_id(static_cast<Types>(T::type), id);
auto &var = variant_set<T>(ir.ids[id], std::forward<P>(args)...);
var.self = id;
return var;
}
template <typename T>
T &get(uint32_t id)
{
return variant_get<T>(ir.ids[id]);
}
template <typename T>
T *maybe_get(uint32_t id)
{
if (ir.ids[id].get_type() == static_cast<Types>(T::type))
return &get<T>(id);
else
return nullptr;
}
template <typename T>
const T &get(uint32_t id) const
{
return variant_get<T>(ir.ids[id]);
}
template <typename T>
const T *maybe_get(uint32_t id) const
{
if (ir.ids[id].get_type() == T::type)
return &get<T>(id);
else
return nullptr;
}
// This must be an ordered data structure so we always pick the same type aliases.
SmallVector<uint32_t> global_struct_cache;
bool types_are_logically_equivalent(const SPIRType &a, const SPIRType &b) const;
bool variable_storage_is_aliased(const SPIRVariable &v) const;
void make_constant_null(uint32_t id, uint32_t type);
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif
+635
View File
@@ -0,0 +1,635 @@
/*
* Copyright 2018-2019 Bradley Austin Davis
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#include "spirv_reflect.hpp"
#include "spirv_glsl.hpp"
#include <iomanip>
using namespace spv;
using namespace SPIRV_CROSS_NAMESPACE;
using namespace std;
namespace simple_json
{
enum class Type
{
Object,
Array,
};
using State = std::pair<Type, bool>;
using Stack = std::stack<State>;
class Stream
{
Stack stack;
StringStream<> buffer;
uint32_t indent{ 0 };
char current_locale_radix_character = '.';
public:
void set_current_locale_radix_character(char c)
{
current_locale_radix_character = c;
}
void begin_json_object();
void end_json_object();
void emit_json_key(const std::string &key);
void emit_json_key_value(const std::string &key, const std::string &value);
void emit_json_key_value(const std::string &key, bool value);
void emit_json_key_value(const std::string &key, uint32_t value);
void emit_json_key_value(const std::string &key, int32_t value);
void emit_json_key_value(const std::string &key, float value);
void emit_json_key_object(const std::string &key);
void emit_json_key_array(const std::string &key);
void begin_json_array();
void end_json_array();
void emit_json_array_value(const std::string &value);
void emit_json_array_value(uint32_t value);
void emit_json_array_value(bool value);
std::string str() const
{
return buffer.str();
}
private:
inline void statement_indent()
{
for (uint32_t i = 0; i < indent; i++)
buffer << " ";
}
template <typename T>
inline void statement_inner(T &&t)
{
buffer << std::forward<T>(t);
}
template <typename T, typename... Ts>
inline void statement_inner(T &&t, Ts &&... ts)
{
buffer << std::forward<T>(t);
statement_inner(std::forward<Ts>(ts)...);
}
template <typename... Ts>
inline void statement(Ts &&... ts)
{
statement_indent();
statement_inner(std::forward<Ts>(ts)...);
buffer << '\n';
}
template <typename... Ts>
void statement_no_return(Ts &&... ts)
{
statement_indent();
statement_inner(std::forward<Ts>(ts)...);
}
};
} // namespace simple_json
using namespace simple_json;
// Hackery to emit JSON without using nlohmann/json C++ library (which requires a
// higher level of compiler compliance than is required by SPIRV-Cross
void Stream::begin_json_array()
{
if (!stack.empty() && stack.top().second)
{
statement_inner(",\n");
}
statement("[");
++indent;
stack.emplace(Type::Array, false);
}
void Stream::end_json_array()
{
if (stack.empty() || stack.top().first != Type::Array)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
{
statement_inner("\n");
}
--indent;
statement_no_return("]");
stack.pop();
if (!stack.empty())
{
stack.top().second = true;
}
}
void Stream::emit_json_array_value(const std::string &value)
{
if (stack.empty() || stack.top().first != Type::Array)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
statement_inner(",\n");
statement_no_return("\"", value, "\"");
stack.top().second = true;
}
void Stream::emit_json_array_value(uint32_t value)
{
if (stack.empty() || stack.top().first != Type::Array)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
statement_inner(",\n");
statement_no_return(std::to_string(value));
stack.top().second = true;
}
void Stream::emit_json_array_value(bool value)
{
if (stack.empty() || stack.top().first != Type::Array)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
statement_inner(",\n");
statement_no_return(value ? "true" : "false");
stack.top().second = true;
}
void Stream::begin_json_object()
{
if (!stack.empty() && stack.top().second)
{
statement_inner(",\n");
}
statement("{");
++indent;
stack.emplace(Type::Object, false);
}
void Stream::end_json_object()
{
if (stack.empty() || stack.top().first != Type::Object)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
{
statement_inner("\n");
}
--indent;
statement_no_return("}");
stack.pop();
if (!stack.empty())
{
stack.top().second = true;
}
}
void Stream::emit_json_key(const std::string &key)
{
if (stack.empty() || stack.top().first != Type::Object)
SPIRV_CROSS_THROW("Invalid JSON state");
if (stack.top().second)
statement_inner(",\n");
statement_no_return("\"", key, "\" : ");
stack.top().second = true;
}
void Stream::emit_json_key_value(const std::string &key, const std::string &value)
{
emit_json_key(key);
statement_inner("\"", value, "\"");
}
void Stream::emit_json_key_value(const std::string &key, uint32_t value)
{
emit_json_key(key);
statement_inner(value);
}
void Stream::emit_json_key_value(const std::string &key, int32_t value)
{
emit_json_key(key);
statement_inner(value);
}
void Stream::emit_json_key_value(const std::string &key, float value)
{
emit_json_key(key);
statement_inner(convert_to_string(value, current_locale_radix_character));
}
void Stream::emit_json_key_value(const std::string &key, bool value)
{
emit_json_key(key);
statement_inner(value ? "true" : "false");
}
void Stream::emit_json_key_object(const std::string &key)
{
emit_json_key(key);
statement_inner("{\n");
++indent;
stack.emplace(Type::Object, false);
}
void Stream::emit_json_key_array(const std::string &key)
{
emit_json_key(key);
statement_inner("[\n");
++indent;
stack.emplace(Type::Array, false);
}
void CompilerReflection::set_format(const std::string &format)
{
if (format != "json")
{
SPIRV_CROSS_THROW("Unsupported format");
}
}
string CompilerReflection::compile()
{
json_stream = std::make_shared<simple_json::Stream>();
json_stream->set_current_locale_radix_character(current_locale_radix_character);
json_stream->begin_json_object();
fixup_type_alias();
reorder_type_alias();
emit_entry_points();
emit_types();
emit_resources();
emit_specialization_constants();
json_stream->end_json_object();
return json_stream->str();
}
void CompilerReflection::emit_types()
{
bool emitted_open_tag = false;
ir.for_each_typed_id<SPIRType>([&](uint32_t, SPIRType &type) {
if (type.basetype == SPIRType::Struct && !type.pointer && type.array.empty())
emit_type(type, emitted_open_tag);
});
if (emitted_open_tag)
{
json_stream->end_json_object();
}
}
void CompilerReflection::emit_type(const SPIRType &type, bool &emitted_open_tag)
{
auto name = type_to_glsl(type);
if (type.type_alias != TypeID(0))
return;
if (!emitted_open_tag)
{
json_stream->emit_json_key_object("types");
emitted_open_tag = true;
}
json_stream->emit_json_key_object("_" + std::to_string(type.self));
json_stream->emit_json_key_value("name", name);
json_stream->emit_json_key_array("members");
// FIXME ideally we'd like to emit the size of a structure as a
// convenience to people parsing the reflected JSON. The problem
// is that there's no implicit size for a type. It's final size
// will be determined by the top level declaration in which it's
// included. So there might be one size for the struct if it's
// included in a std140 uniform block and another if it's included
// in a std430 uniform block.
// The solution is to include *all* potential sizes as a map of
// layout type name to integer, but that will probably require
// some additional logic being written in this class, or in the
// parent CompilerGLSL class.
auto size = type.member_types.size();
for (uint32_t i = 0; i < size; ++i)
{
emit_type_member(type, i);
}
json_stream->end_json_array();
json_stream->end_json_object();
}
void CompilerReflection::emit_type_member(const SPIRType &type, uint32_t index)
{
auto &membertype = get<SPIRType>(type.member_types[index]);
json_stream->begin_json_object();
auto name = to_member_name(type, index);
// FIXME we'd like to emit the offset of each member, but such offsets are
// context dependent. See the comment above regarding structure sizes
json_stream->emit_json_key_value("name", name);
if (membertype.basetype == SPIRType::Struct)
{
json_stream->emit_json_key_value("type", "_" + std::to_string(membertype.self));
}
else
{
json_stream->emit_json_key_value("type", type_to_glsl(membertype));
}
emit_type_member_qualifiers(type, index);
json_stream->end_json_object();
}
void CompilerReflection::emit_type_array(const SPIRType &type)
{
if (!type.array.empty())
{
json_stream->emit_json_key_array("array");
// Note that we emit the zeros here as a means of identifying
// unbounded arrays. This is necessary as otherwise there would
// be no way of differentiating between float[4] and float[4][]
for (const auto &value : type.array)
json_stream->emit_json_array_value(value);
json_stream->end_json_array();
}
}
void CompilerReflection::emit_type_member_qualifiers(const SPIRType &type, uint32_t index)
{
auto flags = combined_decoration_for_member(type, index);
if (flags.get(DecorationRowMajor))
json_stream->emit_json_key_value("row_major", true);
auto &membertype = get<SPIRType>(type.member_types[index]);
emit_type_array(membertype);
auto &memb = ir.meta[type.self].members;
if (index < memb.size())
{
auto &dec = memb[index];
if (dec.decoration_flags.get(DecorationLocation))
json_stream->emit_json_key_value("location", dec.location);
if (dec.decoration_flags.get(DecorationOffset))
json_stream->emit_json_key_value("offset", dec.offset);
}
}
string CompilerReflection::execution_model_to_str(spv::ExecutionModel model)
{
switch (model)
{
case ExecutionModelVertex:
return "vert";
case ExecutionModelTessellationControl:
return "tesc";
case ExecutionModelTessellationEvaluation:
return "tese";
case ExecutionModelGeometry:
return "geom";
case ExecutionModelFragment:
return "frag";
case ExecutionModelGLCompute:
return "comp";
case ExecutionModelRayGenerationNV:
return "rgen";
case ExecutionModelIntersectionNV:
return "rint";
case ExecutionModelAnyHitNV:
return "rahit";
case ExecutionModelClosestHitNV:
return "rchit";
case ExecutionModelMissNV:
return "rmiss";
case ExecutionModelCallableNV:
return "rcall";
default:
return "???";
}
}
// FIXME include things like the local_size dimensions, geometry output vertex count, etc
void CompilerReflection::emit_entry_points()
{
auto entries = get_entry_points_and_stages();
if (!entries.empty())
{
// Needed to make output deterministic.
sort(begin(entries), end(entries), [](const EntryPoint &a, const EntryPoint &b) -> bool {
if (a.execution_model < b.execution_model)
return true;
else if (a.execution_model > b.execution_model)
return false;
else
return a.name < b.name;
});
json_stream->emit_json_key_array("entryPoints");
for (auto &e : entries)
{
json_stream->begin_json_object();
json_stream->emit_json_key_value("name", e.name);
json_stream->emit_json_key_value("mode", execution_model_to_str(e.execution_model));
if (e.execution_model == ExecutionModelGLCompute)
{
const auto &spv_entry = get_entry_point(e.name, e.execution_model);
SpecializationConstant spec_x, spec_y, spec_z;
get_work_group_size_specialization_constants(spec_x, spec_y, spec_z);
json_stream->emit_json_key_array("workgroup_size");
json_stream->emit_json_array_value(spec_x.id != ID(0) ? spec_x.constant_id :
spv_entry.workgroup_size.x);
json_stream->emit_json_array_value(spec_y.id != ID(0) ? spec_y.constant_id :
spv_entry.workgroup_size.y);
json_stream->emit_json_array_value(spec_z.id != ID(0) ? spec_z.constant_id :
spv_entry.workgroup_size.z);
json_stream->end_json_array();
json_stream->emit_json_key_array("workgroup_size_is_spec_constant_id");
json_stream->emit_json_array_value(spec_x.id != ID(0));
json_stream->emit_json_array_value(spec_y.id != ID(0));
json_stream->emit_json_array_value(spec_z.id != ID(0));
json_stream->end_json_array();
}
json_stream->end_json_object();
}
json_stream->end_json_array();
}
}
void CompilerReflection::emit_resources()
{
auto res = get_shader_resources();
emit_resources("subpass_inputs", res.subpass_inputs);
emit_resources("inputs", res.stage_inputs);
emit_resources("outputs", res.stage_outputs);
emit_resources("textures", res.sampled_images);
emit_resources("separate_images", res.separate_images);
emit_resources("separate_samplers", res.separate_samplers);
emit_resources("images", res.storage_images);
emit_resources("ssbos", res.storage_buffers);
emit_resources("ubos", res.uniform_buffers);
emit_resources("push_constants", res.push_constant_buffers);
emit_resources("counters", res.atomic_counters);
emit_resources("acceleration_structures", res.acceleration_structures);
}
void CompilerReflection::emit_resources(const char *tag, const SmallVector<Resource> &resources)
{
if (resources.empty())
{
return;
}
json_stream->emit_json_key_array(tag);
for (auto &res : resources)
{
auto &type = get_type(res.type_id);
auto typeflags = ir.meta[type.self].decoration.decoration_flags;
auto &mask = get_decoration_bitset(res.id);
// If we don't have a name, use the fallback for the type instead of the variable
// for SSBOs and UBOs since those are the only meaningful names to use externally.
// Push constant blocks are still accessed by name and not block name, even though they are technically Blocks.
bool is_push_constant = get_storage_class(res.id) == StorageClassPushConstant;
bool is_block = get_decoration_bitset(type.self).get(DecorationBlock) ||
get_decoration_bitset(type.self).get(DecorationBufferBlock);
ID fallback_id = !is_push_constant && is_block ? ID(res.base_type_id) : ID(res.id);
json_stream->begin_json_object();
if (type.basetype == SPIRType::Struct)
{
json_stream->emit_json_key_value("type", "_" + std::to_string(res.base_type_id));
}
else
{
json_stream->emit_json_key_value("type", type_to_glsl(type));
}
json_stream->emit_json_key_value("name", !res.name.empty() ? res.name : get_fallback_name(fallback_id));
{
bool ssbo_block = type.storage == StorageClassStorageBuffer ||
(type.storage == StorageClassUniform && typeflags.get(DecorationBufferBlock));
if (ssbo_block)
{
auto buffer_flags = get_buffer_block_flags(res.id);
if (buffer_flags.get(DecorationNonReadable))
json_stream->emit_json_key_value("writeonly", true);
if (buffer_flags.get(DecorationNonWritable))
json_stream->emit_json_key_value("readonly", true);
if (buffer_flags.get(DecorationRestrict))
json_stream->emit_json_key_value("restrict", true);
if (buffer_flags.get(DecorationCoherent))
json_stream->emit_json_key_value("coherent", true);
}
}
emit_type_array(type);
{
bool is_sized_block = is_block && (get_storage_class(res.id) == StorageClassUniform ||
get_storage_class(res.id) == StorageClassUniformConstant ||
get_storage_class(res.id) == StorageClassStorageBuffer);
if (is_sized_block)
{
uint32_t block_size = uint32_t(get_declared_struct_size(get_type(res.base_type_id)));
json_stream->emit_json_key_value("block_size", block_size);
}
}
if (type.storage == StorageClassPushConstant)
json_stream->emit_json_key_value("push_constant", true);
if (mask.get(DecorationLocation))
json_stream->emit_json_key_value("location", get_decoration(res.id, DecorationLocation));
if (mask.get(DecorationRowMajor))
json_stream->emit_json_key_value("row_major", true);
if (mask.get(DecorationColMajor))
json_stream->emit_json_key_value("column_major", true);
if (mask.get(DecorationIndex))
json_stream->emit_json_key_value("index", get_decoration(res.id, DecorationIndex));
if (type.storage != StorageClassPushConstant && mask.get(DecorationDescriptorSet))
json_stream->emit_json_key_value("set", get_decoration(res.id, DecorationDescriptorSet));
if (mask.get(DecorationBinding))
json_stream->emit_json_key_value("binding", get_decoration(res.id, DecorationBinding));
if (mask.get(DecorationInputAttachmentIndex))
json_stream->emit_json_key_value("input_attachment_index",
get_decoration(res.id, DecorationInputAttachmentIndex));
if (mask.get(DecorationOffset))
json_stream->emit_json_key_value("offset", get_decoration(res.id, DecorationOffset));
// For images, the type itself adds a layout qualifer.
// Only emit the format for storage images.
if (type.basetype == SPIRType::Image && type.image.sampled == 2)
{
const char *fmt = format_to_glsl(type.image.format);
if (fmt != nullptr)
json_stream->emit_json_key_value("format", std::string(fmt));
}
json_stream->end_json_object();
}
json_stream->end_json_array();
}
void CompilerReflection::emit_specialization_constants()
{
auto specialization_constants = get_specialization_constants();
if (specialization_constants.empty())
return;
json_stream->emit_json_key_array("specialization_constants");
for (const auto spec_const : specialization_constants)
{
auto &c = get<SPIRConstant>(spec_const.id);
auto type = get<SPIRType>(c.constant_type);
json_stream->begin_json_object();
json_stream->emit_json_key_value("id", spec_const.constant_id);
json_stream->emit_json_key_value("type", type_to_glsl(type));
switch (type.basetype)
{
case SPIRType::UInt:
json_stream->emit_json_key_value("default_value", c.scalar());
break;
case SPIRType::Int:
json_stream->emit_json_key_value("default_value", c.scalar_i32());
break;
case SPIRType::Float:
json_stream->emit_json_key_value("default_value", c.scalar_f32());
break;
case SPIRType::Boolean:
json_stream->emit_json_key_value("default_value", c.scalar() != 0);
break;
default:
break;
}
json_stream->end_json_object();
}
json_stream->end_json_array();
}
string CompilerReflection::to_member_name(const SPIRType &type, uint32_t index) const
{
auto *type_meta = ir.find_meta(type.self);
if (type_meta)
{
auto &memb = type_meta->members;
if (index < memb.size() && !memb[index].alias.empty())
return memb[index].alias;
else
return join("_m", index);
}
else
return join("_m", index);
}
@@ -0,0 +1,83 @@
/*
* Copyright 2018-2019 Bradley Austin Davis
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#ifndef SPIRV_CROSS_REFLECT_HPP
#define SPIRV_CROSS_REFLECT_HPP
#include "spirv_glsl.hpp"
#include <utility>
namespace simple_json
{
class Stream;
}
namespace SPIRV_CROSS_NAMESPACE
{
class CompilerReflection : public CompilerGLSL
{
using Parent = CompilerGLSL;
public:
explicit CompilerReflection(std::vector<uint32_t> spirv_)
: Parent(std::move(spirv_))
{
options.vulkan_semantics = true;
}
CompilerReflection(const uint32_t *ir_, size_t word_count)
: Parent(ir_, word_count)
{
options.vulkan_semantics = true;
}
explicit CompilerReflection(const ParsedIR &ir_)
: CompilerGLSL(ir_)
{
options.vulkan_semantics = true;
}
explicit CompilerReflection(ParsedIR &&ir_)
: CompilerGLSL(std::move(ir_))
{
options.vulkan_semantics = true;
}
void set_format(const std::string &format);
std::string compile() override;
private:
static std::string execution_model_to_str(spv::ExecutionModel model);
void emit_entry_points();
void emit_types();
void emit_resources();
void emit_specialization_constants();
void emit_type(const SPIRType &type, bool &emitted_open_tag);
void emit_type_member(const SPIRType &type, uint32_t index);
void emit_type_member_qualifiers(const SPIRType &type, uint32_t index);
void emit_type_array(const SPIRType &type);
void emit_resources(const char *tag, const SmallVector<Resource> &resources);
std::string to_member_name(const SPIRType &type, uint32_t index) const;
std::shared_ptr<simple_json::Stream> json_stream;
};
} // namespace SPIRV_CROSS_NAMESPACE
#endif