Merge commit '7621e2f8dec938cf48181c8b10afc9b01f444e68' into beta

submodules/TelegramUI/Components/DustEffect/BUILD

@@ -0,0 +1,63 @@
+load("@build_bazel_rules_swift//swift:swift.bzl", "swift_library")
+
+load(
+    "@build_bazel_rules_apple//apple:resources.bzl",
+    "apple_resource_bundle",
+    "apple_resource_group",
+)
+load("//build-system/bazel-utils:plist_fragment.bzl",
+    "plist_fragment",
+)
+
+filegroup(
+    name = "DustEffectMetalSources",
+    srcs = glob([
+        "Metal/**/*.metal",
+    ]),
+    visibility = ["//visibility:public"],
+)
+
+plist_fragment(
+    name = "DustEffectMetalSourcesBundleInfoPlist",
+    extension = "plist",
+    template =
+    """
+    <key>CFBundleIdentifier</key>
+    <string>org.telegram.DustEffectMetalSources</string>
+    <key>CFBundleDevelopmentRegion</key>
+    <string>en</string>
+    <key>CFBundleName</key>
+    <string>DustEffect</string>
+    """
+)
+
+apple_resource_bundle(
+    name = "DustEffectMetalSourcesBundle",
+    infoplists = [
+        ":DustEffectMetalSourcesBundleInfoPlist",
+    ],
+    resources = [
+        ":DustEffectMetalSources",
+    ],
+)
+
+swift_library(
+    name = "DustEffect",
+    module_name = "DustEffect",
+    srcs = glob([
+        "Sources/**/*.swift",
+    ]),
+    copts = [
+        "-warnings-as-errors",
+    ],
+    data = [
+        ":DustEffectMetalSourcesBundle",
+    ],
+    deps = [
+        "//submodules/Display",
+        "//submodules/MetalEngine",
+    ],
+    visibility = [
+        "//visibility:public",
+    ],
+)

submodules/TelegramUI/Components/DustEffect/Metal/DustEffectShaders.metal

@@ -0,0 +1,172 @@
+#include <metal_stdlib>
+
+#include "loki_header.metal"
+
+using namespace metal;
+
+struct Rectangle {
+    float2 origin;
+    float2 size;
+};
+
+constant static float2 quadVertices[6] = {
+    float2(0.0, 0.0),
+    float2(1.0, 0.0),
+    float2(0.0, 1.0),
+    float2(1.0, 0.0),
+    float2(0.0, 1.0),
+    float2(1.0, 1.0)
+};
+
+struct QuadVertexOut {
+    float4 position [[position]];
+    float2 uv;
+    float alpha;
+};
+
+float2 mapLocalToScreenCoordinates(const device Rectangle &rect, const device float2 &size, float2 position) {
+    float2 result = float2(rect.origin.x + position.x / size.x * rect.size.x, rect.origin.y + position.y / size.y * rect.size.y);
+    result.x = -1.0 + result.x * 2.0;
+    result.y = -1.0 + result.y * 2.0;
+    
+    return result;
+}
+
+struct Particle {
+    packed_float2 offsetFromBasePosition;
+    packed_float2 velocity;
+    float lifetime;
+};
+
+kernel void dustEffectInitializeParticle(
+    device Particle *particles [[ buffer(0) ]],
+    uint gid [[ thread_position_in_grid ]]
+) {
+    Loki rng = Loki(gid);
+    
+    Particle particle;
+    particle.offsetFromBasePosition = packed_float2(0.0, 0.0);
+    
+    float direction = rng.rand() * (3.14159265 * 2.0);
+    float velocity = (0.1 + rng.rand() * (0.2 - 0.1)) * 420.0;
+    particle.velocity = packed_float2(cos(direction) * velocity, sin(direction) * velocity);
+    
+    particle.lifetime = 0.7 + rng.rand() * (1.5 - 0.7);
+    
+    particles[gid] = particle;
+}
+
+float particleEaseInWindowFunction(float t) {
+    return t;
+}
+
+float particleEaseInValueAt(float fraction, float t) {
+    float windowSize = 0.8;
+
+    float effectiveT = t;
+    float windowStartOffset = -windowSize;
+    float windowEndOffset = 1.0;
+
+    float windowPosition = (1.0 - fraction) * windowStartOffset + fraction * windowEndOffset;
+    float windowT = max(0.0, min(windowSize, effectiveT - windowPosition)) / windowSize;
+    float localT = 1.0 - particleEaseInWindowFunction(windowT);
+
+    return localT;
+}
+
+float2 grad(float2 z ) {
+    // 2D to 1D  (feel free to replace by some other)
+    int n = z.x + z.y * 11111.0;
+
+    // Hugo Elias hash (feel free to replace by another one)
+    n = (n << 13) ^ n;
+    n = (n * (n * n * 15731 + 789221) + 1376312589) >> 16;
+
+    // Perlin style vectors
+    n &= 7;
+    float2 gr = float2(n & 1, n >> 1) * 2.0 - 1.0;
+    return ( n>=6 ) ? float2(0.0, gr.x) :
+           ( n>=4 ) ? float2(gr.x, 0.0) :
+                              gr;
+}
+
+float noise(float2 p ) {
+    float2 i = float2(floor(p));
+    float2 f = fract(p);
+    
+    float2 u = f*f*(3.0-2.0*f); // feel free to replace by a quintic smoothstep instead
+
+    return mix( mix( dot( grad( i+float2(0,0) ), f-float2(0.0,0.0) ),
+                     dot( grad( i+float2(1,0) ), f-float2(1.0,0.0) ), u.x),
+                mix( dot( grad( i+float2(0,1) ), f-float2(0.0,1.0) ),
+                     dot( grad( i+float2(1,1) ), f-float2(1.0,1.0) ), u.x), u.y);
+}
+
+kernel void dustEffectUpdateParticle(
+    device Particle *particles [[ buffer(0) ]],
+    const device uint2 &size [[ buffer(1) ]],
+    const device float &phase [[ buffer(2) ]],
+    const device float &timeStep [[ buffer(3) ]],
+    uint gid [[ thread_position_in_grid ]]
+) {
+    uint count = size.x * size.y;
+    if (gid >= count) {
+        return;
+    }
+    
+    constexpr float easeInDuration = 0.8;
+    float effectFraction = max(0.0, min(easeInDuration, phase)) / easeInDuration;
+    
+    uint particleX = gid % size.x;
+    float particleXFraction = float(particleX) / float(size.x);
+    float particleFraction = particleEaseInValueAt(effectFraction, particleXFraction);
+    
+    Particle particle = particles[gid];
+    particle.offsetFromBasePosition += (particle.velocity * timeStep) * particleFraction;
+    
+    particle.velocity += float2(0.0, timeStep * 120.0) * particleFraction;
+    particle.lifetime = max(0.0, particle.lifetime - timeStep * particleFraction);
+    particles[gid] = particle;
+}
+
+vertex QuadVertexOut dustEffectVertex(
+    const device Rectangle &rect [[ buffer(0) ]],
+    const device float2 &size [[ buffer(1) ]],
+    const device uint2 &particleResolution [[ buffer(2) ]],
+    const device Particle *particles [[ buffer(3) ]],
+    unsigned int vid [[ vertex_id ]],
+    unsigned int particleId [[ instance_id ]]
+) {
+    QuadVertexOut out;
+    
+    float2 quadVertex = quadVertices[vid];
+    
+    uint particleIndexX = particleId % particleResolution.x;
+    uint particleIndexY = particleId / particleResolution.x;
+    
+    Particle particle = particles[particleId];
+    
+    float2 particleSize = size / float2(particleResolution);
+    
+    float2 topLeftPosition = float2(float(particleIndexX) * particleSize.x, float(particleIndexY) * particleSize.y);
+    out.uv = (topLeftPosition + quadVertex * particleSize) / size;
+    
+    topLeftPosition += particle.offsetFromBasePosition;
+    float2 position = topLeftPosition + quadVertex * particleSize;
+    
+    out.position = float4(mapLocalToScreenCoordinates(rect, size, position), 0.0, 1.0);
+    out.alpha = max(0.0, min(0.3, particle.lifetime) / 0.3);
+    
+    return out;
+}
+
+fragment half4 dustEffectFragment(
+    QuadVertexOut in [[stage_in]],
+    texture2d<half, access::sample> inTexture [[ texture(0) ]]
+) {
+    constexpr sampler sampler(coord::normalized, address::clamp_to_edge, filter::linear);
+    
+    half4 color = inTexture.sample(sampler, float2(in.uv.x, 1.0 - in.uv.y));
+    
+    return color * in.alpha;
+}

submodules/TelegramUI/Components/DustEffect/Metal/loki.metal

