xref: /aosp_15_r20/external/angle/src/libANGLE/renderer/metal/shaders/blit.metal (revision 8975f5c5ed3d1c378011245431ada316dfb6f244) 
1//
2// Copyright 2019 The ANGLE Project. All rights reserved.
3// Use of this source code is governed by a BSD-style license that can be
4// found in the LICENSE file.
5//
6// blit.metal: Implements blitting texture content to current frame buffer.
7
8#include "common.h"
9
10using namespace rx::mtl_shader;
11
12// function_constant(0) is already used by common.h
13constant bool kPremultiplyAlpha [[function_constant(1)]];
14constant bool kUnmultiplyAlpha [[function_constant(2)]];
15constant bool kTransformLinearToSrgb [[function_constant(3)]];
16constant int kSourceTextureType [[function_constant(4)]];   // Source color/depth texture type.
17constant int kSourceTexture2Type [[function_constant(5)]];  // Source stencil texture type.
18
19constant bool kSourceTextureType2D      = kSourceTextureType == kTextureType2D;
20constant bool kSourceTextureType2DArray = kSourceTextureType == kTextureType2DArray;
21constant bool kSourceTextureType2DMS    = kSourceTextureType == kTextureType2DMultisample;
22constant bool kSourceTextureTypeCube    = kSourceTextureType == kTextureTypeCube;
23constant bool kSourceTextureType3D      = kSourceTextureType == kTextureType3D;
24
25constant bool kSourceTexture2Type2D      = kSourceTexture2Type == kTextureType2D;
26constant bool kSourceTexture2Type2DArray = kSourceTexture2Type == kTextureType2DArray;
27constant bool kSourceTexture2Type2DMS    = kSourceTexture2Type == kTextureType2DMultisample;
28constant bool kSourceTexture2TypeCube    = kSourceTexture2Type == kTextureTypeCube;
29
30struct BlitParams
31{
32    // xy: lower left, zw: upper right
33    float4 srcTexCoords;
34    int srcLevel;  // Source texture level.
35    int srcLayer;  // Source texture layer.
36    bool dstLuminance;  // destination texture is luminance. Unused by depth & stencil blitting.
37    uint8_t padding[7];
38};
39
40struct BlitVSOut
41{
42    float4 position [[position]];
43    float2 texCoords [[center_no_perspective, user(locn1)]];
44};
45
46vertex BlitVSOut blitVS(unsigned int vid [[vertex_id]], constant BlitParams &options [[buffer(0)]])
47{
48    BlitVSOut output;
49    output.position.xy = select(float2(-1.0f), float2(1.0f), bool2(vid & uint2(2, 1)));
50    output.position.zw = float2(0.0, 1.0);
51    output.texCoords = select(options.srcTexCoords.xy, options.srcTexCoords.zw, bool2(vid & uint2(2, 1)));
52
53    return output;
54}
55
56template <typename SrcTexture2d>
57static uint2 getImageCoords(SrcTexture2d srcTexture, float2 texCoords)
58{
59    uint2 dimens(srcTexture.get_width(), srcTexture.get_height());
60    uint2 coords = uint2(texCoords * float2(dimens));
61
62    return coords;
63}
64
65template <typename T>
66static inline vec<T, 4> blitSampleTextureMS(texture2d_ms<T> srcTexture, float2 texCoords)
67{
68    uint2 coords = getImageCoords(srcTexture, texCoords);
69    return resolveTextureMS(srcTexture, coords);
70}
71
72template <typename T>
73static inline vec<T, 4> blitSampleTexture3D(texture3d<T> srcTexture,
74                                            sampler textureSampler,
75                                            float2 texCoords,
76                                            constant BlitParams &options)
77{
78    uint depth   = srcTexture.get_depth(options.srcLevel);
79    float zCoord = (float(options.srcLayer) + 0.5) / float(depth);
80
81    return srcTexture.sample(textureSampler, float3(texCoords, zCoord), level(options.srcLevel));
82}
83
84// clang-format off
85#define BLIT_COLOR_FS_PARAMS(TYPE)                                                               \
86    BlitVSOut input [[stage_in]],                                                                \
87    texture2d<TYPE> srcTexture2d [[texture(0), function_constant(kSourceTextureType2D)]],        \
88    texture2d_array<TYPE> srcTexture2dArray                                                      \
89    [[texture(0), function_constant(kSourceTextureType2DArray)]],                                \
90    texture2d_ms<TYPE> srcTexture2dMS [[texture(0), function_constant(kSourceTextureType2DMS)]], \
91    texturecube<TYPE> srcTextureCube [[texture(0), function_constant(kSourceTextureTypeCube)]],  \
92    texture3d<TYPE> srcTexture3d [[texture(0), function_constant(kSourceTextureType3D)]],        \
93    sampler textureSampler [[sampler(0)]],                                                       \
94    constant BlitParams &options [[buffer(0)]]
95// clang-format on
96
97#define FORWARD_BLIT_COLOR_FS_PARAMS                                                      \
98    input, srcTexture2d, srcTexture2dArray, srcTexture2dMS, srcTextureCube, srcTexture3d, \
99        textureSampler, options
100
101template <typename T>
102static inline vec<T, 4> blitReadTexture(BLIT_COLOR_FS_PARAMS(T))
103{
104    vec<T, 4> output;
105
106    switch (kSourceTextureType)
107    {
108        case kTextureType2D:
109            output = srcTexture2d.sample(textureSampler, input.texCoords, level(options.srcLevel));
110            break;
111        case kTextureType2DArray:
112            output = srcTexture2dArray.sample(textureSampler, input.texCoords, options.srcLayer,
113                                              level(options.srcLevel));
114            break;
115        case kTextureType2DMultisample:
116            output = blitSampleTextureMS(srcTexture2dMS, input.texCoords);
117            break;
118        case kTextureTypeCube:
119            output = srcTextureCube.sample(textureSampler,
120                                           cubeTexcoords(input.texCoords, options.srcLayer),
121                                           level(options.srcLevel));
122            break;
123        case kTextureType3D:
124            output = blitSampleTexture3D(srcTexture3d, textureSampler, input.texCoords, options);
125            break;
126    }
127
128    if (kTransformLinearToSrgb) {
129        output.x = linearToSRGB(output.x);
130        output.y = linearToSRGB(output.y);
131        output.z = linearToSRGB(output.z);
132    }
133    if (kUnmultiplyAlpha)
134    {
135        if (output.a != 0.0)
136        {
137            output.xyz /= output.a;
138        }
139    }
140    if (kPremultiplyAlpha)
141    {
142        output.xyz *= output.a;
143    }
144
145    if (options.dstLuminance)
146    {
147        output.g = output.b = output.r;
148    }
149
150    return output;
151}
152
153template <typename T>
154static inline MultipleColorOutputs<T> blitFS(BLIT_COLOR_FS_PARAMS(T))
155{
156    vec<T, 4> output = blitReadTexture(FORWARD_BLIT_COLOR_FS_PARAMS);
157
158    return toMultipleColorOutputs(output);
159}
160
161fragment MultipleColorOutputs<float> blitFloatFS(BLIT_COLOR_FS_PARAMS(float))
162{
163    return blitFS(FORWARD_BLIT_COLOR_FS_PARAMS);
164}
165fragment MultipleColorOutputs<int> blitIntFS(BLIT_COLOR_FS_PARAMS(int))
166{
167    return blitFS(FORWARD_BLIT_COLOR_FS_PARAMS);
168}
169fragment MultipleColorOutputs<uint> blitUIntFS(BLIT_COLOR_FS_PARAMS(uint))
170{
171    return blitFS(FORWARD_BLIT_COLOR_FS_PARAMS);
172}
173
174fragment MultipleColorOutputs<uint> copyTextureFloatToUIntFS(BLIT_COLOR_FS_PARAMS(float))
175{
176    float4 inputColor = blitReadTexture<>(FORWARD_BLIT_COLOR_FS_PARAMS);
177    uint4 output = uint4(inputColor * float4(255.0));
178
179    return toMultipleColorOutputs(output);
180}
181
182// Depth & stencil blitting.
183struct FragmentDepthOut
184{
185    float depth [[depth(any)]];
186};
187
188static inline float sampleDepth(
189    texture2d<float> srcTexture2d [[function_constant(kSourceTextureType2D)]],
190    texture2d_array<float> srcTexture2dArray [[function_constant(kSourceTextureType2DArray)]],
191    texture2d_ms<float> srcTexture2dMS [[function_constant(kSourceTextureType2DMS)]],
192    texturecube<float> srcTextureCube [[function_constant(kSourceTextureTypeCube)]],
193    float2 texCoords,
194    constant BlitParams &options)
195{
196    float4 output;
197
198    constexpr sampler textureSampler(mag_filter::nearest, min_filter::nearest);
199
200    switch (kSourceTextureType)
201    {
202        case kTextureType2D:
203            output = srcTexture2d.sample(textureSampler, texCoords, level(options.srcLevel));
204            break;
205        case kTextureType2DArray:
206            output = srcTexture2dArray.sample(textureSampler, texCoords, options.srcLayer,
207                                              level(options.srcLevel));
208            break;
209        case kTextureType2DMultisample:
210            // Always use sample 0 for depth resolve:
211            output = srcTexture2dMS.read(getImageCoords(srcTexture2dMS, texCoords), 0);
212            break;
213        case kTextureTypeCube:
214            output =
215                srcTextureCube.sample(textureSampler, cubeTexcoords(texCoords, options.srcLayer),
216                                      level(options.srcLevel));
217            break;
218    }
219
220    return output.r;
221}
222
223fragment FragmentDepthOut blitDepthFS(BlitVSOut input [[stage_in]],
224                                      texture2d<float> srcTexture2d
225                                      [[texture(0), function_constant(kSourceTextureType2D)]],
226                                      texture2d_array<float> srcTexture2dArray
227                                      [[texture(0), function_constant(kSourceTextureType2DArray)]],
228                                      texture2d_ms<float> srcTexture2dMS
229                                      [[texture(0), function_constant(kSourceTextureType2DMS)]],
230                                      texturecube<float> srcTextureCube
231                                      [[texture(0), function_constant(kSourceTextureTypeCube)]],
232                                      constant BlitParams &options [[buffer(0)]])
233{
234    FragmentDepthOut re;
235
236    re.depth = sampleDepth(srcTexture2d, srcTexture2dArray, srcTexture2dMS, srcTextureCube,
237                           input.texCoords, options);
238
239    return re;
240}
241
242static inline uint32_t sampleStencil(
243    texture2d<uint32_t> srcTexture2d [[function_constant(kSourceTexture2Type2D)]],
244    texture2d_array<uint32_t> srcTexture2dArray [[function_constant(kSourceTexture2Type2DArray)]],
245    texture2d_ms<uint32_t> srcTexture2dMS [[function_constant(kSourceTexture2Type2DMS)]],
246    texturecube<uint32_t> srcTextureCube [[function_constant(kSourceTexture2TypeCube)]],
247    float2 texCoords,
248    int srcLevel,
249    int srcLayer)
250{
251    uint4 output;
252    constexpr sampler textureSampler(mag_filter::nearest, min_filter::nearest);
253
254    switch (kSourceTexture2Type)
255    {
256        case kTextureType2D:
257            output = srcTexture2d.sample(textureSampler, texCoords, level(srcLevel));
258            break;
259        case kTextureType2DArray:
260            output = srcTexture2dArray.sample(textureSampler, texCoords, srcLayer, level(srcLevel));
261            break;
262        case kTextureType2DMultisample:
263            // Always use sample 0 for stencil resolve:
264            output = srcTexture2dMS.read(getImageCoords(srcTexture2dMS, texCoords), 0);
265            break;
266        case kTextureTypeCube:
267            output = srcTextureCube.sample(textureSampler, cubeTexcoords(texCoords, srcLayer),
268                                           level(srcLevel));
269            break;
270    }
271
272    return output.r;
273}
274
275// Write stencil to a buffer
276struct BlitStencilToBufferParams
277{
278    float2 srcStartTexCoords;
279    float2 srcTexCoordSteps;
280    int srcLevel;
281    int srcLayer;
282
283    uint2 dstSize;
284    uint dstBufferRowPitch;
285    // Is multisample resolve needed?
286    bool resolveMS;
287};
288
289kernel void blitStencilToBufferCS(ushort2 gIndices [[thread_position_in_grid]],
290                                  texture2d<uint32_t> srcTexture2d
291                                  [[texture(1), function_constant(kSourceTexture2Type2D)]],
292                                  texture2d_array<uint32_t> srcTexture2dArray
293                                  [[texture(1), function_constant(kSourceTexture2Type2DArray)]],
294                                  texture2d_ms<uint32_t> srcTexture2dMS
295                                  [[texture(1), function_constant(kSourceTexture2Type2DMS)]],
296                                  texturecube<uint32_t> srcTextureCube
297                                  [[texture(1), function_constant(kSourceTexture2TypeCube)]],
298                                  constant BlitStencilToBufferParams &options [[buffer(0)]],
299                                  device uchar *buffer [[buffer(1)]])
300{
301    if (gIndices.x >= options.dstSize.x || gIndices.y >= options.dstSize.y)
302    {
303        return;
304    }
305
306    float2 srcTexCoords = options.srcStartTexCoords + float2(gIndices) * options.srcTexCoordSteps;
307
308    if (kSourceTexture2Type == kTextureType2DMultisample && !options.resolveMS)
309    {
310        uint samples      = srcTexture2dMS.get_num_samples();
311        uint2 imageCoords = getImageCoords(srcTexture2dMS, srcTexCoords);
312        uint bufferOffset = options.dstBufferRowPitch * gIndices.y + samples * gIndices.x;
313
314        for (uint sample = 0; sample < samples; ++sample)
315        {
316            uint stencilPerSample         = srcTexture2dMS.read(imageCoords, sample).r;
317            buffer[bufferOffset + sample] = static_cast<uchar>(stencilPerSample);
318        }
319    }
320    else
321    {
322        uint32_t stencil =
323            sampleStencil(srcTexture2d, srcTexture2dArray, srcTexture2dMS, srcTextureCube,
324                          srcTexCoords, options.srcLevel, options.srcLayer);
325
326        buffer[options.dstBufferRowPitch * gIndices.y + gIndices.x] = static_cast<uchar>(stencil);
327    }
328}
329
330// Fragment's stencil output is only available since Metal 2.1
331@@#if __METAL_VERSION__ >= 210
332
333struct FragmentStencilOut
334{
335    uint32_t stencil [[stencil]];
336};
337
338struct FragmentDepthStencilOut
339{
340    float depth [[depth(any)]];
341    uint32_t stencil [[stencil]];
342};
343
344fragment FragmentStencilOut blitStencilFS(
345    BlitVSOut input [[stage_in]],
346    texture2d<uint32_t> srcTexture2d [[texture(1), function_constant(kSourceTexture2Type2D)]],
347    texture2d_array<uint32_t> srcTexture2dArray
348    [[texture(1), function_constant(kSourceTexture2Type2DArray)]],
349    texture2d_ms<uint32_t> srcTexture2dMS
350    [[texture(1), function_constant(kSourceTexture2Type2DMS)]],
351    texturecube<uint32_t> srcTextureCube [[texture(1), function_constant(kSourceTexture2TypeCube)]],
352    constant BlitParams &options [[buffer(0)]])
353{
354    FragmentStencilOut re;
355
356    re.stencil = sampleStencil(srcTexture2d, srcTexture2dArray, srcTexture2dMS, srcTextureCube,
357                               input.texCoords, options.srcLevel, options.srcLayer);
358
359    return re;
360}
361
362fragment FragmentDepthStencilOut blitDepthStencilFS(
363    BlitVSOut input [[stage_in]],
364    // Source depth texture
365    texture2d<float> srcDepthTexture2d [[texture(0), function_constant(kSourceTextureType2D)]],
366    texture2d_array<float> srcDepthTexture2dArray
367    [[texture(0), function_constant(kSourceTextureType2DArray)]],
368    texture2d_ms<float> srcDepthTexture2dMS
369    [[texture(0), function_constant(kSourceTextureType2DMS)]],
370    texturecube<float> srcDepthTextureCube
371    [[texture(0), function_constant(kSourceTextureTypeCube)]],
372
373    // Source stencil texture
374    texture2d<uint32_t> srcStencilTexture2d
375    [[texture(1), function_constant(kSourceTexture2Type2D)]],
376    texture2d_array<uint32_t> srcStencilTexture2dArray
377    [[texture(1), function_constant(kSourceTexture2Type2DArray)]],
378    texture2d_ms<uint32_t> srcStencilTexture2dMS
379    [[texture(1), function_constant(kSourceTexture2Type2DMS)]],
380    texturecube<uint32_t> srcStencilTextureCube
381    [[texture(1), function_constant(kSourceTexture2TypeCube)]],
382
383    constant BlitParams &options [[buffer(0)]])
384{
385    FragmentDepthStencilOut re;
386
387    re.depth = sampleDepth(srcDepthTexture2d, srcDepthTexture2dArray, srcDepthTexture2dMS,
388                           srcDepthTextureCube, input.texCoords, options);
389    re.stencil =
390        sampleStencil(srcStencilTexture2d, srcStencilTexture2dArray, srcStencilTexture2dMS,
391                      srcStencilTextureCube, input.texCoords, options.srcLevel, options.srcLayer);
392    return re;
393}
394@@#endif  // __METAL_VERSION__ >= 210
395