third_party/angle/src/libANGLE/renderer/metal/shaders/mtl_default_shaders_src_autogen.inc

// GENERATED FILE - DO NOT EDIT.
// Generated by gen_mtl_internal_shaders.py
//
// Copyright 2019 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.
//

// C++ string version of Metal default shaders for debug purpose.




constexpr char default_metallib_src[] = R"(
#include <metal_stdlib>
#include <simd/simd.h>
# 1 "master_source.metal"
# 1 "<built-in>" 1
# 1 "<built-in>" 3
# 374 "<built-in>" 3
# 1 "<command line>" 1
# 1 "<built-in>" 2
# 1 "master_source.metal" 2








# 1 "./clear.metal" 1







# 1 "./common.h" 1
# 22 "./common.h"
using namespace metal;


constant float2 gCorners[6] = {
    float2(-1.0f, 1.0f), float2(1.0f, -1.0f), float2(-1.0f, -1.0f),
    float2(-1.0f, 1.0f), float2(1.0f, 1.0f), float2(1.0f, -1.0f),
};



constant int gTexcoordsIndices[6] = {2, 1, 0, 2, 3, 1};

fragment float4 dummyFS()
{
    return float4(0, 0, 0, 0);
}
# 9 "./clear.metal" 2

struct ClearParams
{
    float4 clearColor;
    float clearDepth;
};

vertex float4 clearVS(unsigned int vid [[ vertex_id ]],
                      constant ClearParams &clearParams [[buffer(0)]])
{
    return float4(gCorners[vid], clearParams.clearDepth, 1.0);
}

fragment float4 clearFS(constant ClearParams &clearParams [[buffer(0)]])
{
    return clearParams.clearColor;
}
# 10 "master_source.metal" 2
# 1 "./blit.metal" 1
# 11 "./blit.metal"
struct BlitParams
{

    float2 srcTexCoords[4];
    int srcLevel;
    bool srcLuminance;
    bool dstFlipViewportY;
    bool dstLuminance;
};

struct BlitVSOut
{
    float4 position [[position]];
    float2 texCoords [[user(locn1)]];
};

vertex BlitVSOut blitVS(unsigned int vid [[ vertex_id ]],
                         constant BlitParams &options [[buffer(0)]])
{
    BlitVSOut output;
    output.position = float4(gCorners[vid], 0.0, 1.0);
    output.texCoords = options.srcTexCoords[gTexcoordsIndices[vid]];

    if (!options.dstFlipViewportY)
    {


        output.position.y = -output.position.y;
    }

    return output;
}

float4 blitSampleTexture(texture2d<float> srcTexture,
                     float2 texCoords,
                     constant BlitParams &options)
{
    constexpr sampler textureSampler(mag_filter::linear,
                                     min_filter::linear);
    float4 output = srcTexture.sample(textureSampler, texCoords, level(options.srcLevel));

    if (options.srcLuminance)
    {
        output.gb = float2(output.r, output.r);
    }

    return output;
}

float4 blitOutput(float4 color, constant BlitParams &options)
{
    float4 ret = color;

    if (options.dstLuminance)
    {
        ret.r = ret.g = ret.b = color.r;
    }

    return ret;
}

fragment float4 blitFS(BlitVSOut input [[stage_in]],
                       texture2d<float> srcTexture [[texture(0)]],
                       constant BlitParams &options [[buffer(0)]])
{
    return blitOutput(blitSampleTexture(srcTexture, input.texCoords, options), options);
}

fragment float4 blitPremultiplyAlphaFS(BlitVSOut input [[stage_in]],
                                       texture2d<float> srcTexture [[texture(0)]],
                                       constant BlitParams &options [[buffer(0)]])
{
    float4 output = blitSampleTexture(srcTexture, input.texCoords, options);
    output.xyz *= output.a;
    return blitOutput(output, options);
}

fragment float4 blitUnmultiplyAlphaFS(BlitVSOut input [[stage_in]],
                                      texture2d<float> srcTexture [[texture(0)]],
                                      constant BlitParams &options [[buffer(0)]])
{
    float4 output = blitSampleTexture(srcTexture, input.texCoords, options);
    if (output.a != 0.0)
    {
        output.xyz *= 1.0 / output.a;
    }
    return blitOutput(output, options);
}
# 11 "master_source.metal" 2
# 1 "./gen_indices.metal" 1








constant bool kSourceBufferAligned[[function_constant(0)]];
constant bool kSourceIndexIsU8[[function_constant(1)]];
constant bool kSourceIndexIsU16[[function_constant(2)]];
constant bool kSourceIndexIsU32[[function_constant(3)]];
constant bool kSourceBufferUnaligned = !kSourceBufferAligned;
constant bool kUseSourceBufferU8 = kSourceIndexIsU8 || kSourceBufferUnaligned;
constant bool kUseSourceBufferU16 = kSourceIndexIsU16 && kSourceBufferAligned;
constant bool kUseSourceBufferU32 = kSourceIndexIsU32 && kSourceBufferAligned;

struct IndexConversionParams
{
    uint32_t srcOffset;
    uint32_t indexCount;
};