@@ -0,0 +1,54 @@
+#include <metal_stdlib>
+#include "loki_header.metal"
+
+unsigned Loki::TausStep(const unsigned z, const int s1, const int s2, const int s3, const unsigned M)
+{
+    unsigned b=(((z << s1) ^ z) >> s2);
+    return (((z & M) << s3) ^ b);
+}
+
+thread Loki::Loki(const unsigned seed1, const unsigned seed2, const unsigned seed3) {
+    unsigned seed = seed1 * 1099087573UL;
+    unsigned seedb = seed2 * 1099087573UL;
+    unsigned seedc = seed3 * 1099087573UL;
+
+    // Round 1: Randomise seed
+    unsigned z1 = TausStep(seed,13,19,12,429496729UL);
+    unsigned z2 = TausStep(seed,2,25,4,4294967288UL);
+    unsigned z3 = TausStep(seed,3,11,17,429496280UL);
+    unsigned z4 = (1664525*seed + 1013904223UL);
+
+    // Round 2: Randomise seed again using second seed
+    unsigned r1 = (z1^z2^z3^z4^seedb);
+
+    z1 = TausStep(r1,13,19,12,429496729UL);
+    z2 = TausStep(r1,2,25,4,4294967288UL);
+    z3 = TausStep(r1,3,11,17,429496280UL);
+    z4 = (1664525*r1 + 1013904223UL);
+
+    // Round 3: Randomise seed again using third seed
+    r1 = (z1^z2^z3^z4^seedc);
+
+    z1 = TausStep(r1,13,19,12,429496729UL);
+    z2 = TausStep(r1,2,25,4,4294967288UL);
+    z3 = TausStep(r1,3,11,17,429496280UL);
+    z4 = (1664525*r1 + 1013904223UL);
+
+    this->seed = (z1^z2^z3^z4) * 2.3283064365387e-10;
+}
+
+thread float Loki::rand() {
+    unsigned hashed_seed = this->seed * 1099087573UL;
+
+    unsigned z1 = TausStep(hashed_seed,13,19,12,429496729UL);
+    unsigned z2 = TausStep(hashed_seed,2,25,4,4294967288UL);
+    unsigned z3 = TausStep(hashed_seed,3,11,17,429496280UL);
+    unsigned z4 = (1664525*hashed_seed + 1013904223UL);
+
+    thread float old_seed = this->seed;
+
+    this->seed = (z1^z2^z3^z4) * 2.3283064365387e-10;
+
+    return old_seed;
+}
+

submodules/TelegramUI/Components/DustEffect/Metal/loki_header.metal

@@ -0,0 +1,45 @@
+/*
+ * Loki Random Number Generator
+ * Copyright (c) 2017 Youssef Victor All rights reserved.
+ *
+ *      Function                        Result
+ *      ------------------------------------------------------------------
+ *
+ *      TausStep                        Combined Tausworthe Generator or
+ *                                      Linear Feedback Shift Register (LFSR)
+ *                                      random number generator. This is a
+ *                                      helper method for rng, which uses
+ *                                      a hybrid approach combining LFSR with
+ *                                      a Linear Congruential Generator (LCG)
+ *                                      in order to produce random numbers with
+ *                                      periods of well over 2^121
+ *
+ *      rand                            A pseudo-random number based on the
+ *                                      method outlined in "Efficient
+ *                                      pseudo-random number generation
+ *                                      for monte-carlo simulations using
+ *                                      graphic processors" by Siddhant
+ *                                      Mohanty et al 2012.
+ *
+ */
+
+#include <metal_stdlib>
+using namespace metal;
+
+#ifndef LOKI
+#define LOKI
+
+
+class Loki {
+private:
+    thread float seed;
+    unsigned TausStep(const unsigned z, const int s1, const int s2, const int s3, const unsigned M);
+
+public:
+    thread Loki(const unsigned seed1, const unsigned seed2 = 1, const unsigned seed3 = 1);
+
+    thread float rand();
+};
+
+#endif
+

submodules/TelegramUI/Components/DustEffect/Sources/DustEffectLayer.swift

@@ -0,0 +1,360 @@
+import Foundation
+import UIKit
+import Display
+import MetalEngine
+import MetalKit
+
+#if DEBUG
+import os
+#endif
+
+#if DEBUG
+class Signposter {
+    func emitEvent(_ string: StaticString) {
+    }
+}
+
+@available(iOS 15.0, *)
+final class SignposterImpl: Signposter {
+    private let signposter = OSSignposter()
+    private let signpostId: OSSignpostID
+    
+    override init() {
+        self.signpostId = self.signposter.makeSignpostID()
+    }
+    
+    override func emitEvent(_ string: StaticString) {
+        self.signposter.emitEvent("Fetch complete.", id: self.signpostId)
+    }
+}
+
+let signposter: Signposter = {
+    if #available(iOS 15.0, *) {
+        return SignposterImpl()
+    } else {
+        return Signposter()
+    }
+}()
+
+#endif
+
+private final class BundleMarker: NSObject {
+}
+
+private var metalLibraryValue: MTLLibrary?
+func metalLibrary(device: MTLDevice) -> MTLLibrary? {
+    if let metalLibraryValue {
+        return metalLibraryValue
+    }
+    
+    let mainBundle = Bundle(for: BundleMarker.self)
+    guard let path = mainBundle.path(forResource: "DustEffectMetalSourcesBundle", ofType: "bundle") else {
+        return nil
+    }
+    guard let bundle = Bundle(path: path) else {
+        return nil
+    }
+    guard let library = try? device.makeDefaultLibrary(bundle: bundle) else {
+        return nil
+    }
+    
+    metalLibraryValue = library
+    return library
+}
+
+public final class DustEffectLayer: MetalEngineSubjectLayer, MetalEngineSubject {
+    public var internalData: MetalEngineSubjectInternalData?
+    
+    private final class Item {
+        let frame: CGRect
+        let texture: MTLTexture
+        
+        var phase: Float = 0
+        var particleBufferIsInitialized: Bool = false
+        var particleBuffer: SharedBuffer?
+        
+        init?(frame: CGRect, image: UIImage) {
+            self.frame = frame
+            
+            guard let cgImage = image.cgImage, let texture = try? MTKTextureLoader(device: MetalEngine.shared.device).newTexture(cgImage: cgImage, options: [.SRGB: false as NSNumber]) else {
+                return nil
+            }
+            self.texture = texture
+        }
+    }
+    
+    private final class RenderState: RenderToLayerState {
+        let pipelineState: MTLRenderPipelineState
+        
+        init?(device: MTLDevice) {
+            guard let library = metalLibrary(device: device) else {
+                return nil
+            }
+            guard let vertexFunction = library.makeFunction(name: "dustEffectVertex"), let fragmentFunction = library.makeFunction(name: "dustEffectFragment") else {
+                return nil
+            }
+            
+            let pipelineDescriptor = MTLRenderPipelineDescriptor()
+            pipelineDescriptor.vertexFunction = vertexFunction
+            pipelineDescriptor.fragmentFunction = fragmentFunction
+            pipelineDescriptor.colorAttachments[0].pixelFormat = .bgra8Unorm
+            pipelineDescriptor.colorAttachments[0].isBlendingEnabled = true
+            pipelineDescriptor.colorAttachments[0].rgbBlendOperation = .add
+            pipelineDescriptor.colorAttachments[0].alphaBlendOperation = .add
+            pipelineDescriptor.colorAttachments[0].sourceRGBBlendFactor = .one
+            pipelineDescriptor.colorAttachments[0].sourceAlphaBlendFactor = .one
+            pipelineDescriptor.colorAttachments[0].destinationRGBBlendFactor = .oneMinusSourceAlpha
+            pipelineDescriptor.colorAttachments[0].destinationAlphaBlendFactor = .one
+            guard let pipelineState = try? device.makeRenderPipelineState(descriptor: pipelineDescriptor) else {
+                return nil
+            }
+            self.pipelineState = pipelineState
+        }
+    }
+    
+    final class DustComputeState: ComputeState {
+        let computePipelineStateInitializeParticle: MTLComputePipelineState
+        let computePipelineStateUpdateParticle: MTLComputePipelineState
+        
+        required init?(device: MTLDevice) {
+            guard let library = metalLibrary(device: device) else {
+                return nil
+            }
+            
+            guard let functionDustEffectInitializeParticle = library.makeFunction(name: "dustEffectInitializeParticle") else {
+                return nil
+            }
+            guard let computePipelineStateInitializeParticle = try? device.makeComputePipelineState(function: functionDustEffectInitializeParticle) else {
+                return nil
+            }
+            self.computePipelineStateInitializeParticle = computePipelineStateInitializeParticle
+            
+            guard let functionDustEffectUpdateParticle = library.makeFunction(name: "dustEffectUpdateParticle") else {
+                return nil
+            }
+            guard let computePipelineStateUpdateParticle = try? device.makeComputePipelineState(function: functionDustEffectUpdateParticle) else {
+                return nil
+            }
+            
+            self.computePipelineStateUpdateParticle = computePipelineStateUpdateParticle
+        }
+    }
+    
+    private var updateLink: SharedDisplayLinkDriver.Link?
+    private var items: [Item] = []
+    private var lastTimeStep: Double = 0.0
+    
+    public var animationSpeed: Float = 1.0
+    public var playsBackwards: Bool = false
+    
+    public var becameEmpty: (() -> Void)?
+    
+    override public init() {
+        super.init()
+        
+        self.isOpaque = false
+        self.backgroundColor = nil
+        
+        self.didEnterHierarchy = { [weak self] in
+            guard let self else {
+                return
+            }
+            self.updateNeedsAnimation()
+        }
+        self.didExitHierarchy = { [weak self] in
+            guard let self else {
+                return
+            }
+            self.updateNeedsAnimation()
+        }
+    }
+    
+    override public init(layer: Any) {
+        super.init(layer: layer)
+    }
+    
+    required public init?(coder: NSCoder) {
+        fatalError("init(coder:) has not been implemented")
+    }
+    
+    private var lastUpdateTimestamp: Double?
+    
+    private func updateItems(deltaTime: Double) {
+        let timestamp = CACurrentMediaTime()
+        let localDeltaTime: Double
+        if let lastUpdateTimestamp = self.lastUpdateTimestamp {
+            localDeltaTime = timestamp - lastUpdateTimestamp
+        } else {
+            localDeltaTime = 0.0
+        }
+        self.lastUpdateTimestamp = timestamp
+        
+        let deltaTimeValue: Double
+        if localDeltaTime <= 0.001 || localDeltaTime >= 0.2 {
+            deltaTimeValue = deltaTime
+        } else {
+            deltaTimeValue = localDeltaTime
+        }
+        
+        self.lastTimeStep = deltaTimeValue
+        
+        var didRemoveItems = false
+        for i in (0 ..< self.items.count).reversed() {
+            self.items[i].phase += Float(deltaTimeValue) * self.animationSpeed / Float(UIView.animationDurationFactor())
+            
+            if self.items[i].phase >= 4.0 {
+                self.items.remove(at: i)
+                didRemoveItems = true
+            }
+        }
+        self.updateNeedsAnimation()
+        
+        if didRemoveItems && self.items.isEmpty {
+            self.becameEmpty?()
+        }
+    }
+    
+    private func updateNeedsAnimation() {
+        if !self.items.isEmpty && self.isInHierarchy {
+            if self.updateLink == nil {
+                self.updateLink = SharedDisplayLinkDriver.shared.add(framesPerSecond: .max, { [weak self] deltaTime in
+                    guard let self else {
+                        return
+                    }
+                    self.updateItems(deltaTime: deltaTime)
+                    self.setNeedsUpdate()
+                })
+            }
+        } else {
+            if self.updateLink != nil {
+                self.updateLink = nil
+            }
+        }
+    }
+    
+    public func addItem(frame: CGRect, image: UIImage) {
+        if let item = Item(frame: frame, image: image) {
+            self.items.append(item)
+            self.updateNeedsAnimation()
+            self.setNeedsUpdate()
+        }
+    }
+    
+    public func update(context: MetalEngineSubjectContext) {
+        if self.bounds.isEmpty {
+            return
+        }
+        
+        let containerSize = self.bounds.size
+        
+        for item in self.items {
+            var itemFrame = item.frame
+            itemFrame.origin.y = containerSize.height - itemFrame.maxY
+            
+            let particleColumnCount = Int(itemFrame.width)
+            let particleRowCount = Int(itemFrame.height)
+            let particleCount = particleColumnCount * particleRowCount
+            
+            if item.particleBuffer == nil {
+                if let particleBuffer = MetalEngine.shared.sharedBuffer(spec: BufferSpec(length: particleCount * 4 * (4 + 1))) {
+                    item.particleBuffer = particleBuffer
+                }
+            }
+        }
+        
+        let lastTimeStep = self.lastTimeStep
+        self.lastTimeStep = 0.0
+        
+        #if DEBUG
+        if lastTimeStep * 1000.0 >= 20.0 {
+            print("Animation Lag: \(lastTimeStep * 1000.0) ms")
+            signposter.emitEvent("AnimationLag")
+        }
+        #endif
+        
+        let _ = context.compute(state: DustComputeState.self, commands: { [weak self] commandBuffer, state in
+            guard let self else {
+                return
+            }
+            guard let computeEncoder = commandBuffer.makeComputeCommandEncoder() else {
+                return
+            }
+            
+            for item in self.items {
+                guard let particleBuffer = item.particleBuffer else {
+                    continue
+                }
+                
+                let itemFrame = item.frame
+                let particleColumnCount = Int(itemFrame.width)
+                let particleRowCount = Int(itemFrame.height)
+                
+                let threadgroupSize = MTLSize(width: 32, height: 1, depth: 1)
+                let threadgroupCount = MTLSize(width: (particleRowCount * particleColumnCount + threadgroupSize.width - 1) / threadgroupSize.width, height: 1, depth: 1)
+                
+                computeEncoder.setBuffer(particleBuffer.buffer, offset: 0, index: 0)
+                
+                if !item.particleBufferIsInitialized {
+                    item.particleBufferIsInitialized = true
+                    computeEncoder.setComputePipelineState(state.computePipelineStateInitializeParticle)
+                    computeEncoder.dispatchThreadgroups(threadgroupCount, threadsPerThreadgroup: threadgroupSize)
+                }
+                
+                if lastTimeStep != 0.0 {
+                    computeEncoder.setComputePipelineState(state.computePipelineStateUpdateParticle)
+                    var particleCount = SIMD2<UInt32>(UInt32(particleColumnCount), UInt32(particleRowCount))
+                    computeEncoder.setBytes(&particleCount, length: 4 * 2, index: 1)
+                    var phase = item.phase
+                    computeEncoder.setBytes(&phase, length: 4, index: 2)
+                    var timeStep: Float = Float(lastTimeStep) / Float(UIView.animationDurationFactor())
+                    timeStep *= 2.0
+                    computeEncoder.setBytes(&timeStep, length: 4, index: 3)
+                    computeEncoder.dispatchThreadgroups(threadgroupCount, threadsPerThreadgroup: threadgroupSize)
+                }
+            }
+            
+            computeEncoder.endEncoding()
+        })
+        
+        context.renderToLayer(spec: RenderLayerSpec(size: RenderSize(width: Int(self.bounds.width * 3.0), height: Int(self.bounds.height * 3.0))), state: RenderState.self, layer: self, commands: { [weak self] encoder, placement in
+            guard let self else {
+                return
+            }
+            
+            for item in self.items {
+                guard let particleBuffer = item.particleBuffer else {
+                    continue
+                }
+                
+                var itemFrame = item.frame
+                itemFrame.origin.y = containerSize.height - itemFrame.maxY
+                
+                let particleColumnCount = Int(itemFrame.width)
+                let particleRowCount = Int(itemFrame.height)
+                let particleCount = particleColumnCount * particleRowCount
+                
+                var effectiveRect = placement.effectiveRect
+                effectiveRect.origin.x += itemFrame.minX / containerSize.width * effectiveRect.width
+                effectiveRect.origin.y += itemFrame.minY / containerSize.height * effectiveRect.height
+                effectiveRect.size.width = itemFrame.width / containerSize.width * effectiveRect.width
+                effectiveRect.size.height = itemFrame.height / containerSize.height * effectiveRect.height
+                
+                var rect = SIMD4<Float>(Float(effectiveRect.minX), Float(effectiveRect.minY), Float(effectiveRect.width), Float(effectiveRect.height))
+                encoder.setVertexBytes(&rect, length: 4 * 4, index: 0)
+                
+                var size = SIMD2<Float>(Float(itemFrame.width), Float(itemFrame.height))
+                encoder.setVertexBytes(&size, length: 4 * 2, index: 1)
+                
+                var particleResolution = SIMD2<UInt32>(UInt32(particleColumnCount), UInt32(particleRowCount))
+                encoder.setVertexBytes(&particleResolution, length: 4 * 2, index: 2)
+                
+                encoder.setVertexBuffer(particleBuffer.buffer, offset: 0, index: 3)
+                
+                encoder.setFragmentTexture(item.texture, index: 0)
+                
+                encoder.drawPrimitives(type: .triangle, vertexStart: 0, vertexCount: 6)
+                encoder.drawPrimitives(type: .triangle, vertexStart: 0, vertexCount: 6, instanceCount: particleCount)
+            }
+        })
+    }
+}