Skip to content

A
angle
Commit 63c91a3b by Le Hoang Quyen Committed by Commit Bot
Options

Metal: autogen for PBO related shaders.

Bug: angleproject:2634 Change-Id: I36e5a6a4cb11d82f666c410b33fdb39ad35a5e80 Reviewed-on: https://chromium-review.googlesource.com/c/angle/angle/+/2336557Reviewed-by: 's avatarJonah Ryan-Davis <jonahr@google.com> Reviewed-by: 's avatarJamie Madill <jmadill@chromium.org> Commit-Queue: Jamie Madill <jmadill@chromium.org>
parent e82f2822
Showing with 1929 additions and 17 deletions
scripts/code_generation_hashes/Metal_default_shaders.json
{ {
"src/libANGLE/renderer/angle_format_map.json":
"aa4a0d3463b76858a75787b9cdec8e98",
"src/libANGLE/renderer/metal/shaders/blit.metal": "src/libANGLE/renderer/metal/shaders/blit.metal":
"f02720a0f3d9624b67bdb808cbc94c7f", "f02720a0f3d9624b67bdb808cbc94c7f",
"src/libANGLE/renderer/metal/shaders/clear.metal": "src/libANGLE/renderer/metal/shaders/clear.metal":
"67da9886363c530132e5bc2199bab2db", "67da9886363c530132e5bc2199bab2db",
"src/libANGLE/renderer/metal/shaders/common.h": "src/libANGLE/renderer/metal/shaders/common.h":
"4260719e5f35107db1029d2fe6e1a732", "0cadef47fb785aa3102acb2a6eae94bb",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_ios_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_ios_autogen.inc":
"f4338f3b943f760d0ba0684415d5ad9d", "242c1dcfed23f2ce130a5813b0845885",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_ios_sim_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_ios_sim_autogen.inc":
"72f85d542aaecf87adfe7cda971af2c4", "472d8b494249c8dd51900fbedf2bfc1e",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_mac_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_debug_mac_autogen.inc":
"41e6633271eca9848076394f9090c155", "71503cf6b999d7d34592851b623bd2ec",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_ios_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_ios_autogen.inc":
"73e40c41a6cdecdd46fb98383c1a03f1", "7bd32594077ff118df55d19607cb03cc",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_ios_sim_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_ios_sim_autogen.inc":
"1b32e75f82db3b3e10d32cca04add8b5", "144649774236fb4c89ea333c6b7c7432",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_mac_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_2_1_mac_autogen.inc":
"d4b8c7ddabab1f4099fea4291e5093ef", "5a88edbf58f094953b65ba76f7dbdbec",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_ios_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_ios_autogen.inc":
"962ff1364859ad38a097a3b6d837a1eb", "d1c9a9922b3dc5761f88e63f3f7715b6",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_ios_sim_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_ios_sim_autogen.inc":
"e34a72c1a16ecb43357da1b192328859", "821c685f4ba29301b27eb4cff53fd6d9",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_mac_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_debug_mac_autogen.inc":
"652937413a0ff7885683dcfaf8c16941", "4e7dd43bf0f6403a551052c6d83b4abe",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_ios_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_ios_autogen.inc":
"7d7d55cacb18961cc04839cf5aa95c7a", "e7421348d7b6168f78645f904b438129",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_ios_sim_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_ios_sim_autogen.inc":
"31372593af055f15070ebe6f50b19237", "eb0cb7ef0f1fcd26d2747cee596c4d1f",
"src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_mac_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_mac_autogen.inc":
"a0032256c254737bbdfb58395a7c3f6a", "bce93a1c3e8283b96a41a0088f4d83ad",
"src/libANGLE/renderer/metal/shaders/compiled/mtl_default_shaders_autogen.inc": "src/libANGLE/renderer/metal/shaders/compiled/mtl_default_shaders_autogen.inc":
"634a127f4e94f6bc3123e89850d010ee", "634a127f4e94f6bc3123e89850d010ee",
"src/libANGLE/renderer/metal/shaders/constants.h": "src/libANGLE/renderer/metal/shaders/constants.h":
"dad1a869a1095be669b7da5651901d38", "dad1a869a1095be669b7da5651901d38",
"src/libANGLE/renderer/metal/shaders/copy_buffer.metal":
"97b258edbade1ed088e4c03a1102f974",
"src/libANGLE/renderer/metal/shaders/format_autogen.h":
"2d9f0752644dffd14423fa0c03ca8e65",
"src/libANGLE/renderer/metal/shaders/gen_indices.metal": "src/libANGLE/renderer/metal/shaders/gen_indices.metal":
"87a76d5e12825111c0595f69e79f5d20", "87a76d5e12825111c0595f69e79f5d20",
"src/libANGLE/renderer/metal/shaders/gen_mipmap.metal": "src/libANGLE/renderer/metal/shaders/gen_mipmap.metal":
"54dca94c48bead446624079070b9b309", "54dca94c48bead446624079070b9b309",
"src/libANGLE/renderer/metal/shaders/gen_mtl_internal_shaders.py": "src/libANGLE/renderer/metal/shaders/gen_mtl_internal_shaders.py":
"baaa20430675d8909fa4c61f8a926eb8", "691460118fccd9c7fd0c451adaa2b330",
"src/libANGLE/renderer/metal/shaders/visibility.metal": "src/libANGLE/renderer/metal/shaders/visibility.metal":
"998d705656c63849edd0187cd8062fc7" "998d705656c63849edd0187cd8062fc7"
} }
\ No newline at end of file
src/libANGLE/renderer/metal/BUILD.gn
...@@ -72,6 +72,7 @@ _metal_backend_sources = [ ...@@ -72,6 +72,7 @@ _metal_backend_sources = [
"shaders/compiled/compiled_default_metallib_mac_autogen.inc", "shaders/compiled/compiled_default_metallib_mac_autogen.inc",
"shaders/compiled/mtl_default_shaders_autogen.inc", "shaders/compiled/mtl_default_shaders_autogen.inc",
"shaders/constants.h", "shaders/constants.h",
"shaders/format_autogen.h",
] ]
config("angle_metal_backend_config") { config("angle_metal_backend_config") {
......
src/libANGLE/renderer/metal/shaders/common.h
...@@ -134,6 +134,128 @@ static inline float4 linearToSRGB(float4 color) ...@@ -134,6 +134,128 @@ static inline float4 linearToSRGB(float4 color)
return float4(linearToSRGB(color.r), linearToSRGB(color.g), linearToSRGB(color.b), color.a); return float4(linearToSRGB(color.r), linearToSRGB(color.g), linearToSRGB(color.b), color.a);
} }
template <typename Short>
static inline Short bytesToShort(constant uchar *input, uint offset)
{
Short inputLo = input[offset];
Short inputHi = input[offset + 1];
// Little endian conversion:
return inputLo | (inputHi << 8);
}
template <typename Int>
static inline Int bytesToInt(constant uchar *input, uint offset)
{
Int input0 = input[offset];
Int input1 = input[offset + 1];
Int input2 = input[offset + 2];
Int input3 = input[offset + 3];
// Little endian conversion:
return input0 | (input1 << 8) | (input2 << 16) | (input3 << 24);
}
template <typename Short>
static inline void shortToBytes(Short val, uint offset, device uchar *output)
{
ushort valUnsigned = as_type<ushort>(val);
output[offset] = valUnsigned & 0xff;
output[offset + 1] = (valUnsigned >> 8) & 0xff;
}
template <typename Int>
static inline void intToBytes(Int val, uint offset, device uchar *output)
{
uint valUnsigned = as_type<uint>(val);
output[offset] = valUnsigned & 0xff;
output[offset + 1] = (valUnsigned >> 8) & 0xff;
output[offset + 2] = (valUnsigned >> 16) & 0xff;
output[offset + 3] = (valUnsigned >> 24) & 0xff;
}
static inline void floatToBytes(float val, uint offset, device uchar *output)
{
intToBytes(as_type<uint>(val), offset, output);
}
static inline void int24bitToBytes(uint val, uint offset, device uchar *output)
{
output[offset] = val & 0xff;
output[offset + 1] = (val >> 8) & 0xff;
output[offset + 2] = (val >> 16) & 0xff;
}
template <unsigned int inputBitCount, unsigned int inputBitStart, typename T>
static inline T getShiftedData(T input)
{
static_assert(inputBitCount + inputBitStart <= (sizeof(T) * 8),
"T must have at least as many bits as inputBitCount + inputBitStart.");
const T mask = (1 << inputBitCount) - 1;
return (input >> inputBitStart) & mask;
}
template <unsigned int inputBitCount, unsigned int inputBitStart, typename T>
static inline T shiftData(T input)
{
static_assert(inputBitCount + inputBitStart <= (sizeof(T) * 8),
"T must have at least as many bits as inputBitCount + inputBitStart.");
const T mask = (1 << inputBitCount) - 1;
return (input & mask) << inputBitStart;
}
template <unsigned int inputBitCount, typename T>
static inline float normalizedToFloat(T input)
{
static_assert(inputBitCount <= (sizeof(T) * 8),
"T must have more bits than or same bits as inputBitCount.");
static_assert(inputBitCount <= 23, "Only single precision is supported");
constexpr float inverseMax = 1.0f / ((1 << inputBitCount) - 1);
return input * inverseMax;
}
template <typename T>
static inline float normalizedToFloat(T input)
{
return normalizedToFloat<sizeof(T) * 8, T>(input);
}
template <>
inline float normalizedToFloat(short input)
{
constexpr float inverseMax = 1.0f / 0x7fff;
return static_cast<float>(input) * inverseMax;
}
template <>
inline float normalizedToFloat(int input)
{
constexpr float inverseMax = 1.0f / 0x7fffffff;
return static_cast<float>(input) * inverseMax;
}
template <>
inline float normalizedToFloat(uint input)
{
constexpr float inverseMax = 1.0f / 0xffffffff;
return static_cast<float>(input) * inverseMax;
}
template <unsigned int outputBitCount, typename T>
static inline T floatToNormalized(float input)
{
static_assert(outputBitCount <= (sizeof(T) * 8),
"T must have more bits than or same bits as inputBitCount.");
static_assert(outputBitCount <= 23, "Only single precision is supported");
return static_cast<T>(((1 << outputBitCount) - 1) * input + 0.5f);
}
template <typename T>
static inline T floatToNormalized(float input)
{
return floatToNormalized<sizeof(T) * 8, T>(input);
}
} // namespace mtl_shader } // namespace mtl_shader
} // namespace rx } // namespace rx
......
src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_ios_autogen.inc
This source diff could not be displayed because it is too large. You can view the blob instead.
src/libANGLE/renderer/metal/shaders/compiled/compiled_default_metallib_mac_autogen.inc
This source diff could not be displayed because it is too large. You can view the blob instead.
src/libANGLE/renderer/metal/shaders/copy_buffer.metal 0 → 100644
//
// Copyright 2020 The ANGLE Project. All rights reserved.
// Use of this source code is governed by a BSD-style license that can be
// found in the LICENSE file.
//
// copy_buffer.metal: implements compute shader that copy formatted data from buffer to texture,
// and from texture to buffer.
// NOTE(hqle): This file is a bit hard to read but there are a lot of repeated works, and it would
// be a pain to implement without the use of macros.
//
#include <metal_pack>
#include "common.h"
#include "format_autogen.h"
using namespace rx::mtl_shader;
constant int kCopyFormatType [[function_constant(1)]];
/* -------- copy pixel data between buffer and texture ---------*/
constant int kCopyTextureType [[function_constant(2)]];
constant bool kCopyTextureType2D = kCopyTextureType == kTextureType2D;
constant bool kCopyTextureType2DArray = kCopyTextureType == kTextureType2DArray;
constant bool kCopyTextureType2DMS = kCopyTextureType == kTextureType2DMultisample;
constant bool kCopyTextureTypeCube = kCopyTextureType == kTextureTypeCube;
constant bool kCopyTextureType3D = kCopyTextureType == kTextureType3D;
struct CopyPixelParams
{
uint3 copySize;
uint3 textureOffset;
uint bufferStartOffset;
uint pixelSize;
uint bufferRowPitch;
uint bufferDepthPitch;
};
struct WritePixelParams
{
uint2 copySize;
uint2 textureOffset;
uint bufferStartOffset;
uint pixelSize;
uint bufferRowPitch;
uint textureLevel;
uint textureLayer;
bool reverseTextureRowOrder;
};
// clang-format off
#define TEXTURE_PARAMS(TYPE, ACCESS, NAME_PREFIX) \
texture2d<TYPE, ACCESS> NAME_PREFIX##Texture2d \
[[texture(0), function_constant(kCopyTextureType2D)]], \
texture2d_array<TYPE, ACCESS> NAME_PREFIX##Texture2dArray \
[[texture(0), function_constant(kCopyTextureType2DArray)]], \
texture3d<TYPE, ACCESS> NAME_PREFIX##Texture3d \
[[texture(0), function_constant(kCopyTextureType3D)]], \
texturecube<TYPE, ACCESS> NAME_PREFIX##TextureCube \
[[texture(0), function_constant(kCopyTextureTypeCube)]]
#define FORWARD_TEXTURE_PARAMS(NAME_PREFIX) \
NAME_PREFIX##Texture2d, \
NAME_PREFIX##Texture2dArray, \
NAME_PREFIX##Texture3d, \
NAME_PREFIX##TextureCube
// Params for reading from buffer to texture
#define DEST_TEXTURE_PARAMS(TYPE) TEXTURE_PARAMS(TYPE, access::write, dst)
#define FORWARD_DEST_TEXTURE_PARAMS FORWARD_TEXTURE_PARAMS(dst)
#define COMMON_READ_KERNEL_PARAMS(TEXTURE_TYPE) \
ushort3 gIndices [[thread_position_in_grid]], \
constant CopyPixelParams &options[[buffer(0)]], \
constant uchar *buffer [[buffer(1)]], \
DEST_TEXTURE_PARAMS(TEXTURE_TYPE)
#define COMMON_READ_FUNC_PARAMS \
uint bufferOffset, \
constant uchar *buffer
#define FORWARD_COMMON_READ_FUNC_PARAMS bufferOffset, buffer
// Params for writing to buffer by coping from texture.
// (NOTE: it has additional multisample source texture parameter)
#define SRC_TEXTURE_PARAMS(TYPE) \
TEXTURE_PARAMS(TYPE, access::read, src), \
texture2d_ms<TYPE, access::read> srcTexture2dMS \
[[texture(0), function_constant(kCopyTextureType2DMS)]] \
#define FORWARD_SRC_TEXTURE_PARAMS FORWARD_TEXTURE_PARAMS(src), srcTexture2dMS
#define COMMON_WRITE_KERNEL_PARAMS(TEXTURE_TYPE) \
ushort2 gIndices [[thread_position_in_grid]], \
constant WritePixelParams &options[[buffer(0)]], \
SRC_TEXTURE_PARAMS(TEXTURE_TYPE), \
device uchar *buffer [[buffer(1)]] \
#define COMMON_WRITE_FUNC_PARAMS(TYPE) \
ushort2 gIndices, \
constant WritePixelParams &options,\
uint bufferOffset, \
vec<TYPE, 4> color, \
device uchar *buffer \
#define COMMON_WRITE_FLOAT_FUNC_PARAMS COMMON_WRITE_FUNC_PARAMS(float)
#define COMMON_WRITE_SINT_FUNC_PARAMS COMMON_WRITE_FUNC_PARAMS(int)
#define COMMON_WRITE_UINT_FUNC_PARAMS COMMON_WRITE_FUNC_PARAMS(uint)
#define FORWARD_COMMON_WRITE_FUNC_PARAMS gIndices, options, bufferOffset, color, buffer
// clang-format on
// Write to texture code based on texture type:
template <typename T>
static inline void textureWrite(ushort3 gIndices,
constant CopyPixelParams &options,
vec<T, 4> color,
DEST_TEXTURE_PARAMS(T))
{
uint3 writeIndices = options.textureOffset + uint3(gIndices);
switch (kCopyTextureType)
{
case kTextureType2D:
dstTexture2d.write(color, writeIndices.xy);
break;
case kTextureType2DArray:
dstTexture2dArray.write(color, writeIndices.xy, writeIndices.z);
break;
case kTextureType3D:
dstTexture3d.write(color, writeIndices);
break;
case kTextureTypeCube:
dstTextureCube.write(color, writeIndices.xy, writeIndices.z);
break;
}
}
// Read from texture code based on texture type:
template <typename T>
static inline vec<T, 4> textureRead(ushort2 gIndices,
constant WritePixelParams &options,
SRC_TEXTURE_PARAMS(T))
{
vec<T, 4> color;
uint2 coords = uint2(gIndices);
if (options.reverseTextureRowOrder)
{
coords.y = options.copySize.y - 1 - gIndices.y;
}
coords += options.textureOffset;
switch (kCopyTextureType)
{
case kTextureType2D:
color = srcTexture2d.read(coords.xy, options.textureLevel);
break;
case kTextureType2DArray:
color = srcTexture2dArray.read(coords.xy, options.textureLayer, options.textureLevel);
break;
case kTextureType2DMultisample:
color = resolveTextureMS(srcTexture2dMS, coords.xy);
break;
case kTextureType3D:
color = srcTexture3d.read(uint3(coords, options.textureLayer), options.textureLevel);
break;
case kTextureTypeCube:
color = srcTextureCube.read(coords.xy, options.textureLayer, options.textureLevel);
break;
}
return color;
}
// Calculate offset into buffer:
#define CALC_BUFFER_READ_OFFSET(pixelSize) \
options.bufferStartOffset + (gIndices.z * options.bufferDepthPitch + \
gIndices.y * options.bufferRowPitch + gIndices.x * pixelSize)
#define CALC_BUFFER_WRITE_OFFSET(pixelSize) \
options.bufferStartOffset + (gIndices.y * options.bufferRowPitch + gIndices.x * pixelSize)
// Per format handling code:
#define READ_FORMAT_SWITCH_CASE(format) \
case FormatID::format: { \
auto color = read##format(FORWARD_COMMON_READ_FUNC_PARAMS); \
textureWrite(gIndices, options, color, FORWARD_DEST_TEXTURE_PARAMS); \
} \
break;
#define WRITE_FORMAT_SWITCH_CASE(format) \
case FormatID::format: { \
auto color = textureRead(gIndices, options, FORWARD_SRC_TEXTURE_PARAMS); \
write##format(FORWARD_COMMON_WRITE_FUNC_PARAMS); \
} \
break;
#define READ_KERNEL_GUARD \
if (gIndices.x >= options.copySize.x || gIndices.y >= options.copySize.y || \
gIndices.z >= options.copySize.z) \
{ \
return; \
}
#define WRITE_KERNEL_GUARD \
if (gIndices.x >= options.copySize.x || gIndices.y >= options.copySize.y) \
{ \
return; \
}
// R5G6B5
static inline float4 readR5G6B5_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
ushort src = bytesToShort<ushort>(buffer, bufferOffset);
color.r = normalizedToFloat<5>(getShiftedData<5, 11>(src));
color.g = normalizedToFloat<6>(getShiftedData<6, 5>(src));
color.b = normalizedToFloat<5>(getShiftedData<5, 0>(src));
color.a = 1.0;
return color;
}
static inline void writeR5G6B5_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
ushort dst = shiftData<5, 11>(floatToNormalized<5, ushort>(color.r)) |
shiftData<6, 5>(floatToNormalized<6, ushort>(color.g)) |
shiftData<5, 0>(floatToNormalized<5, ushort>(color.b));
shortToBytes(dst, bufferOffset, buffer);
}
// R4G4B4A4
static inline float4 readR4G4B4A4_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
ushort src = bytesToShort<ushort>(buffer, bufferOffset);
color.r = normalizedToFloat<4>(getShiftedData<4, 12>(src));
color.g = normalizedToFloat<4>(getShiftedData<4, 8>(src));
color.b = normalizedToFloat<4>(getShiftedData<4, 4>(src));
color.a = normalizedToFloat<4>(getShiftedData<4, 0>(src));
return color;
}
static inline void writeR4G4B4A4_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
ushort dst = shiftData<4, 12>(floatToNormalized<4, ushort>(color.r)) |
shiftData<4, 8>(floatToNormalized<4, ushort>(color.g)) |
shiftData<4, 4>(floatToNormalized<4, ushort>(color.b)) |
shiftData<4, 0>(floatToNormalized<4, ushort>(color.a));
;
shortToBytes(dst, bufferOffset, buffer);
}
// R5G5B5A1
static inline float4 readR5G5B5A1_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
ushort src = bytesToShort<ushort>(buffer, bufferOffset);
color.r = normalizedToFloat<5>(getShiftedData<5, 11>(src));
color.g = normalizedToFloat<5>(getShiftedData<5, 6>(src));
color.b = normalizedToFloat<5>(getShiftedData<5, 1>(src));
color.a = normalizedToFloat<1>(getShiftedData<1, 0>(src));
return color;
}
static inline void writeR5G5B5A1_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
ushort dst = shiftData<5, 11>(floatToNormalized<5, ushort>(color.r)) |
shiftData<5, 6>(floatToNormalized<5, ushort>(color.g)) |
shiftData<5, 1>(floatToNormalized<5, ushort>(color.b)) |
shiftData<1, 0>(floatToNormalized<1, ushort>(color.a));
;
shortToBytes(dst, bufferOffset, buffer);
}
// R10G10B10A2_SINT
static inline int4 readR10G10B10A2_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
int src = bytesToInt<int>(buffer, bufferOffset);
constexpr int3 rgbSignMask(0x200); // 1 set at the 9 bit
constexpr int3 negativeMask(0xfffffc00); // All bits from 10 to 31 set to 1
constexpr int alphaSignMask = 0x2;
constexpr int alphaNegMask = 0xfffffffc;
color.r = getShiftedData<10, 0>(src);
color.g = getShiftedData<10, 10>(src);
color.b = getShiftedData<10, 20>(src);
int3 isRgbNegative = (color.rgb & rgbSignMask) >> 9;
color.rgb = (isRgbNegative * negativeMask) | color.rgb;
color.a = getShiftedData<2, 30>(src);
int isAlphaNegative = color.a & alphaSignMask >> 1;
color.a = (isAlphaNegative * alphaNegMask) | color.a;
return color;
}
// R10G10B10A2_UINT
static inline uint4 readR10G10B10A2_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
uint src = bytesToInt<uint>(buffer, bufferOffset);
color.r = getShiftedData<10, 0>(src);
color.g = getShiftedData<10, 10>(src);
color.b = getShiftedData<10, 20>(src);
color.a = getShiftedData<2, 30>(src);
return color;
}
// R8G8B8A8 generic
static inline float4 readR8G8B8A8(COMMON_READ_FUNC_PARAMS, bool isSRGB)
{
float4 color;
uint src = bytesToInt<uint>(buffer, bufferOffset);
if (isSRGB)
{
color = unpack_unorm4x8_srgb_to_float(src);
}
else
{
color = unpack_unorm4x8_to_float(src);
}
return color;
}
static inline void writeR8G8B8A8(COMMON_WRITE_FLOAT_FUNC_PARAMS, bool isSRGB)
{
uint dst;
if (isSRGB)
{
dst = pack_float_to_srgb_unorm4x8(color);
}
else
{
dst = pack_float_to_unorm4x8(color);
}
intToBytes(dst, bufferOffset, buffer);
}
static inline float4 readR8G8B8(COMMON_READ_FUNC_PARAMS, bool isSRGB)
{
float4 color;
color.r = normalizedToFloat<uchar>(buffer[bufferOffset]);
color.g = normalizedToFloat<uchar>(buffer[bufferOffset + 1]);
color.b = normalizedToFloat<uchar>(buffer[bufferOffset + 2]);
color.a = 1.0;
if (isSRGB)
{
color = sRGBtoLinear(color);
}
return color;
}
static inline void writeR8G8B8(COMMON_WRITE_FLOAT_FUNC_PARAMS, bool isSRGB)
{
color.a = 1.0;
uint dst;
if (isSRGB)
{
dst = pack_float_to_srgb_unorm4x8(color);
}
else
{
dst = pack_float_to_unorm4x8(color);
}
int24bitToBytes(dst, bufferOffset, buffer);
}
// RGBA8_SNORM
static inline float4 readR8G8B8A8_SNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
uint src = bytesToInt<uint>(buffer, bufferOffset);
color = unpack_snorm4x8_to_float(src);
return color;
}
static inline void writeR8G8B8A8_SNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
uint dst = pack_float_to_snorm4x8(color);
intToBytes(dst, bufferOffset, buffer);
}
// RGB8_SNORM
static inline float4 readR8G8B8_SNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<7, char>(buffer[bufferOffset]);
color.g = normalizedToFloat<7, char>(buffer[bufferOffset + 1]);
color.b = normalizedToFloat<7, char>(buffer[bufferOffset + 2]);
color.a = 1.0;
return color;
}
static inline void writeR8G8B8_SNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
uint dst = pack_float_to_snorm4x8(color);
int24bitToBytes(dst, bufferOffset, buffer);
}
// RGBA8
static inline float4 readR8G8B8A8_UNORM(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8A8(FORWARD_COMMON_READ_FUNC_PARAMS, false);
}
static inline void writeR8G8B8A8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
return writeR8G8B8A8(FORWARD_COMMON_WRITE_FUNC_PARAMS, false);
}
static inline float4 readR8G8B8A8_UNORM_SRGB(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8A8(FORWARD_COMMON_READ_FUNC_PARAMS, true);
}
static inline void writeR8G8B8A8_UNORM_SRGB(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
return writeR8G8B8A8(FORWARD_COMMON_WRITE_FUNC_PARAMS, true);
}
// BGRA8
static inline float4 readB8G8R8A8_UNORM(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8A8(FORWARD_COMMON_READ_FUNC_PARAMS, false).bgra;
}
static inline void writeB8G8R8A8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
color.rgba = color.bgra;
return writeR8G8B8A8(FORWARD_COMMON_WRITE_FUNC_PARAMS, false);
}
static inline float4 readB8G8R8A8_UNORM_SRGB(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8A8(FORWARD_COMMON_READ_FUNC_PARAMS, true).bgra;
}
static inline void writeB8G8R8A8_UNORM_SRGB(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
color.rgba = color.bgra;
return writeR8G8B8A8(FORWARD_COMMON_WRITE_FUNC_PARAMS, true);
}
// RGB8
static inline float4 readR8G8B8_UNORM(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8(FORWARD_COMMON_READ_FUNC_PARAMS, false);
}
static inline void writeR8G8B8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
return writeR8G8B8(FORWARD_COMMON_WRITE_FUNC_PARAMS, false);
}
static inline float4 readR8G8B8_UNORM_SRGB(COMMON_READ_FUNC_PARAMS)
{
return readR8G8B8(FORWARD_COMMON_READ_FUNC_PARAMS, true);
}
static inline void writeR8G8B8_UNORM_SRGB(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
return writeR8G8B8(FORWARD_COMMON_WRITE_FUNC_PARAMS, true);
}
// L8
static inline float4 readL8_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = float3(normalizedToFloat<uchar>(buffer[bufferOffset]));
color.a = 1.0;
return color;
}
static inline void writeL8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = floatToNormalized<uchar>(color.r);
}
// A8
static inline void writeA8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = floatToNormalized<uchar>(color.a);
}
// L8A8
static inline float4 readL8A8_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = float3(normalizedToFloat<uchar>(buffer[bufferOffset]));
color.a = normalizedToFloat<uchar>(buffer[bufferOffset + 1]);
return color;
}
static inline void writeL8A8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = floatToNormalized<uchar>(color.r);
buffer[bufferOffset + 1] = floatToNormalized<uchar>(color.a);
}
// R8
static inline float4 readR8_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<uchar>(buffer[bufferOffset]);
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = floatToNormalized<uchar>(color.r);
}
static inline float4 readR8_SNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<7, char>(buffer[bufferOffset]);
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR8_SNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = as_type<uchar>(floatToNormalized<7, char>(color.r));
}
// R8_SINT
static inline int4 readR8_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = as_type<char>(buffer[bufferOffset]);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR8_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
}
// R8_UINT
static inline uint4 readR8_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = as_type<uchar>(buffer[bufferOffset]);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR8_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
}
// R8G8
static inline float4 readR8G8_UNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<uchar>(buffer[bufferOffset]);
color.g = normalizedToFloat<uchar>(buffer[bufferOffset + 1]);
color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR8G8_UNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = floatToNormalized<uchar>(color.r);
buffer[bufferOffset + 1] = floatToNormalized<uchar>(color.g);
}
static inline float4 readR8G8_SNORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<7, char>(buffer[bufferOffset]);
color.g = normalizedToFloat<7, char>(buffer[bufferOffset + 1]);
color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR8G8_SNORM(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
buffer[bufferOffset] = as_type<uchar>(floatToNormalized<7, char>(color.r));
buffer[bufferOffset + 1] = as_type<uchar>(floatToNormalized<7, char>(color.g));
}
// RG8_SINT
static inline int4 readR8G8_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = as_type<char>(buffer[bufferOffset]);
color.g = as_type<char>(buffer[bufferOffset + 1]);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR8G8_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
buffer[bufferOffset + 1] = static_cast<uchar>(color.g);
}
// RG8_UINT
static inline uint4 readR8G8_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = as_type<uchar>(buffer[bufferOffset]);
color.g = as_type<uchar>(buffer[bufferOffset + 1]);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR8G8_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
buffer[bufferOffset + 1] = static_cast<uchar>(color.g);
}
// R8G8B8_SINT
static inline int4 readR8G8B8_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = as_type<char>(buffer[bufferOffset]);
color.g = as_type<char>(buffer[bufferOffset + 1]);
color.b = as_type<char>(buffer[bufferOffset + 2]);
color.a = 1;
return color;
}
// R8G8B8_UINT
static inline uint4 readR8G8B8_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = as_type<uchar>(buffer[bufferOffset]);
color.g = as_type<uchar>(buffer[bufferOffset + 1]);
color.b = as_type<uchar>(buffer[bufferOffset + 2]);
color.a = 1;
return color;
}
// R8G8G8A8_SINT
static inline int4 readR8G8B8A8_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = as_type<char>(buffer[bufferOffset]);
color.g = as_type<char>(buffer[bufferOffset + 1]);
color.b = as_type<char>(buffer[bufferOffset + 2]);
color.a = as_type<char>(buffer[bufferOffset + 3]);
return color;
}
static inline void writeR8G8B8A8_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
buffer[bufferOffset + 1] = static_cast<uchar>(color.g);
buffer[bufferOffset + 2] = static_cast<uchar>(color.b);
buffer[bufferOffset + 3] = static_cast<uchar>(color.a);
}
// R8G8G8A8_UINT
static inline uint4 readR8G8B8A8_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = as_type<uchar>(buffer[bufferOffset]);
color.g = as_type<uchar>(buffer[bufferOffset + 1]);
color.b = as_type<uchar>(buffer[bufferOffset + 2]);
color.a = as_type<uchar>(buffer[bufferOffset + 3]);
return color;
}
static inline void writeR8G8B8A8_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
buffer[bufferOffset] = static_cast<uchar>(color.r);
buffer[bufferOffset + 1] = static_cast<uchar>(color.g);
buffer[bufferOffset + 2] = static_cast<uchar>(color.b);
buffer[bufferOffset + 3] = static_cast<uchar>(color.a);
}
// R16_FLOAT
static inline float4 readR16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.r)), bufferOffset, buffer);
}
// R16_NORM
template <typename ShortType>
static inline float4 readR16_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset));
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
#define readR16_SNORM readR16_NORM<short>
#define readR16_UNORM readR16_NORM<ushort>
// R16_SINT
static inline int4 readR16_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToShort<short>(buffer, bufferOffset);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR16_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
shortToBytes(static_cast<short>(color.r), bufferOffset, buffer);
}
// R16_UINT
static inline uint4 readR16_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToShort<ushort>(buffer, bufferOffset);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR16_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
shortToBytes(static_cast<ushort>(color.r), bufferOffset, buffer);
}
// A16_FLOAT
static inline float4 readA16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.a = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.rgb = 0.0;
return color;
}
static inline void writeA16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.a)), bufferOffset, buffer);
}
// L16_FLOAT
static inline float4 readL16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.a = 1.0;
return color;
}
static inline void writeL16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.r)), bufferOffset, buffer);
}
// L16A16_FLOAT
static inline float4 readL16A16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.a = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 2));
return color;
}
static inline void writeL16A16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.r)), bufferOffset, buffer);
shortToBytes(as_type<ushort>(static_cast<half>(color.a)), bufferOffset + 2, buffer);
}
// R16G16_FLOAT
static inline float4 readR16G16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.g = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 2));
color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR16G16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.r)), bufferOffset, buffer);
shortToBytes(as_type<ushort>(static_cast<half>(color.g)), bufferOffset + 2, buffer);
}
// R16G16_NORM
template <typename ShortType>
static inline float4 readR16G16_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset));
color.g = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 2));
color.b = 0.0;
color.a = 1.0;
return color;
}
#define readR16G16_SNORM readR16G16_NORM<short>
#define readR16G16_UNORM readR16G16_NORM<ushort>
// R16G16_SINT
static inline int4 readR16G16_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToShort<short>(buffer, bufferOffset);
color.g = bytesToShort<short>(buffer, bufferOffset + 2);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR16G16_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
shortToBytes(static_cast<short>(color.r), bufferOffset, buffer);
shortToBytes(static_cast<short>(color.g), bufferOffset + 2, buffer);
}
// R16G16_UINT
static inline uint4 readR16G16_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToShort<ushort>(buffer, bufferOffset);
color.g = bytesToShort<ushort>(buffer, bufferOffset + 2);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR16G16_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
shortToBytes(static_cast<ushort>(color.r), bufferOffset, buffer);
shortToBytes(static_cast<ushort>(color.g), bufferOffset + 2, buffer);
}
// R16G16B16_FLOAT
static inline float4 readR16G16B16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.g = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 2));
color.b = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 4));
color.a = 1.0;
return color;
}
// R16G16B16_NORM
template <typename ShortType>
static inline float4 readR16G16B16_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset));
color.g = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 2));
color.b = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 4));
color.a = 1.0;
return color;
}
#define readR16G16B16_SNORM readR16G16B16_NORM<short>
#define readR16G16B16_UNORM readR16G16B16_NORM<ushort>
// R16G16B16_SINT
static inline int4 readR16G16B16_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToShort<short>(buffer, bufferOffset);
color.g = bytesToShort<short>(buffer, bufferOffset + 2);
color.b = bytesToShort<short>(buffer, bufferOffset + 4);
color.a = 1;
return color;
}
// R16G16B16_UINT
static inline uint4 readR16G16B16_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToShort<ushort>(buffer, bufferOffset);
color.g = bytesToShort<ushort>(buffer, bufferOffset + 2);
color.b = bytesToShort<ushort>(buffer, bufferOffset + 4);
color.a = 1;
return color;
}
// R16G16B16A16_FLOAT
static inline float4 readR16G16B16A16_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset));
color.g = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 2));
color.b = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 4));
color.a = as_type<half>(bytesToShort<ushort>(buffer, bufferOffset + 6));
return color;
}
static inline void writeR16G16B16A16_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
shortToBytes(as_type<ushort>(static_cast<half>(color.r)), bufferOffset, buffer);
shortToBytes(as_type<ushort>(static_cast<half>(color.g)), bufferOffset + 2, buffer);
shortToBytes(as_type<ushort>(static_cast<half>(color.b)), bufferOffset + 4, buffer);
shortToBytes(as_type<ushort>(static_cast<half>(color.a)), bufferOffset + 6, buffer);
}
// R16G16B16A16_NORM
template <typename ShortType>
static inline float4 readR16G16B16A16_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset));
color.g = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 2));
color.b = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 4));
color.a = normalizedToFloat<ShortType>(bytesToShort<ShortType>(buffer, bufferOffset + 6));
return color;
}
#define readR16G16B16A16_SNORM readR16G16B16A16_NORM<short>
#define readR16G16B16A16_UNORM readR16G16B16A16_NORM<ushort>
// R16G16B16A16_SINT
static inline int4 readR16G16B16A16_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToShort<short>(buffer, bufferOffset);
color.g = bytesToShort<short>(buffer, bufferOffset + 2);
color.b = bytesToShort<short>(buffer, bufferOffset + 4);
color.a = bytesToShort<short>(buffer, bufferOffset + 6);
return color;
}
static inline void writeR16G16B16A16_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
shortToBytes(static_cast<short>(color.r), bufferOffset, buffer);
shortToBytes(static_cast<short>(color.g), bufferOffset + 2, buffer);
shortToBytes(static_cast<short>(color.b), bufferOffset + 4, buffer);
shortToBytes(static_cast<short>(color.a), bufferOffset + 6, buffer);
}
// R16G16B16A16_UINT
static inline uint4 readR16G16B16A16_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToShort<ushort>(buffer, bufferOffset);
color.g = bytesToShort<ushort>(buffer, bufferOffset + 2);
color.b = bytesToShort<ushort>(buffer, bufferOffset + 4);
color.a = bytesToShort<ushort>(buffer, bufferOffset + 6);
return color;
}
static inline void writeR16G16B16A16_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
shortToBytes(static_cast<ushort>(color.r), bufferOffset, buffer);
shortToBytes(static_cast<ushort>(color.g), bufferOffset + 2, buffer);
shortToBytes(static_cast<ushort>(color.b), bufferOffset + 4, buffer);
shortToBytes(static_cast<ushort>(color.a), bufferOffset + 6, buffer);
}
// R32_FLOAT
static inline float4 readR32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.r), bufferOffset, buffer);
}
// R32_NORM
template <typename IntType>
static inline float4 readR32_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset));
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
#define readR32_SNORM readR32_NORM<int>
#define readR32_UNORM readR32_NORM<uint>
// A32_FLOAT
static inline float4 readA32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.a = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.rgb = 0.0;
return color;
}
static inline void writeA32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.a), bufferOffset, buffer);
}
// L32_FLOAT
static inline float4 readL32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.a = 1.0;
return color;
}
static inline void writeL32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.r), bufferOffset, buffer);
}
// R32_SINT
static inline int4 readR32_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToInt<int>(buffer, bufferOffset);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR32_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
}
// R32_FIXED
static inline float4 readR32_FIXED(COMMON_READ_FUNC_PARAMS)
{
float4 color;
constexpr float kDivisor = 1.0f / (1 << 16);
color.r = bytesToInt<int>(buffer, bufferOffset) * kDivisor;
color.g = color.b = 0.0;
color.a = 1.0;
return color;
}
// R32_UINT
static inline uint4 readR32_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToInt<uint>(buffer, bufferOffset);
color.g = color.b = 0;
color.a = 1;
return color;
}
static inline void writeR32_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
}
// L32A32_FLOAT
static inline float4 readL32A32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.rgb = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.a = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 4));
return color;
}
static inline void writeL32A32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.r), bufferOffset, buffer);
intToBytes(as_type<uint>(color.a), bufferOffset + 4, buffer);
}
// R32G32_FLOAT
static inline float4 readR32G32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.g = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 4));
color.b = 0.0;
color.a = 1.0;
return color;
}
static inline void writeR32G32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.r), bufferOffset, buffer);
intToBytes(as_type<uint>(color.g), bufferOffset + 4, buffer);
}
// R32G32_NORM
template <typename IntType>
static inline float4 readR32G32_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset));
color.g = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 4));
color.b = 0.0;
color.a = 1.0;
return color;
}
#define readR32G32_SNORM readR32G32_NORM<int>
#define readR32G32_UNORM readR32G32_NORM<uint>
// R32G32_SINT
static inline int4 readR32G32_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToInt<int>(buffer, bufferOffset);
color.g = bytesToInt<int>(buffer, bufferOffset + 4);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR32G32_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
intToBytes(color.g, bufferOffset + 4, buffer);
}
// R32G32_FIXED
static inline float4 readR32G32_FIXED(COMMON_READ_FUNC_PARAMS)
{
float4 color;
constexpr float kDivisor = 1.0f / (1 << 16);
color.r = bytesToInt<int>(buffer, bufferOffset) * kDivisor;
color.g = bytesToInt<int>(buffer, bufferOffset + 4) * kDivisor;
color.b = 0.0;
color.a = 1.0;
return color;
}
// R32G32_UINT
static inline uint4 readR32G32_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToInt<uint>(buffer, bufferOffset);
color.g = bytesToInt<uint>(buffer, bufferOffset + 4);
color.b = 0;
color.a = 1;
return color;
}
static inline void writeR32G32_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
intToBytes(color.g, bufferOffset + 4, buffer);
}
// R32G32B32_FLOAT
static inline float4 readR32G32B32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.g = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 4));
color.b = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 8));
color.a = 1.0;
return color;
}
// R32G32B32_NORM
template <typename IntType>
static inline float4 readR32G32B32_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset));
color.g = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 4));
color.b = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 8));
color.a = 1.0;
return color;
}
#define readR32G32B32_SNORM readR32G32B32_NORM<int>
#define readR32G32B32_UNORM readR32G32B32_NORM<uint>
// R32G32B32_SINT
static inline int4 readR32G32B32_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToInt<int>(buffer, bufferOffset);
color.g = bytesToInt<int>(buffer, bufferOffset + 4);
color.b = bytesToInt<int>(buffer, bufferOffset + 8);
color.a = 1;
return color;
}
// R32G32B32_FIXED
static inline float4 readR32G32B32_FIXED(COMMON_READ_FUNC_PARAMS)
{
float4 color;
constexpr float kDivisor = 1.0f / (1 << 16);
color.r = bytesToInt<int>(buffer, bufferOffset) * kDivisor;
color.g = bytesToInt<int>(buffer, bufferOffset + 4) * kDivisor;
color.b = bytesToInt<int>(buffer, bufferOffset + 8) * kDivisor;
color.a = 1.0;
return color;
}
// R32G32B32_UINT
static inline uint4 readR32G32B32_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToInt<uint>(buffer, bufferOffset);
color.g = bytesToInt<uint>(buffer, bufferOffset + 4);
color.b = bytesToInt<uint>(buffer, bufferOffset + 8);
color.a = 1;
return color;
}
// R32G32B32A32_FLOAT
static inline float4 readR32G32B32A32_FLOAT(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = as_type<float>(bytesToInt<uint>(buffer, bufferOffset));
color.g = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 4));
color.b = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 8));
color.a = as_type<float>(bytesToInt<uint>(buffer, bufferOffset + 12));
return color;
}
static inline void writeR32G32B32A32_FLOAT(COMMON_WRITE_FLOAT_FUNC_PARAMS)
{
intToBytes(as_type<uint>(color.r), bufferOffset, buffer);
intToBytes(as_type<uint>(color.g), bufferOffset + 4, buffer);
intToBytes(as_type<uint>(color.b), bufferOffset + 8, buffer);
intToBytes(as_type<uint>(color.a), bufferOffset + 12, buffer);
}
// R32G32B32A32_NORM
template <typename IntType>
static inline float4 readR32G32B32A32_NORM(COMMON_READ_FUNC_PARAMS)
{
float4 color;
color.r = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset));
color.g = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 4));
color.b = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 8));
color.a = normalizedToFloat<IntType>(bytesToInt<IntType>(buffer, bufferOffset + 12));
return color;
}
#define readR32G32B32A32_SNORM readR32G32B32A32_NORM<int>
#define readR32G32B32A32_UNORM readR32G32B32A32_NORM<uint>
// R32G32B32A32_SINT
static inline int4 readR32G32B32A32_SINT(COMMON_READ_FUNC_PARAMS)
{
int4 color;
color.r = bytesToInt<int>(buffer, bufferOffset);
color.g = bytesToInt<int>(buffer, bufferOffset + 4);
color.b = bytesToInt<int>(buffer, bufferOffset + 8);
color.a = bytesToInt<int>(buffer, bufferOffset + 12);
return color;
}
static inline void writeR32G32B32A32_SINT(COMMON_WRITE_SINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
intToBytes(color.g, bufferOffset + 4, buffer);
intToBytes(color.b, bufferOffset + 8, buffer);
intToBytes(color.a, bufferOffset + 12, buffer);
}
// R32G32B32A32_FIXED
static inline float4 readR32G32B32A32_FIXED(COMMON_READ_FUNC_PARAMS)
{
float4 color;
constexpr float kDivisor = 1.0f / (1 << 16);
color.r = bytesToInt<int>(buffer, bufferOffset) * kDivisor;
color.g = bytesToInt<int>(buffer, bufferOffset + 4) * kDivisor;
color.b = bytesToInt<int>(buffer, bufferOffset + 8) * kDivisor;
color.a = bytesToInt<int>(buffer, bufferOffset + 12) * kDivisor;
return color;
}
// R32G32B32A32_UINT
static inline uint4 readR32G32B32A32_UINT(COMMON_READ_FUNC_PARAMS)
{
uint4 color;
color.r = bytesToInt<uint>(buffer, bufferOffset);
color.g = bytesToInt<uint>(buffer, bufferOffset + 4);
color.b = bytesToInt<uint>(buffer, bufferOffset + 8);
color.a = bytesToInt<uint>(buffer, bufferOffset + 12);
return color;
}
static inline void writeR32G32B32A32_UINT(COMMON_WRITE_UINT_FUNC_PARAMS)
{
intToBytes(color.r, bufferOffset, buffer);
intToBytes(color.g, bufferOffset + 4, buffer);
intToBytes(color.b, bufferOffset + 8, buffer);
intToBytes(color.a, bufferOffset + 12, buffer);
}
#define ALIAS_READ_SINT_FUNC(FORMAT) \
static inline int4 read##FORMAT##_SSCALED(COMMON_READ_FUNC_PARAMS) \
{ \
return read##FORMAT##_SINT(FORWARD_COMMON_READ_FUNC_PARAMS); \
}
#define ALIAS_READ_UINT_FUNC(FORMAT) \
static inline uint4 read##FORMAT##_USCALED(COMMON_READ_FUNC_PARAMS) \
{ \
return read##FORMAT##_UINT(FORWARD_COMMON_READ_FUNC_PARAMS); \
}
#define ALIAS_READ_INT_FUNC(FORMAT) \
ALIAS_READ_SINT_FUNC(FORMAT) \
ALIAS_READ_UINT_FUNC(FORMAT)
#define ALIAS_READ_INT_FUNCS(BITS) \
ALIAS_READ_INT_FUNC(R##BITS) \
ALIAS_READ_INT_FUNC(R##BITS##G##BITS) \
ALIAS_READ_INT_FUNC(R##BITS##G##BITS##B##BITS) \
ALIAS_READ_INT_FUNC(R##BITS##G##BITS##B##BITS##A##BITS)
ALIAS_READ_INT_FUNCS(8)
ALIAS_READ_INT_FUNCS(16)
ALIAS_READ_INT_FUNCS(32)
ALIAS_READ_INT_FUNC(R10G10B10A2)
// Copy pixels from buffer to texture
kernel void readFromBufferToFloatTexture(COMMON_READ_KERNEL_PARAMS(float))
{
READ_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R5G6B5_UNORM) \
PROC(R8G8B8A8_UNORM) \
PROC(R8G8B8A8_UNORM_SRGB) \
PROC(R8G8B8A8_SNORM) \
PROC(B8G8R8A8_UNORM) \
PROC(B8G8R8A8_UNORM_SRGB) \
PROC(R8G8B8_UNORM) \
PROC(R8G8B8_UNORM_SRGB) \
PROC(R8G8B8_SNORM) \
PROC(L8_UNORM) \
PROC(L8A8_UNORM) \
PROC(R5G5B5A1_UNORM) \
PROC(R4G4B4A4_UNORM) \
PROC(R8_UNORM) \
PROC(R8_SNORM) \
PROC(R8G8_UNORM) \
PROC(R8G8_SNORM) \
PROC(R16_FLOAT) \
PROC(A16_FLOAT) \
PROC(L16_FLOAT) \
PROC(L16A16_FLOAT) \
PROC(R16G16_FLOAT) \
PROC(R16G16B16_FLOAT) \
PROC(R16G16B16A16_FLOAT) \
PROC(R32_FLOAT) \
PROC(A32_FLOAT) \
PROC(L32_FLOAT) \
PROC(L32A32_FLOAT) \
PROC(R32G32_FLOAT) \
PROC(R32G32B32_FLOAT) \
PROC(R32G32B32A32_FLOAT)
uint bufferOffset = CALC_BUFFER_READ_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(READ_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
kernel void readFromBufferToIntTexture(COMMON_READ_KERNEL_PARAMS(int))
{
READ_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R8_SINT) \
PROC(R8G8_SINT) \
PROC(R8G8B8_SINT) \
PROC(R8G8B8A8_SINT) \
PROC(R16_SINT) \
PROC(R16G16_SINT) \
PROC(R16G16B16_SINT) \
PROC(R16G16B16A16_SINT) \
PROC(R32_SINT) \
PROC(R32G32_SINT) \
PROC(R32G32B32_SINT) \
PROC(R32G32B32A32_SINT)
uint bufferOffset = CALC_BUFFER_READ_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(READ_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
kernel void readFromBufferToUIntTexture(COMMON_READ_KERNEL_PARAMS(uint))
{
READ_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R8_UINT) \
PROC(R8G8_UINT) \
PROC(R8G8B8_UINT) \
PROC(R8G8B8A8_UINT) \
PROC(R16_UINT) \
PROC(R16G16_UINT) \
PROC(R16G16B16_UINT) \
PROC(R16G16B16A16_UINT) \
PROC(R32_UINT) \
PROC(R32G32_UINT) \
PROC(R32G32B32_UINT) \
PROC(R32G32B32A32_UINT)
uint bufferOffset = CALC_BUFFER_READ_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(READ_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
// Copy pixels from texture to buffer
kernel void writeFromFloatTextureToBuffer(COMMON_WRITE_KERNEL_PARAMS(float))
{
WRITE_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R5G6B5_UNORM) \
PROC(R8G8B8A8_UNORM) \
PROC(R8G8B8A8_UNORM_SRGB) \
PROC(R8G8B8A8_SNORM) \
PROC(B8G8R8A8_UNORM) \
PROC(B8G8R8A8_UNORM_SRGB) \
PROC(R8G8B8_UNORM) \
PROC(R8G8B8_UNORM_SRGB) \
PROC(R8G8B8_SNORM) \
PROC(L8_UNORM) \
PROC(A8_UNORM) \
PROC(L8A8_UNORM) \
PROC(R5G5B5A1_UNORM) \
PROC(R4G4B4A4_UNORM) \
PROC(R8_UNORM) \
PROC(R8_SNORM) \
PROC(R8G8_UNORM) \
PROC(R8G8_SNORM) \
PROC(R16_FLOAT) \
PROC(A16_FLOAT) \
PROC(L16_FLOAT) \
PROC(L16A16_FLOAT) \
PROC(R16G16_FLOAT) \
PROC(R16G16B16A16_FLOAT) \
PROC(R32_FLOAT) \
PROC(A32_FLOAT) \
PROC(L32_FLOAT) \
PROC(L32A32_FLOAT) \
PROC(R32G32_FLOAT) \
PROC(R32G32B32A32_FLOAT)
uint bufferOffset = CALC_BUFFER_WRITE_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(WRITE_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
kernel void writeFromIntTextureToBuffer(COMMON_WRITE_KERNEL_PARAMS(int))
{
WRITE_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R8_SINT) \
PROC(R8G8_SINT) \
PROC(R8G8B8A8_SINT) \
PROC(R16_SINT) \
PROC(R16G16_SINT) \
PROC(R16G16B16A16_SINT) \
PROC(R32_SINT) \
PROC(R32G32_SINT) \
PROC(R32G32B32A32_SINT)
uint bufferOffset = CALC_BUFFER_WRITE_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(WRITE_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
kernel void writeFromUIntTextureToBuffer(COMMON_WRITE_KERNEL_PARAMS(uint))
{
WRITE_KERNEL_GUARD
#define SUPPORTED_FORMATS(PROC) \
PROC(R8_UINT) \
PROC(R8G8_UINT) \
PROC(R8G8B8A8_UINT) \
PROC(R16_UINT) \
PROC(R16G16_UINT) \
PROC(R16G16B16A16_UINT) \
PROC(R32_UINT) \
PROC(R32G32_UINT) \
PROC(R32G32B32A32_UINT)
uint bufferOffset = CALC_BUFFER_WRITE_OFFSET(options.pixelSize);
switch (kCopyFormatType)
{
SUPPORTED_FORMATS(WRITE_FORMAT_SWITCH_CASE)
}
#undef SUPPORTED_FORMATS
}
src/libANGLE/renderer/metal/shaders/format_autogen.h 0 → 100644
// GENERATED FILE - DO NOT EDIT.
// Generated by gen_mtl_internal_shaders.py
//
// Copyright 2020 The ANGLE Project Authors. All rights reserved.
// Use of this source code is governed by a BSD-style license that can be
// found in the LICENSE file.
//
namespace rx
{
namespace mtl_shader
{
namespace FormatID
{
enum
{
NONE,
A16_FLOAT,
A1R5G5B5_UNORM,
A2R10G10B10_SINT_VERTEX,
A2R10G10B10_SNORM_VERTEX,
A2R10G10B10_SSCALED_VERTEX,
A2R10G10B10_UINT_VERTEX,
A2R10G10B10_UNORM_VERTEX,
A2R10G10B10_USCALED_VERTEX,
A32_FLOAT,
A8_UNORM,
ASTC_10x10_SRGB_BLOCK,
ASTC_10x10_UNORM_BLOCK,
ASTC_10x5_SRGB_BLOCK,
ASTC_10x5_UNORM_BLOCK,
ASTC_10x6_SRGB_BLOCK,
ASTC_10x6_UNORM_BLOCK,
ASTC_10x8_SRGB_BLOCK,
ASTC_10x8_UNORM_BLOCK,
ASTC_12x10_SRGB_BLOCK,
ASTC_12x10_UNORM_BLOCK,
ASTC_12x12_SRGB_BLOCK,
ASTC_12x12_UNORM_BLOCK,
ASTC_3x3x3_UNORM_BLOCK,
ASTC_3x3x3_UNORM_SRGB_BLOCK,
ASTC_4x3x3_UNORM_BLOCK,
ASTC_4x3x3_UNORM_SRGB_BLOCK,
ASTC_4x4_SRGB_BLOCK,
ASTC_4x4_UNORM_BLOCK,
ASTC_4x4x3_UNORM_BLOCK,
ASTC_4x4x3_UNORM_SRGB_BLOCK,
ASTC_4x4x4_UNORM_BLOCK,
ASTC_4x4x4_UNORM_SRGB_BLOCK,
ASTC_5x4_SRGB_BLOCK,
ASTC_5x4_UNORM_BLOCK,
ASTC_5x4x4_UNORM_BLOCK,
ASTC_5x4x4_UNORM_SRGB_BLOCK,
ASTC_5x5_SRGB_BLOCK,
ASTC_5x5_UNORM_BLOCK,
ASTC_5x5x4_UNORM_BLOCK,
ASTC_5x5x4_UNORM_SRGB_BLOCK,
ASTC_5x5x5_UNORM_BLOCK,
ASTC_5x5x5_UNORM_SRGB_BLOCK,
ASTC_6x5_SRGB_BLOCK,
ASTC_6x5_UNORM_BLOCK,
ASTC_6x5x5_UNORM_BLOCK,
ASTC_6x5x5_UNORM_SRGB_BLOCK,
ASTC_6x6_SRGB_BLOCK,
ASTC_6x6_UNORM_BLOCK,
ASTC_6x6x5_UNORM_BLOCK,
ASTC_6x6x5_UNORM_SRGB_BLOCK,
ASTC_6x6x6_UNORM_BLOCK,
ASTC_6x6x6_UNORM_SRGB_BLOCK,
ASTC_8x5_SRGB_BLOCK,
ASTC_8x5_UNORM_BLOCK,
ASTC_8x6_SRGB_BLOCK,
ASTC_8x6_UNORM_BLOCK,
ASTC_8x8_SRGB_BLOCK,
ASTC_8x8_UNORM_BLOCK,
B10G10R10A2_UNORM,
B4G4R4A4_UNORM,
B5G5R5A1_UNORM,
B5G6R5_UNORM,
B8G8R8A8_TYPELESS,
B8G8R8A8_TYPELESS_SRGB,
B8G8R8A8_UNORM,
B8G8R8A8_UNORM_SRGB,
B8G8R8X8_UNORM,
BC1_RGBA_UNORM_BLOCK,
BC1_RGBA_UNORM_SRGB_BLOCK,
BC1_RGB_UNORM_BLOCK,
BC1_RGB_UNORM_SRGB_BLOCK,
BC2_RGBA_UNORM_BLOCK,
BC2_RGBA_UNORM_SRGB_BLOCK,
BC3_RGBA_UNORM_BLOCK,
BC3_RGBA_UNORM_SRGB_BLOCK,
BC4_RED_SNORM_BLOCK,
BC4_RED_UNORM_BLOCK,
BC5_RG_SNORM_BLOCK,
BC5_RG_UNORM_BLOCK,
BPTC_RGBA_UNORM_BLOCK,
BPTC_RGB_SIGNED_FLOAT_BLOCK,
BPTC_RGB_UNSIGNED_FLOAT_BLOCK,
BPTC_SRGB_ALPHA_UNORM_BLOCK,
D16_UNORM,
D24_UNORM_S8_UINT,
D24_UNORM_X8_UINT,
D32_FLOAT,
D32_FLOAT_S8X24_UINT,
D32_UNORM,
EAC_R11G11_SNORM_BLOCK,
EAC_R11G11_UNORM_BLOCK,
EAC_R11_SNORM_BLOCK,
EAC_R11_UNORM_BLOCK,
ETC1_LOSSY_DECODE_R8G8B8_UNORM_BLOCK,
ETC1_R8G8B8_UNORM_BLOCK,
ETC2_R8G8B8A1_SRGB_BLOCK,
ETC2_R8G8B8A1_UNORM_BLOCK,
ETC2_R8G8B8A8_SRGB_BLOCK,
ETC2_R8G8B8A8_UNORM_BLOCK,
ETC2_R8G8B8_SRGB_BLOCK,
ETC2_R8G8B8_UNORM_BLOCK,
L16A16_FLOAT,
L16_FLOAT,
L32A32_FLOAT,
L32_FLOAT,
L8A8_UNORM,
L8_UNORM,
PVRTC1_RGBA_2BPP_UNORM_BLOCK,
PVRTC1_RGBA_2BPP_UNORM_SRGB_BLOCK,
PVRTC1_RGBA_4BPP_UNORM_BLOCK,
PVRTC1_RGBA_4BPP_UNORM_SRGB_BLOCK,
PVRTC1_RGB_2BPP_UNORM_BLOCK,
PVRTC1_RGB_2BPP_UNORM_SRGB_BLOCK,
PVRTC1_RGB_4BPP_UNORM_BLOCK,
PVRTC1_RGB_4BPP_UNORM_SRGB_BLOCK,
R10G10B10A2_SINT,
R10G10B10A2_SNORM,
R10G10B10A2_SSCALED,
R10G10B10A2_UINT,
R10G10B10A2_UNORM,
R10G10B10A2_USCALED,
R10G10B10X2_UNORM,
R11G11B10_FLOAT,
R16G16B16A16_FLOAT,
R16G16B16A16_SINT,
R16G16B16A16_SNORM,
R16G16B16A16_SSCALED,
R16G16B16A16_UINT,
R16G16B16A16_UNORM,
R16G16B16A16_USCALED,
R16G16B16_FLOAT,
R16G16B16_SINT,
R16G16B16_SNORM,
R16G16B16_SSCALED,
R16G16B16_UINT,
R16G16B16_UNORM,
R16G16B16_USCALED,
R16G16_FLOAT,
R16G16_SINT,
R16G16_SNORM,
R16G16_SSCALED,
R16G16_UINT,
R16G16_UNORM,
R16G16_USCALED,
R16_FLOAT,
R16_SINT,
R16_SNORM,
R16_SSCALED,
R16_UINT,
R16_UNORM,
R16_USCALED,
R32G32B32A32_FIXED,
R32G32B32A32_FLOAT,
R32G32B32A32_SINT,
R32G32B32A32_SNORM,
R32G32B32A32_SSCALED,
R32G32B32A32_UINT,
R32G32B32A32_UNORM,
R32G32B32A32_USCALED,
R32G32B32_FIXED,
R32G32B32_FLOAT,
R32G32B32_SINT,
R32G32B32_SNORM,
R32G32B32_SSCALED,
R32G32B32_UINT,
R32G32B32_UNORM,
R32G32B32_USCALED,
R32G32_FIXED,
R32G32_FLOAT,
R32G32_SINT,
R32G32_SNORM,
R32G32_SSCALED,
R32G32_UINT,
R32G32_UNORM,
R32G32_USCALED,
R32_FIXED,
R32_FLOAT,
R32_SINT,
R32_SNORM,
R32_SSCALED,
R32_UINT,
R32_UNORM,
R32_USCALED,
R4G4B4A4_UNORM,
R5G5B5A1_UNORM,
R5G6B5_UNORM,
R8G8B8A8_SINT,
R8G8B8A8_SNORM,
R8G8B8A8_SSCALED,
R8G8B8A8_TYPELESS,
R8G8B8A8_TYPELESS_SRGB,
R8G8B8A8_UINT,
R8G8B8A8_UNORM,
R8G8B8A8_UNORM_SRGB,
R8G8B8A8_USCALED,
R8G8B8_SINT,
R8G8B8_SNORM,
R8G8B8_SSCALED,
R8G8B8_UINT,
R8G8B8_UNORM,
R8G8B8_UNORM_SRGB,
R8G8B8_USCALED,
R8G8_SINT,
R8G8_SNORM,
R8G8_SSCALED,
R8G8_UINT,
R8G8_UNORM,
R8G8_USCALED,
R8_SINT,
R8_SNORM,
R8_SSCALED,
R8_UINT,
R8_UNORM,
R8_UNORM_SRGB,
R8_USCALED,
R9G9B9E5_SHAREDEXP,
S8_UINT,
X2R10G10B10_SINT_VERTEX,
X2R10G10B10_SNORM_VERTEX,
X2R10G10B10_SSCALED_VERTEX,
X2R10G10B10_UINT_VERTEX,
X2R10G10B10_UNORM_VERTEX,
X2R10G10B10_USCALED_VERTEX,
};
}
} // namespace mtl_shader
} // namespace rx
src/libANGLE/renderer/metal/shaders/gen_mtl_internal_shaders.py
...@@ -12,6 +12,9 @@ import sys ...@@ -12,6 +12,9 @@ import sys
import json import json
from datetime import datetime from datetime import datetime
sys.path.append('../..')
import angle_format
template_header_boilerplate = """// GENERATED FILE - DO NOT EDIT. template_header_boilerplate = """// GENERATED FILE - DO NOT EDIT.
// Generated by {script_name} // Generated by {script_name}
// //
...@@ -161,9 +164,35 @@ def gen_precompiled_shaders(mac_version, ios_version, variable_name, additional_ ...@@ -161,9 +164,35 @@ def gen_precompiled_shaders(mac_version, ios_version, variable_name, additional_
os.system('rm -rfv compiled/*.metallib') os.system('rm -rfv compiled/*.metallib')
def gen_shader_enums_code(angle_formats):
code = "namespace rx\n"
code += "{\n"
code += "namespace mtl_shader\n"
code += "{\n"
code += "\n"
code += "namespace FormatID\n"
code += "{\n"
code += "enum\n"
code += "{\n"
code += " NONE,\n"
for angle_format in sorted(angle_formats):
if angle_format == 'NONE': # NONE already moved to the beginning of enum declaration
continue
code += " " + angle_format + ",\n"
code += "};\n\n"
code += "}\n"
code += "\n"
code += "}\n"
code += "}\n"
return code
def main(): def main():
src_files = [ src_files = [
'blit.metal', 'clear.metal', 'gen_indices.metal', 'gen_mipmap.metal', 'visibility.metal' 'blit.metal', 'clear.metal', 'gen_indices.metal', 'gen_mipmap.metal', 'copy_buffer.metal',
'visibility.metal'
] ]
# yapf: disable # yapf: disable
...@@ -185,8 +214,9 @@ def main(): ...@@ -185,8 +214,9 @@ def main():
# auto_script parameters. # auto_script parameters.
if len(sys.argv) > 1: if len(sys.argv) > 1:
inputs = src_files + ['common.h', 'constants.h'] inputs = ['../../angle_format_map.json'] + src_files + ['common.h', 'constants.h']
outputs = ['compiled/mtl_default_shaders_autogen.inc'] + os_specific_autogen_files outputs = ['format_autogen.h', 'compiled/mtl_default_shaders_autogen.inc'
] + os_specific_autogen_files
if sys.argv[1] == 'inputs': if sys.argv[1] == 'inputs':
print ','.join(inputs) print ','.join(inputs)
...@@ -202,6 +232,15 @@ def main(): ...@@ -202,6 +232,15 @@ def main():
boilerplate_code = template_header_boilerplate.format( boilerplate_code = template_header_boilerplate.format(
script_name=sys.argv[0], copyright_year=datetime.today().year) script_name=sys.argv[0], copyright_year=datetime.today().year)
# -------- Generate shader constants -----------
angle_to_gl = angle_format.load_inverse_table('../../angle_format_map.json')
shader_formats_autogen = gen_shader_enums_code(angle_to_gl.keys())
shader_autogen_header = boilerplate_code + shader_formats_autogen
with open('format_autogen.h', 'wt') as out_file:
out_file.write(shader_autogen_header)
out_file.close()
# -------- Compile shaders ----------- # -------- Compile shaders -----------
# boilerplate code # boilerplate code
os.system("echo \"{0}\" > compiled/mtl_default_shaders_autogen.inc".format(boilerplate_code)) os.system("echo \"{0}\" > compiled/mtl_default_shaders_autogen.inc".format(boilerplate_code))
......
Markdown is supported
0% or
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment