diff --git a/ch10/ch10new.playground/Contents.swift b/ch10/ch10new.playground/Contents.swift new file mode 100644 index 0000000..9435df8 --- /dev/null +++ b/ch10/ch10new.playground/Contents.swift @@ -0,0 +1,103 @@ + +import MetalKit +import PlaygroundSupport + +guard let device = MTLCreateSystemDefaultDevice(), + let commandQueue = device.makeCommandQueue() else { + fatalError("Metal is not supported on this device") +} + +let textureDescriptor = MTLTextureDescriptor.texture2DDescriptor(pixelFormat: .rgba8Unorm, + width: 600, + height: 600, + mipmapped: false) +textureDescriptor.usage = [.shaderWrite, .shaderRead] +guard let texture = device.makeTexture(descriptor: textureDescriptor) else { + fatalError("Failed to create texture") +} + +let shaderSource = """ +#include +using namespace metal; + +float dist(float2 point, float2 center, float radius) +{ + return length(point - center) - radius; +} + +kernel void compute(texture2d output [[texture(0)]], + uint2 gid [[thread_position_in_grid]]) +{ + int width = output.get_width(); + int height = output.get_height(); + float2 uv = float2(gid) / float2(width, height); + uv = uv * 2.0 - 1.0; + float distToCircle = dist(uv, float2(0), 0.5); + float distToCircle2 = dist(uv, float2(-0.1, 0.1), 0.5); + bool inside = distToCircle2 < 0; + output.write(inside ? float4(0) : float4(1, 0.7, 0, 1) * (1 - distToCircle), gid); +} +""" + +guard let library = try? device.makeLibrary(source: shaderSource, options: nil), + let kernelFunction = library.makeFunction(name: "compute"), + let computePipelineState = try? device.makeComputePipelineState(function: kernelFunction) else { + fatalError("Failed to create compute pipeline") +} + +guard let commandBuffer = commandQueue.makeCommandBuffer(), + let computeEncoder = commandBuffer.makeComputeCommandEncoder() else { + fatalError("Failed to create command buffer or compute encoder") +} + +let w = computePipelineState.threadExecutionWidth +let h = computePipelineState.maxTotalThreadsPerThreadgroup / w +let threadsPerThreadgroup = MTLSize(width: w, height: h, depth: 1) +let threadgroupsPerGrid = MTLSize(width: (texture.width + w - 1) / w, + height: (texture.height + h - 1) / h, + depth: 1) +computeEncoder.setComputePipelineState(computePipelineState) +computeEncoder.setTexture(texture, index: 0) +computeEncoder.dispatchThreadgroups(threadgroupsPerGrid, threadsPerThreadgroup: threadsPerThreadgroup) +computeEncoder.endEncoding() +commandBuffer.commit() +commandBuffer.waitUntilCompleted() + +class MetalView: MTKView { + var texture: MTLTexture! + + init(frame frameRect: CGRect, device: MTLDevice?, texture: MTLTexture?) { + self.texture = texture + super.init(frame: frameRect, device: device) + self.isPaused = true + self.enableSetNeedsDisplay = true + self.drawableSize = CGSize(width: 600, height: 600) // Set the drawable size here + } + + required init(coder: NSCoder) { fatalError("init(coder:) has not been implemented") } + + override func draw(_ rect: CGRect) { + guard let drawable = self.currentDrawable else { return } + if drawable.texture.width != texture.width || drawable.texture.height != texture.height { + print("Drawable texture size does not match Metal texture size.") + return + } + guard let commandBuffer = device?.makeCommandQueue()?.makeCommandBuffer(), + let blitEncoder = commandBuffer.makeBlitCommandEncoder() else { return } + blitEncoder.copy(from: texture, + sourceSlice: 0, + sourceLevel: 0, + sourceOrigin: MTLOrigin(x: 0, y: 0, z: 0), + sourceSize: MTLSize(width: texture.width, height: texture.height, depth: 1), + to: drawable.texture, + destinationSlice: 0, + destinationLevel: 0, + destinationOrigin: MTLOrigin(x: 0, y: 0, z: 0)) + blitEncoder.endEncoding() + commandBuffer.present(drawable) + commandBuffer.commit() + } +} + +let view = MetalView(frame: CGRect(x: 0, y: 0, width: 600, height: 600), device: device, texture: texture) +PlaygroundPage.current.liveView = view diff --git a/ch10/ch10new.playground/Resources/Shaders.metal b/ch10/ch10new.playground/Resources/Shaders.metal new file mode 100644 index 0000000..6a1f371 --- /dev/null +++ b/ch10/ch10new.playground/Resources/Shaders.metal @@ -0,0 +1,21 @@ + +#include +using namespace metal; + +float dist(float2 point, float2 center, float radius) +{ + return length(point - center) - radius; +} + +kernel void compute(texture2d output [[texture(0)]], + uint2 gid [[thread_position_in_grid]]) +{ + int width = output.get_width(); + int height = output.get_height(); + float2 uv = float2(gid) / float2(width, height); + uv = uv * 2.0 - 1.0; + float distToCircle = dist(uv, float2(0), 0.5); + float distToCircle2 = dist(uv, float2(-0.1, 0.1), 0.5); + bool inside = distToCircle2 < 0; + output.write(inside ? float4(0) : float4(1, 0.7, 0, 1) * (1 - distToCircle), gid); +} diff --git a/ch10/ch10new.playground/Sources/MetalView.swift b/ch10/ch10new.playground/Sources/MetalView.swift new file mode 100644 index 0000000..abd99e3 --- /dev/null +++ b/ch10/ch10new.playground/Sources/MetalView.swift @@ -0,0 +1,48 @@ + +import MetalKit + +public class MetalView: NSObject, MTKViewDelegate { + + public var device: MTLDevice! + var queue: MTLCommandQueue! + var cps: MTLComputePipelineState! + + override public init() { + super.init() + registerShaders() + } + + func registerShaders() { + device = MTLCreateSystemDefaultDevice()! + queue = device.makeCommandQueue() + guard let path = Bundle.main.path(forResource: "Shaders", ofType: "metal") else { + print("No shader file found") + return + } + do { + let input = try String(contentsOfFile: path, encoding: String.Encoding.utf8) + let library = try device.makeLibrary(source: input, options: nil) + let kernel = library.makeFunction(name: "compute")! + cps = try device.makeComputePipelineState(function: kernel) + } catch let e { + Swift.print("\(e)") + } + } + + public func mtkView(_ view: MTKView, drawableSizeWillChange size: CGSize) {} + + public func draw(in view: MTKView) { + if let drawable = view.currentDrawable, + let commandBuffer = queue.makeCommandBuffer(), + let commandEncoder = commandBuffer.makeComputeCommandEncoder() { + commandEncoder.setComputePipelineState(cps) + commandEncoder.setTexture(drawable.texture, index: 0) + let threadGroupCount = MTLSizeMake(8, 8, 1) + let threadGroups = MTLSizeMake(drawable.texture.width / threadGroupCount.width, drawable.texture.height / threadGroupCount.height, 1) + commandEncoder.dispatchThreadgroups(threadGroups, threadsPerThreadgroup: threadGroupCount) + commandEncoder.endEncoding() + commandBuffer.present(drawable) + commandBuffer.commit() + } + } +} diff --git a/ch10/ch10new.playground/contents.xcplayground b/ch10/ch10new.playground/contents.xcplayground new file mode 100644 index 0000000..06828af --- /dev/null +++ b/ch10/ch10new.playground/contents.xcplayground @@ -0,0 +1,4 @@ + + + + \ No newline at end of file diff --git a/ch10/ch10new.playground/playground.xcworkspace/contents.xcworkspacedata b/ch10/ch10new.playground/playground.xcworkspace/contents.xcworkspacedata new file mode 100644 index 0000000..919434a --- /dev/null +++ b/ch10/ch10new.playground/playground.xcworkspace/contents.xcworkspacedata @@ -0,0 +1,7 @@ + + + + + diff --git a/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/marius.xcuserdatad/UserInterfaceState.xcuserstate b/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/marius.xcuserdatad/UserInterfaceState.xcuserstate new file mode 100644 index 0000000..f40c78e Binary files /dev/null and b/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/marius.xcuserdatad/UserInterfaceState.xcuserstate differ diff --git a/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/mhorga.xcuserdatad/UserInterfaceState.xcuserstate b/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/mhorga.xcuserdatad/UserInterfaceState.xcuserstate new file mode 100644 index 0000000..7c005c2 Binary files /dev/null and b/ch10/ch10new.playground/playground.xcworkspace/xcuserdata/mhorga.xcuserdatad/UserInterfaceState.xcuserstate differ