เอาล่ะ ตอบคำถามของฉันเองที่นี่ตามคำแนะนำของผู้แสดงความคิดเห็น ลองเขียนเชเดอร์ของตัวเองดูสิ!
นี่คือรหัสเชเดอร์:
#include <metal_stdlib>
using namespace metal;
/*
hadamardProduct:
Perform an element-wise multiplication (hadamard product) of the two input matrices A and B, store the result in C
*/
kernel void hadamardProductKernel(
texture_buffer<float, access::read> A [[texture(0)]],
texture_buffer<float, access::read> B [[texture(1)]],
texture_buffer<float, access::write> C [[texture(2)]],
uint gid [[thread_position_in_grid]]) {
// C[i,j] = A[i,j] * B[i,j]
C.write(A.read(gid) * B.read(gid), gid);
}
และความรวดเร็วที่รันเชเดอร์บนเมทริกซ์ 4x4 สองตัว:
import Foundation
import Metal
import MetalKit
guard
let gpu = MTLCreateSystemDefaultDevice(),
let commandQueue = gpu.makeCommandQueue(),
let commandBuffer = commandQueue.makeCommandBuffer(),
let defaultLibrary = gpu.makeDefaultLibrary(),
let kernelFunction = defaultLibrary.makeFunction(name: "hadamardProductKernel")
else {exit(1)}
// Create the matrices to multiply (as row-major matrices)
var A:[Float] = [2,0,0,0,
0,2,0,0,
0,0,2,0,
0,0,0,2]
var B:[Float] = [1,0,0,0,
0,2,0,0,
0,0,3,0,
0,0,0,4]
let A_buffer = gpu.makeTexture(descriptor: MTLTextureDescriptor.textureBufferDescriptor(with: .r32Float,
width: 16,
resourceOptions: .storageModeManaged,
usage: .shaderRead))
let B_buffer = gpu.makeTexture(descriptor: MTLTextureDescriptor.textureBufferDescriptor(with: .r32Float,
width: 16,
resourceOptions: .storageModeManaged,
usage: .shaderRead))
let C_buffer = gpu.makeTexture(descriptor: MTLTextureDescriptor.textureBufferDescriptor(with: .r32Float,
width: 16,
resourceOptions: .storageModeManaged,
usage: .shaderWrite))
A_buffer?.replace(region: MTLRegionMake1D(0, 16),
mipmapLevel: 0,
withBytes: UnsafeRawPointer(A),
bytesPerRow: 64)
B_buffer?.replace(region: MTLRegionMake1D(0, 16),
mipmapLevel: 0,
withBytes: UnsafeRawPointer(B),
bytesPerRow: 64)
let computePipelineState = try gpu.makeComputePipelineState(function: kernelFunction)
let computeEncoder = commandBuffer.makeComputeCommandEncoder()
computeEncoder?.setComputePipelineState(computePipelineState)
computeEncoder?.setTexture(A_buffer, index: 0)
computeEncoder?.setTexture(B_buffer, index: 1)
computeEncoder?.setTexture(C_buffer, index: 2)
let threadGroupSize = MTLSize(width: 16, height: 1, depth: 1)
let threadGroupCount = MTLSize(width: 1, height: 1, depth: 1)
computeEncoder?.dispatchThreadgroups(threadGroupCount, threadsPerThreadgroup: threadGroupSize)
computeEncoder?.endEncoding()
commandBuffer.commit()
commandBuffer.waitUntilCompleted()
print("done")
ขอขอบคุณความคิดเห็นใด ๆ ที่เชื่อมโยงไปยังแหล่งข้อมูลเพื่อการเรียนรู้เพิ่มเติมเกี่ยวกับสิ่งนี้
person
mrplants
schedule
09.01.2019
MPSMatrix
ได้รับการสนับสนุนโดยMTLBuffer
ดังนั้นคุณจึงสามารถเขียนเคอร์เนลเล็กๆ น้อยๆ เพื่อทำการคูณได้ ฟังก์ชันเคอร์เนลจะใช้พารามิเตอร์บัฟเฟอร์สามตัว (สองเข้าหนึ่งออก) คุณอาจต้องระมัดระวังในการจัดทำดัชนีหากก้าวย่างของคุณไม่ตรงกับจำนวนคอลัมน์ของคุณ แต่นี่เป็นเรื่องเล็กน้อย (และ ขนานกันอย่างเขินอาย) - person warrenm   schedule 08.01.2019MPSImageMultiply
ด้วย - person Matthijs Hollemans   schedule 08.01.2019