inline ushort getIndexAligned(constant ushort *inputAligned, uint offset, uint idx)
{
    return inputAligned[offset / 2 + idx];
}
inline uint getIndexAligned(constant uint *inputAligned, uint offset, uint idx)
{
    return inputAligned[offset / 4 + idx];
}
inline uchar getIndexAligned(constant uchar *input, uint offset, uint idx)
{
    return input[offset + idx];
}
inline ushort getIndexUnalignedU16(constant uchar *input, uint offset, uint idx)
{
    ushort inputLo = input[offset + 2 * idx];
    ushort inputHi = input[offset + 2 * idx + 1];

    return inputLo | (inputHi << 8);
}
inline uint getIndexUnalignedU32(constant uchar *input, uint offset, uint idx)
{
    uint input0 = input[offset + 4 * idx];
    uint input1 = input[offset + 4 * idx + 1];
    uint input2 = input[offset + 4 * idx + 2];
    uint input3 = input[offset + 4 * idx + 3];

    return input0 | (input1 << 8) | (input2 << 16) | (input3 << 24);
}

kernel void convertIndexU8ToU16(uint idx[[thread_position_in_grid]],
                                constant IndexConversionParams &options[[buffer(0)]],
                                constant uchar *input[[buffer(1)]],
                                device ushort *output[[buffer(2)]])
{
    if (idx >= options.indexCount) { return; };
    output[idx] = getIndexAligned(input, options.srcOffset, idx);
}

kernel void convertIndexU16(
    uint idx[[thread_position_in_grid]],
    constant IndexConversionParams &options[[buffer(0)]],
    constant uchar *input[[ buffer(1), function_constant(kSourceBufferUnaligned) ]],
    constant ushort *inputAligned[[ buffer(1), function_constant(kSourceBufferAligned) ]],
    device ushort *output[[buffer(2)]])
{
    if (idx >= options.indexCount) { return; };

    ushort value;
    if (kSourceBufferAligned)
    {
        value = getIndexAligned(inputAligned, options.srcOffset, idx);
    }
    else
    {
        value = getIndexUnalignedU16(input, options.srcOffset, idx);
    }
    output[idx] = value;
}

kernel void convertIndexU32(
    uint idx[[thread_position_in_grid]],
    constant IndexConversionParams &options[[buffer(0)]],
    constant uchar *input[[ buffer(1), function_constant(kSourceBufferUnaligned) ]],
    constant uint *inputAligned[[ buffer(1), function_constant(kSourceBufferAligned) ]],
    device uint *output[[buffer(2)]])
{
    if (idx >= options.indexCount) { return; };

    uint value;
    if (kSourceBufferAligned)
    {
        value = getIndexAligned(inputAligned, options.srcOffset, idx);
    }
    else
    {
        value = getIndexUnalignedU32(input, options.srcOffset, idx);
    }
    output[idx] = value;
}

struct TriFanArrayParams
{
    uint firstVertex;
    uint vertexCountFrom3rd;
};
kernel void genTriFanIndicesFromArray(uint idx[[thread_position_in_grid]],
                                      constant TriFanArrayParams &options[[buffer(0)]],
                                      device uint *output[[buffer(2)]])
{
    if (idx >= options.vertexCountFrom3rd) { return; };

    uint vertexIdx = options.firstVertex + 2 + idx;

    output[3 * idx] = options.firstVertex;
    output[3 * idx + 1] = vertexIdx - 1;
    output[3 * idx + 2] = vertexIdx;
}

inline uint getIndexU32(uint offset,
                        uint idx,
                        constant uchar *inputU8[[function_constant(kUseSourceBufferU8)]],
                        constant ushort *inputU16[[function_constant(kUseSourceBufferU16)]],
                        constant uint *inputU32[[function_constant(kUseSourceBufferU32)]])
{
    if (kUseSourceBufferU8)
    {
        if (kSourceIndexIsU16)
        {
            return getIndexUnalignedU16(inputU8, offset, idx);
        }
        else if (kSourceIndexIsU32)
        {
            return getIndexUnalignedU32(inputU8, offset, idx);
        }
        return getIndexAligned(inputU8, offset, idx);
    }
    else if (kUseSourceBufferU16)
    {
        return getIndexAligned(inputU16, offset, idx);
    }
    else if (kUseSourceBufferU32)
    {
        return getIndexAligned(inputU32, offset, idx);
    }
    return 0;
}



kernel void genTriFanIndicesFromElements(
    uint idx[[thread_position_in_grid]],
    constant IndexConversionParams &options[[buffer(0)]],
    constant uchar *inputU8[[ buffer(1), function_constant(kUseSourceBufferU8) ]],
    constant ushort *inputU16[[ buffer(1), function_constant(kUseSourceBufferU16) ]],
    constant uint *inputU32[[ buffer(1), function_constant(kUseSourceBufferU32) ]],
    device uint *output[[buffer(2)]])
{
    if (idx >= options.indexCount) { return; };

    uint elemIdx = 2 + idx;

    output[3 * idx] = getIndexU32(options.srcOffset, 0, inputU8, inputU16, inputU32);
    output[3 * idx + 1] = getIndexU32(options.srcOffset, elemIdx - 1, inputU8, inputU16, inputU32);
    output[3 * idx + 2] = getIndexU32(options.srcOffset, elemIdx, inputU8, inputU16, inputU32);
}
# 12 "master_source.metal" 2

)";