diff --git a/CMakeLists.txt b/CMakeLists.txt index 431f29a..eb3eec8 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -42,8 +42,10 @@ add_framework(Metal SOURCES src/Metal/Metal.mm src/Metal/MTLBuffer.mm + src/Metal/MTLArgument.mm src/Metal/MTLCaptureManager.mm src/Metal/MTLCommandBuffer.mm + src/Metal/MTLBlitCommandEncoder.mm src/Metal/MTLCommandQueue.mm src/Metal/MTLCompileOptions.mm src/Metal/MTLComputeCommandEncoder.mm @@ -97,6 +99,10 @@ if (BUILD_METAL) target_compile_definitions(Metal PUBLIC DARLING_METAL_ENABLED=1 ) + # MTLDevice.mm reads the Vulkan device properties from Indium's private headers. + target_link_libraries(Metal PRIVATE + indium_private + ) endif() # diff --git a/include/Metal/MTLCommandBuffer.h b/include/Metal/MTLCommandBuffer.h index ef43df8..9045a84 100644 --- a/include/Metal/MTLCommandBuffer.h +++ b/include/Metal/MTLCommandBuffer.h @@ -16,6 +16,7 @@ METAL_DECLARATIONS_BEGIN @protocol MTLComputeCommandEncoder; @protocol MTLCommandQueue; @protocol MTLRenderCommandEncoder; +@protocol MTLBlitCommandEncoder; @class MTLComputePassDescriptor; @class MTLRenderPassDescriptor; @@ -36,9 +37,11 @@ typedef void (^MTLCommandBufferHandler)(id); - (id)computeCommandEncoderWithDescriptor: (MTLComputePassDescriptor*)computePassDescriptor; - (id)computeCommandEncoderWithDispatchType: (MTLDispatchType)dispatchType; - (id)computeCommandEncoder; +- (id)blitCommandEncoder; - (id)renderCommandEncoderWithDescriptor: (MTLRenderPassDescriptor*)renderPassDescriptor; +- (void)addScheduledHandler:(MTLCommandBufferHandler)block; - (void)addCompletedHandler: (MTLCommandBufferHandler)block; - (void)waitUntilCompleted; - (void)presentDrawable: (id)drawable; diff --git a/include/Metal/MTLComputeCommandEncoder.h b/include/Metal/MTLComputeCommandEncoder.h index 06f4e51..a8e7436 100644 --- a/include/Metal/MTLComputeCommandEncoder.h +++ b/include/Metal/MTLComputeCommandEncoder.h @@ -13,6 +13,9 @@ METAL_DECLARATIONS_BEGIN @protocol MTLComputeCommandEncoder; @protocol MTLCounterSampleBuffer; @protocol MTLBuffer; +@protocol MTLTexture; +@protocol MTLSamplerState; +@protocol MTLDepthStencilState; @class MTLComputePassDescriptor; @class MTLComputePassSampleBufferAttachmentDescriptorArray; @@ -53,6 +56,9 @@ MTL_EXPORT @property(readonly) MTLDispatchType dispatchType; +- (void)setTexture:(id)texture atIndex:(NSUInteger)index; +- (void)setSamplerState:(id)sampler atIndex:(NSUInteger)index; +- (void)setSamplerState:(id)sampler lodMinClamp:(float)lodMinClamp lodMaxClamp:(float)lodMaxClamp atIndex:(NSUInteger)index; - (void)setComputePipelineState:(id)state; - (void)setBuffer: (id)buffer diff --git a/include/Metal/MTLDepthStencilDescriptor.h b/include/Metal/MTLDepthStencilDescriptor.h index f22a57d..6b6715d 100644 --- a/include/Metal/MTLDepthStencilDescriptor.h +++ b/include/Metal/MTLDepthStencilDescriptor.h @@ -3,10 +3,17 @@ #ifndef _METAL_MTLDEPTHSTENCILDESCRIPTOR_H_ #define _METAL_MTLDEPTHSTENCILDESCRIPTOR_H_ - -#import - -@interface MTLDepthStencilDescriptor : NSObject +#import +@protocol MTLDevice; +@interface MTLDepthStencilDescriptor : NSObject +@property(nonatomic) MTLCompareFunction depthCompareFunction; +@property(nonatomic) BOOL depthWriteEnabled; +@property(nonatomic, copy) MTLStencilDescriptor* frontFaceStencil; +@property(nonatomic, copy) MTLStencilDescriptor* backFaceStencil; +@property(nonatomic, copy) NSString* label; @end - -#endif // _METAL_MTLDEPTHSTENCILDESCRIPTOR_H_ +@protocol MTLDepthStencilState +@property(readonly) id device; +@property(readonly) NSString* label; +@end +#endif diff --git a/include/Metal/MTLDevice.h b/include/Metal/MTLDevice.h index a3efc96..299c038 100644 --- a/include/Metal/MTLDevice.h +++ b/include/Metal/MTLDevice.h @@ -18,9 +18,18 @@ METAL_DECLARATIONS_BEGIN @protocol MTLBuffer; @protocol MTLLibrary; @protocol MTLRenderPipelineState; +@protocol MTLSamplerState; +@protocol MTLDepthStencilState; +@class MTLDepthStencilDescriptor; +@class MTLSamplerDescriptor; +@class MTLTextureDescriptor; +@protocol MTLTexture; @class MTLComputePipelineDescriptor; -@class MTLAutoreleasedComputePipelineReflection; +@class MTLComputePipelineReflection; +@class MTLRenderPipelineReflection; +typedef MTLComputePipelineReflection* MTLAutoreleasedComputePipelineReflection; +typedef MTLRenderPipelineReflection* MTLAutoreleasedRenderPipelineReflection; @class MTLRenderPipelineDescriptor; typedef NS_OPTIONS(NSUInteger, MTLPipelineOption) { @@ -42,8 +51,31 @@ MTL_EXPORT NSArray>* MTLCopyAllDevices(void); MTL_EXPORT NSArray>* MTLCopyAllDevicesWithObserver(id* observer, MTLDeviceNotificationHandler handler); MTL_EXPORT void MTLRemoveDeviceObserver(id observer); +typedef NS_ENUM(NSInteger, MTLReadWriteTextureTier) { + MTLReadWriteTextureTierNone = 0, + MTLReadWriteTextureTier1 = 1, + MTLReadWriteTextureTier2 = 2, +}; + @protocol MTLDevice +@property (readonly) MTLReadWriteTextureTier readWriteTextureSupport; + +/* darling-lol M57: additional device queries ANGLE's Metal backend probes. */ +@property (readonly) BOOL hasUnifiedMemory; +@property (readonly) NSUInteger maxBufferLength; +@property (readonly) NSUInteger recommendedMaxWorkingSetSize; +@property (readonly) BOOL isLowPower; +@property (readonly) BOOL isRemovable; +@property (readonly) BOOL isHeadless; +@property (readonly) NSString* name; +@property (readonly) uint64_t registryID; +- (BOOL)isDepth24Stencil8PixelFormatSupported; +- (BOOL)supportsFamily: (NSUInteger)family; +- (BOOL)supportsFeatureSet: (NSUInteger)featureSet; +- (BOOL)supportsTextureSampleCount: (NSUInteger)sampleCount; +- (id)newCommandQueueWithMaxCommandBufferCount: (NSUInteger)maxCommandBufferCount; + - (id)newComputePipelineStateWithDescriptor: (MTLComputePipelineDescriptor*)descriptor options: (MTLPipelineOption)options reflection: (MTLAutoreleasedComputePipelineReflection*)reflection @@ -57,10 +89,14 @@ MTL_EXPORT void MTLRemoveDeviceObserver(id observer); reflection: (MTLAutoreleasedComputePipelineReflection*)reflection error: (NSError**)error; +- (id)newRenderPipelineStateWithDescriptor:(MTLRenderPipelineDescriptor*)descriptor options:(MTLPipelineOption)options reflection:(MTLAutoreleasedRenderPipelineReflection*)reflection error:(NSError**)error; - (id)newRenderPipelineStateWithDescriptor: (MTLRenderPipelineDescriptor*)descriptor error: (NSError**)error; +- (id)newTextureWithDescriptor:(MTLTextureDescriptor*)descriptor; - (id)newCommandQueue; +- (id)newDepthStencilStateWithDescriptor:(MTLDepthStencilDescriptor*)descriptor; +- (id)newSamplerStateWithDescriptor:(MTLSamplerDescriptor*)descriptor; - (id)newBufferWithLength: (NSUInteger)length options: (MTLResourceOptions)options; diff --git a/include/Metal/MTLRenderCommandEncoder.h b/include/Metal/MTLRenderCommandEncoder.h index 5a3845c..caeab7c 100644 --- a/include/Metal/MTLRenderCommandEncoder.h +++ b/include/Metal/MTLRenderCommandEncoder.h @@ -12,6 +12,9 @@ METAL_DECLARATIONS_BEGIN @protocol MTLBuffer; +@protocol MTLTexture; +@protocol MTLSamplerState; +@protocol MTLDepthStencilState; @protocol MTLRasterizationRateMap; @protocol MTLTexture; @protocol MTLRenderPipelineState; @@ -26,6 +29,13 @@ METAL_DECLARATIONS_BEGIN struct MTLSamplePosition; typedef struct MTLSamplePosition MTLSamplePosition; +typedef NS_OPTIONS(NSUInteger, MTLBarrierScope) { + MTLBarrierScopeBuffers = 1, MTLBarrierScopeTextures = 2, MTLBarrierScopeRenderTargets = 4, +}; +typedef NS_OPTIONS(NSUInteger, MTLRenderStages) { + MTLRenderStageVertex = 1, MTLRenderStageFragment = 2, MTLRenderStageTile = 4, + MTLRenderStageObject = 8, MTLRenderStageMesh = 16, +}; typedef NS_ENUM(NSUInteger, MTLLoadAction) { MTLLoadActionDontCare = 0, MTLLoadActionLoad = 1, @@ -98,10 +108,10 @@ typedef struct MTLViewport { } MTLViewport; typedef struct MTLScissorRect { - size_t height; - size_t width; - size_t x; - size_t y; + NSUInteger x; + NSUInteger y; + NSUInteger width; + NSUInteger height; } MTLScissorRect; NS_INLINE @@ -190,6 +200,14 @@ MTL_EXPORT @protocol MTLRenderCommandEncoder +- (void)setVertexTexture:(id)texture atIndex:(NSUInteger)index; +- (void)setFragmentTexture:(id)texture atIndex:(NSUInteger)index; +- (void)setVertexSamplerState:(id)sampler atIndex:(NSUInteger)index; +- (void)setFragmentSamplerState:(id)sampler atIndex:(NSUInteger)index; +- (void)setDepthBias:(float)depthBias slopeScale:(float)slopeScale clamp:(float)clamp; +- (void)setDepthStencilState:(id)state; +- (void)memoryBarrierWithScope:(MTLBarrierScope)scope afterStages:(MTLRenderStages)after beforeStages:(MTLRenderStages)before; +- (void)setStencilReferenceValue:(uint32_t)value; - (void)setRenderPipelineState: (id)pipelineState; - (void)setTriangleFillMode: (MTLTriangleFillMode)fillMode; - (void)setFrontFacingWinding: (MTLWinding)frontFacingWinding; diff --git a/include/Metal/MTLSamplerDescriptor.h b/include/Metal/MTLSamplerDescriptor.h index f049017..c8026d3 100644 --- a/include/Metal/MTLSamplerDescriptor.h +++ b/include/Metal/MTLSamplerDescriptor.h @@ -1,12 +1,46 @@ -// SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 - -#ifndef _METAL_MTLSAMPLERDESCRIPTOR_H_ -#define _METAL_MTLSAMPLERDESCRIPTOR_H_ - +#pragma once #import - -@interface MTLSamplerDescriptor : NSObject +typedef NS_ENUM(NSUInteger, MTLSamplerMinMagFilter) { + MTLSamplerMinMagFilterNearest = 0, + MTLSamplerMinMagFilterLinear = 1, +}; +typedef NS_ENUM(NSUInteger, MTLSamplerMipFilter) { + MTLSamplerMipFilterNotMipmapped = 0, + MTLSamplerMipFilterNearest = 1, + MTLSamplerMipFilterLinear = 2, +}; +typedef NS_ENUM(NSUInteger, MTLSamplerAddressMode) { + MTLSamplerAddressModeClampToEdge = 0, + MTLSamplerAddressModeMirrorClampToEdge = 1, + MTLSamplerAddressModeRepeat = 2, + MTLSamplerAddressModeMirrorRepeat = 3, + MTLSamplerAddressModeClampToZero = 4, + MTLSamplerAddressModeClampToBorderColor = 5, +}; +typedef NS_ENUM(NSUInteger, MTLSamplerBorderColor) { + MTLSamplerBorderColorTransparentBlack = 0, + MTLSamplerBorderColorOpaqueBlack = 1, + MTLSamplerBorderColorOpaqueWhite = 2, +}; +@protocol MTLDevice; +@protocol MTLSamplerState +@property(readonly) id device; +@property(readonly) NSString* label; +@end +@interface MTLSamplerDescriptor : NSObject +@property MTLSamplerMinMagFilter minFilter; +@property MTLSamplerMinMagFilter magFilter; +@property MTLSamplerMipFilter mipFilter; +@property NSUInteger maxAnisotropy; +@property MTLSamplerAddressMode sAddressMode; +@property MTLSamplerAddressMode tAddressMode; +@property MTLSamplerAddressMode rAddressMode; +@property MTLSamplerBorderColor borderColor; +@property BOOL normalizedCoordinates; +@property float lodMinClamp; +@property float lodMaxClamp; +@property BOOL supportArgumentBuffers; +@property NSUInteger compareFunction; +@property(copy) NSString* label; @end - -#endif // _METAL_MTLSAMPLERDESCRIPTOR_H_ diff --git a/include/Metal/MTLStencilDescriptor.h b/include/Metal/MTLStencilDescriptor.h index 9dd6f91..ebe0baf 100644 --- a/include/Metal/MTLStencilDescriptor.h +++ b/include/Metal/MTLStencilDescriptor.h @@ -3,10 +3,33 @@ #ifndef _METAL_MTLSTENCILDESCRIPTOR_H_ #define _METAL_MTLSTENCILDESCRIPTOR_H_ - #import - -@interface MTLStencilDescriptor : NSObject +typedef NS_ENUM(NSUInteger, MTLCompareFunction) { + MTLCompareFunctionNever = 0, + MTLCompareFunctionLess = 1, + MTLCompareFunctionEqual = 2, + MTLCompareFunctionLessEqual = 3, + MTLCompareFunctionGreater = 4, + MTLCompareFunctionNotEqual = 5, + MTLCompareFunctionGreaterEqual = 6, + MTLCompareFunctionAlways = 7, +}; +typedef NS_ENUM(NSUInteger, MTLStencilOperation) { + MTLStencilOperationKeep = 0, + MTLStencilOperationZero = 1, + MTLStencilOperationReplace = 2, + MTLStencilOperationIncrementClamp = 3, + MTLStencilOperationDecrementClamp = 4, + MTLStencilOperationInvert = 5, + MTLStencilOperationIncrementWrap = 6, + MTLStencilOperationDecrementWrap = 7, +}; +@interface MTLStencilDescriptor : NSObject +@property(nonatomic) MTLCompareFunction stencilCompareFunction; +@property(nonatomic) MTLStencilOperation stencilFailureOperation; +@property(nonatomic) MTLStencilOperation depthFailureOperation; +@property(nonatomic) MTLStencilOperation depthStencilPassOperation; +@property(nonatomic) uint32_t readMask; +@property(nonatomic) uint32_t writeMask; @end - -#endif // _METAL_MTLSTENCILDESCRIPTOR_H_ +#endif diff --git a/include/Metal/MTLTextureDescriptor.h b/include/Metal/MTLTextureDescriptor.h index 959d552..02e147f 100644 --- a/include/Metal/MTLTextureDescriptor.h +++ b/include/Metal/MTLTextureDescriptor.h @@ -1,12 +1,22 @@ -// SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 - -#ifndef _METAL_MTLTEXTUREDESCRIPTOR_H_ -#define _METAL_MTLTEXTUREDESCRIPTOR_H_ - -#import - -@interface MTLTextureDescriptor : NSObject +#pragma once +#import +@interface MTLTextureDescriptor : NSObject +@property MTLTextureType textureType; +@property MTLPixelFormat pixelFormat; +@property NSUInteger width; +@property NSUInteger height; +@property NSUInteger depth; +@property NSUInteger mipmapLevelCount; +@property NSUInteger sampleCount; +@property NSUInteger arrayLength; +@property MTLResourceOptions resourceOptions; +@property BOOL allowGPUOptimizedContents; +@property MTLTextureUsage usage; +@property MTLTextureSwizzleChannels swizzle; +@property MTLCPUCacheMode cpuCacheMode; +@property MTLStorageMode storageMode; +@property MTLHazardTrackingMode hazardTrackingMode; ++ (instancetype)texture2DDescriptorWithPixelFormat:(MTLPixelFormat)format width:(NSUInteger)width height:(NSUInteger)height mipmapped:(BOOL)mipmapped; ++ (instancetype)textureCubeDescriptorWithPixelFormat:(MTLPixelFormat)format size:(NSUInteger)size mipmapped:(BOOL)mipmapped; @end - -#endif // _METAL_MTLTEXTUREDESCRIPTOR_H_ diff --git a/include/Metal/MTLVertexDescriptor.h b/include/Metal/MTLVertexDescriptor.h index 2084d6f..366dfa7 100644 --- a/include/Metal/MTLVertexDescriptor.h +++ b/include/Metal/MTLVertexDescriptor.h @@ -3,10 +3,90 @@ #ifndef _METAL_MTLVERTEXDESCRIPTOR_H_ #define _METAL_MTLVERTEXDESCRIPTOR_H_ - #import - -@interface MTLVertexDescriptor : NSObject +typedef NS_ENUM(NSUInteger, MTLVertexFormat) { + MTLVertexFormatInvalid = 0, + MTLVertexFormatUChar2 = 1, + MTLVertexFormatUChar3 = 2, + MTLVertexFormatUChar4 = 3, + MTLVertexFormatChar2 = 4, + MTLVertexFormatChar3 = 5, + MTLVertexFormatChar4 = 6, + MTLVertexFormatUChar2Normalized = 7, + MTLVertexFormatUChar3Normalized = 8, + MTLVertexFormatUChar4Normalized = 9, + MTLVertexFormatChar2Normalized = 10, + MTLVertexFormatChar3Normalized = 11, + MTLVertexFormatChar4Normalized = 12, + MTLVertexFormatUShort2 = 13, + MTLVertexFormatUShort3 = 14, + MTLVertexFormatUShort4 = 15, + MTLVertexFormatShort2 = 16, + MTLVertexFormatShort3 = 17, + MTLVertexFormatShort4 = 18, + MTLVertexFormatUShort2Normalized = 19, + MTLVertexFormatUShort3Normalized = 20, + MTLVertexFormatUShort4Normalized = 21, + MTLVertexFormatShort2Normalized = 22, + MTLVertexFormatShort3Normalized = 23, + MTLVertexFormatShort4Normalized = 24, + MTLVertexFormatHalf2 = 25, + MTLVertexFormatHalf3 = 26, + MTLVertexFormatHalf4 = 27, + MTLVertexFormatFloat = 28, + MTLVertexFormatFloat2 = 29, + MTLVertexFormatFloat3 = 30, + MTLVertexFormatFloat4 = 31, + MTLVertexFormatInt = 32, + MTLVertexFormatInt2 = 33, + MTLVertexFormatInt3 = 34, + MTLVertexFormatInt4 = 35, + MTLVertexFormatUInt = 36, + MTLVertexFormatUInt2 = 37, + MTLVertexFormatUInt3 = 38, + MTLVertexFormatUInt4 = 39, + MTLVertexFormatInt1010102Normalized = 40, + MTLVertexFormatUInt1010102Normalized = 41, + MTLVertexFormatUChar4Normalized_BGRA = 42, + MTLVertexFormatUChar = 45, + MTLVertexFormatChar = 46, + MTLVertexFormatUCharNormalized = 47, + MTLVertexFormatCharNormalized = 48, + MTLVertexFormatUShort = 49, + MTLVertexFormatShort = 50, + MTLVertexFormatUShortNormalized = 51, + MTLVertexFormatShortNormalized = 52, + MTLVertexFormatHalf = 53, +}; +typedef NS_ENUM(NSUInteger, MTLVertexStepFunction) { + MTLVertexStepFunctionConstant = 0, + MTLVertexStepFunctionPerVertex = 1, + MTLVertexStepFunctionPerInstance = 2, + MTLVertexStepFunctionPerPatch = 3, + MTLVertexStepFunctionPerPatchControlPoint = 4, +}; +@interface MTLVertexBufferLayoutDescriptor : NSObject +@property(nonatomic) NSUInteger stride; +@property(nonatomic) MTLVertexStepFunction stepFunction; +@property(nonatomic) NSUInteger stepRate; @end - -#endif // _METAL_MTLVERTEXDESCRIPTOR_H_ +@interface MTLVertexBufferLayoutDescriptorArray : NSObject +- (MTLVertexBufferLayoutDescriptor*)objectAtIndexedSubscript:(NSUInteger)index; +- (void)setObject:(MTLVertexBufferLayoutDescriptor*)descriptor atIndexedSubscript:(NSUInteger)index; +@end +@interface MTLVertexAttributeDescriptor : NSObject +@property(nonatomic) MTLVertexFormat format; +@property(nonatomic) NSUInteger offset; +@property(nonatomic) NSUInteger bufferIndex; +@end +@interface MTLVertexAttributeDescriptorArray : NSObject +- (MTLVertexAttributeDescriptor*)objectAtIndexedSubscript:(NSUInteger)index; +- (void)setObject:(MTLVertexAttributeDescriptor*)descriptor atIndexedSubscript:(NSUInteger)index; +@end +@interface MTLVertexDescriptor : NSObject ++ (instancetype)vertexDescriptor; +@property(readonly) MTLVertexBufferLayoutDescriptorArray* layouts; +@property(readonly) MTLVertexAttributeDescriptorArray* attributes; +- (void)reset; +@end +#endif diff --git a/include/Metal/Metal.h b/include/Metal/Metal.h index 3c63a1d..a9c0dc7 100644 --- a/include/Metal/Metal.h +++ b/include/Metal/Metal.h @@ -5,6 +5,8 @@ #define _METAL_METAL_H_ #import +#import +#import #import #import #import diff --git a/private-include/Metal/MTLBufferInternal.h b/private-include/Metal/MTLBufferInternal.h index 0a620b2..79cf40a 100644 --- a/private-include/Metal/MTLBufferInternal.h +++ b/private-include/Metal/MTLBufferInternal.h @@ -13,7 +13,7 @@ @interface MTLBufferInternal : NSObject #if DARLING_METAL_ENABLED -@property(readonly) std::shared_ptr buffer; +@property(nonatomic, readonly) std::shared_ptr buffer; - (instancetype)initWithBuffer: (std::shared_ptr)buffer device: (id)device @@ -21,5 +21,8 @@ #endif @end +// Direct access for the encoders: the property getter is a message send that returns the shared_ptr by value. +const std::shared_ptr& MTLBufferGetIndium(id object); + #endif // _METAL_MTLBUFFERINTERNAL_H_ diff --git a/private-include/Metal/MTLRenderPipelineInternal.h b/private-include/Metal/MTLRenderPipelineInternal.h index 2463e97..750a245 100644 --- a/private-include/Metal/MTLRenderPipelineInternal.h +++ b/private-include/Metal/MTLRenderPipelineInternal.h @@ -39,7 +39,7 @@ METAL_DECLARATIONS_BEGIN @interface MTLRenderPipelineStateInternal : NSObject #if DARLING_METAL_ENABLED -@property(readonly) std::shared_ptr state; +@property(nonatomic, readonly) std::shared_ptr state; - (instancetype)initWithState: (std::shared_ptr)state device: (id)device @@ -47,6 +47,9 @@ METAL_DECLARATIONS_BEGIN #endif @end +// Direct access for the encoders: the property getter is a message send that returns the shared_ptr by value. +const std::shared_ptr& MTLRenderPipelineGetIndium(id object); + METAL_DECLARATIONS_END diff --git a/private-include/Metal/MTLTextureInternal.h b/private-include/Metal/MTLTextureInternal.h index 09572bf..f8e2977 100644 --- a/private-include/Metal/MTLTextureInternal.h +++ b/private-include/Metal/MTLTextureInternal.h @@ -37,7 +37,7 @@ MTL_EXPORT @interface MTLTextureInternal : NSObject #if DARLING_METAL_ENABLED -@property(readonly) std::shared_ptr texture; +@property(nonatomic, readonly) std::shared_ptr texture; - (instancetype)initWithTexture: (std::shared_ptr)texture device: (id)device @@ -45,5 +45,8 @@ MTL_EXPORT #endif @end +// Direct access for the encoders: the property getter is a message send that returns the shared_ptr by value. +const std::shared_ptr& MTLTextureGetIndium(id object); + #endif // _METAL_MTLTEXTUREINTERNAL_H_ diff --git a/src/Metal/MTLBuffer.mm b/src/Metal/MTLBuffer.mm index e3382a8..1182427 100644 --- a/src/Metal/MTLBuffer.mm +++ b/src/Metal/MTLBuffer.mm @@ -11,6 +11,10 @@ #if DARLING_METAL_ENABLED @synthesize buffer = _buffer; +const std::shared_ptr& MTLBufferGetIndium(id object) { + static const std::shared_ptr none; + return object ? ((MTLBufferInternal*) object)->_buffer : none; +} @synthesize device = _device; @synthesize resourceOptions = _resourceOptions; @synthesize label = _label; diff --git a/src/Metal/MTLCommandBuffer.mm b/src/Metal/MTLCommandBuffer.mm index 7dedaa0..5eca991 100644 --- a/src/Metal/MTLCommandBuffer.mm +++ b/src/Metal/MTLCommandBuffer.mm @@ -1,7 +1,18 @@ // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 +#import +#import +#import +#include +#include +#include +#include +#include +#include +#include #import +#import #import #import #import @@ -9,6 +20,199 @@ #import #import + +#if DARLING_METAL_ENABLED +// Diagnostic (MTLDebugEnabled only): remember the most recent render pass of a command buffer, then once the +// pass is finished read its first color target back and log how many draws it had and whether any pixel is lit. +struct MTLDebugPass { + id commandBuffer = nil; + id texture = nil; + unsigned long drawsAtStart = 0; + unsigned long ordinal = 0; + unsigned long load = 0; +}; +static thread_local MTLDebugPass debugPass; + +static void MTLDebugFlushPass(id commandBuffer) +{ + if (!debugPass.texture) { + return; + } + id texture = debugPass.texture; + debugPass.texture = nil; + if (debugPass.commandBuffer != commandBuffer) { + [texture release]; + return; + } + unsigned long draws = MTLTraceDrawTotal() - debugPass.drawsAtStart; + unsigned long ordinal = debugPass.ordinal, load = debugPass.load; + NSUInteger width = texture.width, height = texture.height; + unsigned long format = texture.pixelFormat; + bool readable = (format == 80 || format == 81 || format == 70 || format == 71) && width >= 64; + static unsigned long logged = 0; + // Full-screen passes are logged only at the start; every other size keeps being logged (the 3D scene passes come later) + static unsigned long smallLogged = 0, fullLogged = 0; + bool full = width == 1920; + bool interestingSmall = !full && smallLogged < 400; + if (interestingSmall) ++smallLogged; + bool skipFull = full && ++fullLogged > 60; + if (!readable || skipFull || (ordinal > 400 && width < 1280 && !interestingSmall) || (!interestingSmall && ++logged > 600)) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + if (ordinal <= 400) { + fprintf(log, "pass-end #%lu %lux%lu fmt=%lu draws=%lu (not read)\n", ordinal, (unsigned long) width, (unsigned long) height, format, draws); + } + fclose(log); + } + [texture release]; + return; + } + id readback = [[commandBuffer device] newBufferWithLength: width * height * 4 options: MTLResourceStorageModeShared]; + id blit = [commandBuffer blitCommandEncoder]; + [blit copyFromTexture: texture sourceSlice: 0 sourceLevel: 0 sourceOrigin: MTLOriginMake(0, 0, 0) + sourceSize: MTLSizeMake(width, height, 1) toBuffer: readback destinationOffset: 0 + destinationBytesPerRow: width * 4 destinationBytesPerImage: width * height * 4]; + [blit endEncoding]; + [commandBuffer addCompletedHandler: ^(id finished) { + const uint8_t* bytes = (const uint8_t*) readback.contents; + unsigned long lit = 0, maxChannel = 0; + for (NSUInteger i = 0; i < width * height * 4; i += 4) { + unsigned char top = std::max(bytes[i], std::max(bytes[i + 1], bytes[i + 2])); + if (top) ++lit; + if (top > maxChannel) maxChannel = top; + } + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "pass-end #%lu %lux%lu fmt=%lu load=%lu draws=%lu litPixels=%lu max=%lu\n", ordinal, + (unsigned long) width, (unsigned long) height, format, load, draws, lit, maxChannel); + fclose(log); + } + [readback release]; + [texture release]; + }]; +} +#endif + +#if DARLING_METAL_ENABLED +// Diagnostic (/tmp/metal-debug-timing): remember the render passes of a command buffer and report them when it is slow. +struct MTLTimingPass { + std::string description; + id encoder = nil; + unsigned long drawsAtStart = 0; + unsigned long draws = 0; +}; +static thread_local std::vector timingPasses; +static thread_local id timingCommandBuffer = nil; + +static void MTLTimingFinishPass() +{ + if (!timingPasses.empty() && timingPasses.back().encoder) { + MTLTimingPass& pass = timingPasses.back(); + pass.draws = MTLTraceDrawTotal() - pass.drawsAtStart; + NSString* label = [(id) pass.encoder label]; + if (label) pass.description += std::string(" '") + [label UTF8String] + "'"; + [pass.encoder release]; + pass.encoder = nil; + } +} +#endif + +#if DARLING_METAL_ENABLED +// Diagnostic snapshot: while /tmp/metal-snapshot exists, save the output of the render passes (as .ppm in the guest /tmp, +// first 60 per process) so each pass of a frame can be looked at. Cheap while the flag file is absent. +struct MTLSnapPass { + id commandBuffer = nil; + id texture = nil; + unsigned long drawsAtStart = 0; +}; +static thread_local MTLSnapPass snapPass; + +// The first line of /tmp/metal-snapshot names the capture; a new name starts a new capture of up to 30 passes. +static std::mutex snapTagLock; +static std::string snapTag; +static unsigned long snapSavedForTag = 0; + +static bool MTLSnapshotArmed() +{ + static std::atomic calls { 0 }; + static std::atomic armed { false }; + // re-read a few times a second: this runs for every encoder and commit, and a guest fopen is several host syscalls + if (calls.fetch_add(1) % 2048 == 0) { + bool present = false; + if (FILE* file = fopen("/tmp/metal-snapshot", "r")) { + char text[48] = {}; + if (fgets(text, sizeof(text), file)) { + std::string tag(text); + while (!tag.empty() && (tag.back() == '\n' || tag.back() == ' ')) tag.pop_back(); + std::lock_guard guard(snapTagLock); + if (tag != snapTag) { + snapTag = tag; + snapSavedForTag = 0; + } + } + present = true; + fclose(file); + } + armed = present; + } + return armed; +} + +static void MTLSnapFlush(id commandBuffer) +{ + if (!snapPass.texture) { + return; + } + id texture = snapPass.texture; + snapPass.texture = nil; + unsigned long draws = MTLTraceDrawTotal() - snapPass.drawsAtStart; + NSUInteger width = texture.width, height = texture.height; + unsigned long format = texture.pixelFormat; + bool readable = (format == 80 || format == 81 || format == 70 || format == 71) && width >= 16 && width <= 2048 && snapPass.commandBuffer == commandBuffer; + std::string tag; + unsigned long number = 0; + { + std::lock_guard guard(snapTagLock); + if (!readable || snapSavedForTag >= 30) { + readable = false; + } else { + number = snapSavedForTag++; + tag = snapTag; + } + } + if (!readable) { + [texture release]; + return; + } + id readback = [[commandBuffer device] newBufferWithLength: width * height * 4 options: MTLResourceStorageModeShared]; + id blit = [commandBuffer blitCommandEncoder]; + [blit copyFromTexture: texture sourceSlice: 0 sourceLevel: 0 sourceOrigin: MTLOriginMake(0, 0, 0) + sourceSize: MTLSizeMake(width, height, 1) toBuffer: readback destinationOffset: 0 + destinationBytesPerRow: width * 4 destinationBytesPerImage: width * height * 4]; + [blit endEncoding]; + [commandBuffer addCompletedHandler: ^(id finished) { + const uint8_t* bytes = (const uint8_t*) readback.contents; + char path[128]; + snprintf(path, sizeof(path), "/tmp/snap-%s-%02lu-%lux%lu-draws%lu-tex%p.ppm", tag.c_str(), number, (unsigned long) width, (unsigned long) height, draws, (void*) texture); + if (FILE* out = fopen(path, "wb")) { + fprintf(out, "P6\n%lu %lu\n255\n", (unsigned long) width, (unsigned long) height); + bool bgra = format == 80 || format == 81; + std::vector row(width * 3); + for (NSUInteger y = 0; y < height; ++y) { + for (NSUInteger x = 0; x < width; ++x) { + const uint8_t* px = bytes + (y * width + x) * 4; + row[x * 3 + 0] = bgra ? px[2] : px[0]; + row[x * 3 + 1] = px[1]; + row[x * 3 + 2] = bgra ? px[0] : px[2]; + } + fwrite(row.data(), 1, row.size(), out); + } + fclose(out); + } + [readback release]; + [texture release]; + }]; +} +#endif + #if DARLING_METAL_ENABLED // used to take care of RR while passing the block around in C++ code struct MTLCommandBufferHandlerWrapper { @@ -65,7 +269,7 @@ struct MTLCommandBufferHandlerWrapper { if (self != nil) { _commandBuffer = commandBuffer; _device = [commandQueue.device retain]; - _commandQueue = [_commandQueue retain]; + _commandQueue = [commandQueue retain]; } return self; } @@ -78,6 +282,13 @@ struct MTLCommandBufferHandlerWrapper { [super dealloc]; } +- (id)blitCommandEncoder +{ + auto encoder = _commandBuffer->blitCommandEncoder(); + if (!encoder) return nil; + return [[[MTLBlitCommandEncoderInternal alloc] initWithEncoder:encoder device:_device] autorelease]; +} + - (id)computeCommandEncoderWithDescriptor: (MTLComputePassDescriptor*)computePassDescriptor { auto encoder = _commandBuffer->computeCommandEncoder(computePassDescriptor.asIndiumDescriptor); @@ -107,11 +318,91 @@ struct MTLCommandBufferHandlerWrapper { - (id)renderCommandEncoderWithDescriptor: (MTLRenderPassDescriptor*)renderPassDescriptor { + MTLTraceCount("renderPass"); + if (MTLDebugEnabled()) { + MTLRenderPassColorAttachmentDescriptor* sizeColor = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + char name[64]; + snprintf(name, sizeof(name), "pass_%lux%lu_f%lu", sizeColor.texture ? (unsigned long) sizeColor.texture.width : 0UL, sizeColor.texture ? (unsigned long) sizeColor.texture.height : 0UL, sizeColor.texture ? (unsigned long) sizeColor.texture.pixelFormat : 0UL); + MTLTraceCount(name); + } + if (MTLDebugEnabled()) { + MTLDebugFlushPass(self); + } + // Diagnostic: with /tmp/metal-debug-magenta present, every pass clears its first color target to magenta + if (MTLDebugEnabled() && access("/tmp/metal-debug-magenta", F_OK) == 0) { + MTLRenderPassColorAttachmentDescriptor* debugColor = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + if (debugColor.texture && debugColor.loadAction == MTLLoadActionClear) { + debugColor.clearColor = MTLClearColorMake(1, 0, 1, 1); + } + } + { + // Diagnostic: describe the first render passes (and every 500th) in the trace log + static unsigned long passes = 0; + ++passes; + if (MTLDebugEnabled() && (passes <= 40 || passes % 500 == 0)) { + MTLRenderPassColorAttachmentDescriptor* color = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + id texture = color.texture; + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "pass #%lu color0: tex=%s %lux%lu fmt=%lu load=%lu store=%lu clear=(%.2f %.2f %.2f %.2f) depth=%d\n", passes, + texture ? "yes" : "none", texture ? (unsigned long) texture.width : 0UL, texture ? (unsigned long) texture.height : 0UL, + texture ? (unsigned long) texture.pixelFormat : 0UL, (unsigned long) color.loadAction, (unsigned long) color.storeAction, + color.clearColor.red, color.clearColor.green, color.clearColor.blue, color.clearColor.alpha, + renderPassDescriptor.depthAttachment.texture ? 1 : 0); + fclose(log); + } + } + } auto encoder = _commandBuffer->renderCommandEncoder([renderPassDescriptor asIndiumDescriptor]); if (!encoder) { return nil; } - return [[[MTLRenderCommandEncoderInternal alloc] initWithEncoder: encoder device: _device] autorelease]; + if (MTLDebugEnabled()) { + MTLRenderPassColorAttachmentDescriptor* color = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + static unsigned long ordinal = 0; + debugPass.commandBuffer = self; + debugPass.texture = [color.texture retain]; + debugPass.drawsAtStart = MTLTraceDrawTotal(); + debugPass.ordinal = ++ordinal; + debugPass.load = (unsigned long) color.loadAction; + } + id result = [[[MTLRenderCommandEncoderInternal alloc] initWithEncoder: encoder device: _device] autorelease]; + if (MTLTimingEnabled()) { + MTLTimingFinishPass(); + if (timingCommandBuffer != self) { + timingPasses.clear(); + timingCommandBuffer = self; + } + MTLRenderPassColorAttachmentDescriptor* color = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + char text[160]; + snprintf(text, sizeof(text), "render %lux%lu fmt=%lu load=%lu depth=%d", color.texture ? (unsigned long) color.texture.width : 0UL, color.texture ? (unsigned long) color.texture.height : 0UL, + color.texture ? (unsigned long) color.texture.pixelFormat : 0UL, (unsigned long) color.loadAction, renderPassDescriptor.depthAttachment.texture ? 1 : 0); + { + char sizeName[64]; + snprintf(sizeName, sizeof(sizeName), "pass_%lux%lu_f%lu_load%lu", color.texture ? (unsigned long) color.texture.width : 0UL, color.texture ? (unsigned long) color.texture.height : 0UL, + color.texture ? (unsigned long) color.texture.pixelFormat : 0UL, (unsigned long) color.loadAction); + MTLTraceTiming(sizeName, 0); + } + MTLTimingPass pass; + pass.description = text; + pass.encoder = [result retain]; + pass.drawsAtStart = MTLTraceDrawTotal(); + timingPasses.push_back(pass); + } + if (MTLSnapshotArmed()) { + MTLSnapFlush(self); + MTLRenderPassColorAttachmentDescriptor* snapColor = [renderPassDescriptor.colorAttachments objectAtIndexedSubscript: 0]; + if (snapColor.texture) { + snapPass.commandBuffer = self; + snapPass.texture = [snapColor.texture retain]; + snapPass.drawsAtStart = MTLTraceDrawTotal(); + } + } + return result; +} + +- (void)addScheduledHandler:(MTLCommandBufferHandler)block +{ + _commandBuffer->addScheduledHandler(MTLCommandBufferHandlerWrapper(block, self)); } - (void)addCompletedHandler: (MTLCommandBufferHandler)block @@ -121,17 +412,107 @@ struct MTLCommandBufferHandlerWrapper { - (void)waitUntilCompleted { + double waitStart = MTLTimingEnabled() ? MTLTraceNow() : 0; _commandBuffer->waitUntilCompleted(); + if (MTLTimingEnabled()) { + MTLTraceTiming("waitUntilCompleted", MTLTraceNow() - waitStart); + } } - (void)presentDrawable: (id)drawable { + if (MTLDebugEnabled()) { + static unsigned long presents = 0; + id drawableTexture = [(id)drawable respondsToSelector: @selector(texture)] ? [(id)drawable texture] : nil; + if (++presents <= 40) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "present #%lu lastPassTexture=%p drawableTexture=%p same=%d\n", presents, debugPass.texture, drawableTexture, debugPass.texture == drawableTexture); + fclose(log); + } + } + MTLDebugFlushPass(self); + } + // Diagnostic: with /tmp/metal-debug-present-magenta present, paint the drawable magenta right before it is + // presented, to tell a broken presentation path from content that is genuinely black. + if (MTLDebugEnabled() && access("/tmp/metal-debug-present-magenta", F_OK) == 0 && [(id)drawable respondsToSelector: @selector(texture)]) { + MTLRenderPassDescriptor* debugPass = [MTLRenderPassDescriptor renderPassDescriptor]; + MTLRenderPassColorAttachmentDescriptor* debugColor = [debugPass.colorAttachments objectAtIndexedSubscript: 0]; + debugColor.texture = [(id)drawable texture]; + debugColor.loadAction = MTLLoadActionClear; + debugColor.storeAction = MTLStoreActionStore; + debugColor.clearColor = MTLClearColorMake(1, 0, 1, 1); + id debugEncoder = [self renderCommandEncoderWithDescriptor: debugPass]; + [debugEncoder endEncoding]; + + // ... and read the drawable back on the Vulkan side, to see what it holds before it crosses to GL + static unsigned long debugPresents = 0; + ++debugPresents; + if (debugPresents <= 5 || debugPresents % 200 == 0) { + id texture = debugColor.texture; + NSUInteger width = texture.width, height = texture.height; + id readback = [[self device] newBufferWithLength: width * height * 4 options: MTLResourceStorageModeShared]; + id debugBlit = [self blitCommandEncoder]; + [debugBlit copyFromTexture: texture sourceSlice: 0 sourceLevel: 0 sourceOrigin: MTLOriginMake(0, 0, 0) + sourceSize: MTLSizeMake(width, height, 1) toBuffer: readback destinationOffset: 0 + destinationBytesPerRow: width * 4 destinationBytesPerImage: width * height * 4]; + [debugBlit endEncoding]; + unsigned long ordinal = debugPresents; + [self addCompletedHandler: ^(id finished) { + const uint8_t* bytes = (const uint8_t*) readback.contents; + unsigned long nonZero = 0; + for (NSUInteger i = 0; i < width * height * 4; i += 4) { + if (bytes[i] | bytes[i + 1] | bytes[i + 2]) ++nonZero; + } + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "vulkan-readback #%lu %lux%lu nonZeroPixels=%lu first=(%u %u %u %u) fmt=%lu\n", ordinal, + (unsigned long) width, (unsigned long) height, nonZero, bytes[0], bytes[1], bytes[2], bytes[3], (unsigned long) texture.pixelFormat); + fclose(log); + } + [readback release]; + }]; + } + } _commandBuffer->presentDrawable(((id)drawable).drawable); } - (void)commit { + MTLTraceCount("commit"); + if (MTLSnapshotArmed()) { + MTLSnapFlush(self); + } + double commitStart = MTLTimingEnabled() ? MTLTraceNow() : 0; + if (MTLTimingEnabled()) { + std::string report; + if (timingCommandBuffer == self) { + MTLTimingFinishPass(); + for (const auto& pass: timingPasses) { + char draws[32]; + snprintf(draws, sizeof(draws), " draws=%lu; ", pass.draws); + report += pass.description + draws; + } + timingPasses.clear(); + timingCommandBuffer = nil; + } + _commandBuffer->addCompletedHandler(MTLCommandBufferHandlerWrapper(^(id finished) { + double elapsed = MTLTraceNow() - commitStart; + MTLTraceTiming("gpuCommitToComplete", elapsed); + static std::atomic reported { 0 }; + if (elapsed > 0.3 && reported.fetch_add(1) < 40) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "slow-cmdbuf %.0fms label=%s passes: %s\n", elapsed * 1000, [[finished label] UTF8String] ? [[finished label] UTF8String] : "-", report.c_str()); + fclose(log); + } + } + }, self)); + } + if (MTLDebugEnabled()) { + MTLDebugFlushPass(self); + } _commandBuffer->commit(); + if (MTLTimingEnabled()) { + MTLTraceTiming("commitCall", MTLTraceNow() - commitStart); + } } #else diff --git a/src/Metal/MTLComputeCommandEncoder.mm b/src/Metal/MTLComputeCommandEncoder.mm index 208a0e9..b13ad20 100644 --- a/src/Metal/MTLComputeCommandEncoder.mm +++ b/src/Metal/MTLComputeCommandEncoder.mm @@ -1,3 +1,6 @@ +#import +#import +#import // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 @@ -196,6 +199,15 @@ MTL_UNSUPPORTED_CLASS [super dealloc]; } +- (void)setTexture:(id)texture atIndex:(NSUInteger)index { + _encoder->setTexture(((MTLTextureInternal*)texture).texture, index); +} +- (void)setSamplerState:(id)sampler atIndex:(NSUInteger)index { + _encoder->setSamplerState(((MTLSamplerStateInternal*)sampler).sampler, index); +} +- (void)setSamplerState:(id)sampler lodMinClamp:(float)lodMinClamp lodMaxClamp:(float)lodMaxClamp atIndex:(NSUInteger)index { + _encoder->setSamplerState(((MTLSamplerStateInternal*)sampler).sampler, lodMinClamp, lodMaxClamp, index); +} - (void)endEncoding { _encoder->endEncoding(); @@ -249,6 +261,16 @@ MTL_UNSUPPORTED_CLASS - (void)dispatchThreadgroups: (MTLSize)threadgroupsPerGrid threadsPerThreadgroup: (MTLSize)threadsPerThreadgroup { + MTLTraceCount("dispatch"); + if (MTLTimingEnabled() || MTLDebugEnabled()) { + static unsigned long logged = 0; + if (++logged <= 40) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "dispatch #%lu groups=(%lu %lu %lu) threads=(%lu %lu %lu)\n", logged, (unsigned long) threadgroupsPerGrid.width, (unsigned long) threadgroupsPerGrid.height, (unsigned long) threadgroupsPerGrid.depth, (unsigned long) threadsPerThreadgroup.width, (unsigned long) threadsPerThreadgroup.height, (unsigned long) threadsPerThreadgroup.depth); + fclose(log); + } + } + } _encoder->dispatchThreadgroups(MTLSizeToIndium(threadgroupsPerGrid), MTLSizeToIndium(threadsPerThreadgroup)); } diff --git a/src/Metal/MTLDepthStencilDescriptor.mm b/src/Metal/MTLDepthStencilDescriptor.mm index 2a57796..0cbedf2 100644 --- a/src/Metal/MTLDepthStencilDescriptor.mm +++ b/src/Metal/MTLDepthStencilDescriptor.mm @@ -1,24 +1,58 @@ // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 -#import -#import -#import - +#import +#import +#import @implementation MTLDepthStencilDescriptor - +@synthesize depthCompareFunction = _depthCompareFunction; +@synthesize depthWriteEnabled = _depthWriteEnabled; +@synthesize frontFaceStencil = _frontFaceStencil; +@synthesize backFaceStencil = _backFaceStencil; +@synthesize label = _label; +- (instancetype)init { + self = [super init]; + if (self) _depthCompareFunction = MTLCompareFunctionAlways; + return self; +} +- (void)dealloc { [_frontFaceStencil release]; [_backFaceStencil release]; [_label release]; [super dealloc]; } +- (id)copyWithZone:(NSZone*)zone { + MTLDepthStencilDescriptor* copy = [[[self class] allocWithZone:zone] init]; + copy.depthCompareFunction = _depthCompareFunction; + copy.depthWriteEnabled = _depthWriteEnabled; + copy.frontFaceStencil = _frontFaceStencil; + copy.backFaceStencil = _backFaceStencil; + copy.label = _label; + return copy; +} #if DARLING_METAL_ENABLED - -- (NSMethodSignature *)methodSignatureForSelector:(SEL)aSelector -{ - return [NSMethodSignature signatureWithObjCTypes: "v@:"]; +- (Indium::DepthStencilDescriptor)asIndiumDescriptor { + Indium::DepthStencilDescriptor result; + result.depthCompareFunction = static_cast(_depthCompareFunction); + result.depthWriteEnabled = _depthWriteEnabled; + if (_frontFaceStencil) result.frontFaceStencil = [_frontFaceStencil asIndiumDescriptor]; + if (_backFaceStencil) result.backFaceStencil = [_backFaceStencil asIndiumDescriptor]; + return result; } - -- (void)forwardInvocation:(NSInvocation *)anInvocation -{ - NSLog(@"Stub called: %@ in %@", NSStringFromSelector([anInvocation selector]), [self class]); +#endif +@end +@implementation MTLDepthStencilStateInternal +#if DARLING_METAL_ENABLED +@synthesize state = _state; +const std::shared_ptr& MTLDepthStencilGetIndium(id object) { + static const std::shared_ptr none; + return object ? ((MTLDepthStencilStateInternal*) object)->_state : none; } - +@synthesize device = _device; +@synthesize label = _label; +- (instancetype)initWithState:(std::shared_ptr)state device:(id)device label:(NSString*)label { + self = [super init]; + if (self) { _state = state; _device = [device retain]; _label = [label copy]; } + return self; +} +- (void)dealloc { [_device release]; [_label release]; [super dealloc]; } +#else +@dynamic device, label; +MTL_UNSUPPORTED_CLASS #endif - @end diff --git a/src/Metal/MTLDevice.mm b/src/Metal/MTLDevice.mm index f58184a..2c45608 100644 --- a/src/Metal/MTLDevice.mm +++ b/src/Metal/MTLDevice.mm @@ -2,14 +2,32 @@ // SPDX-License-Identifier: MPL-2.0 #import +#import +#import #import #import #import #import #import +#import +#import +#import #import #import +#import +#import +#include +#include +#include +#include +#import +#include #import +#import +#if DARLING_METAL_ENABLED +#include +#include +#endif MTL_EXTERN const MTLDeviceNotificationName MTLDeviceWasAddedNotification = @"MTLDeviceWasAdded"; MTL_EXTERN const MTLDeviceNotificationName MTLDeviceRemovalRequestedNotification = @"MTLDeviceRemovalRequested"; @@ -88,6 +106,37 @@ void MTLRemoveDeviceObserver(id observer) { [observer release]; }; + +#if DARLING_METAL_ENABLED +// Diagnostic: report pipeline creation failures (the game just skips the work when a pipeline is nil). +static void MTLDebugPipelineFailure(const char* what) +{ + if (!(MTLDebugEnabled() || MTLTimingEnabled())) { + return; + } + std::string message = "unknown"; + try { throw; } catch (const std::exception& e) { message = e.what(); } catch (...) {} + MTLTraceCount(what); + static unsigned long logged = 0; + if (++logged <= 80) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "pipeline-failure %s: %s\n", what, message.c_str()); + fclose(log); + } + } +} +#endif + +#if DARLING_METAL_ENABLED +std::mutex& MTLDebugReflectionLock() { static std::mutex lock; return lock; } +std::unordered_map& MTLDebugReflectionMap() { static std::unordered_map map; return map; } +std::string MTLDebugReflectionFor(void* pso) +{ + std::lock_guard guard(MTLDebugReflectionLock()); + auto it = MTLDebugReflectionMap().find(pso); + return it == MTLDebugReflectionMap().end() ? std::string("(none)") : it->second; +} +#endif @implementation MTLDeviceInternal #if DARLING_METAL_ENABLED @@ -96,6 +145,9 @@ void MTLRemoveDeviceObserver(id observer) { NSThread* _pollingThread; NSCondition* _threadExitCondition; BOOL _threadIsRunning; + NSString* _name; + uint64_t _workingSetSize; + BOOL _sharesSystemMemory; } @synthesize device = _device; @@ -116,6 +168,50 @@ void MTLRemoveDeviceObserver(id observer) { [_threadExitCondition unlock]; } +// Real host GPU facts, computed once per device in -initWithDevice: (games may query them per frame). +- (void)loadHostDeviceInfo +{ + // Overrides are read once per process. + static const char* const nameOverride = getenv("DARLING_METAL_GPU_NAME"); + static const uint64_t vramOverride = MTLDeviceInfo::parseMegabytes(getenv("DARLING_METAL_VRAM_MB")); + static const bool macName = [] { + const char* value = getenv("DARLING_METAL_MAC_GPU_NAME"); + return value && value[0] == '1' && value[1] == '\0'; + }(); + + auto priv = dynamic_cast(_device.get()); + std::string vulkanName; + try { + vulkanName = _device ? _device->name() : ""; + } catch (...) { + vulkanName = ""; + } + const char* mac = macName && priv ? MTLDeviceInfo::macGpuName(priv->properties().vendorID) : nullptr; + _name = [[NSString alloc] initWithUTF8String: MTLDeviceInfo::resolveName(nameOverride, vulkanName.c_str(), mac).c_str()]; + if (_name == nil) { + _name = [[NSString alloc] initWithUTF8String: MTLDeviceInfo::kFallbackName]; + } + + uint64_t workingSet = 0; + _sharesSystemMemory = NO; + if (priv) { + const VkPhysicalDeviceType type = priv->properties().deviceType; + _sharesSystemMemory = type == VK_PHYSICAL_DEVICE_TYPE_INTEGRATED_GPU || type == VK_PHYSICAL_DEVICE_TYPE_CPU; + const VkPhysicalDeviceMemoryProperties memory = priv->memoryProperties(); + MTLDeviceInfo::Heap heaps[VK_MAX_MEMORY_HEAPS]; + const size_t count = memory.memoryHeapCount < VK_MAX_MEMORY_HEAPS ? memory.memoryHeapCount : VK_MAX_MEMORY_HEAPS; + for (size_t i = 0; i < count; ++i) { + heaps[i] = { memory.memoryHeaps[i].size, (memory.memoryHeaps[i].flags & VK_MEMORY_HEAP_DEVICE_LOCAL_BIT) != 0 }; + } + workingSet = MTLDeviceInfo::workingSetBytes(heaps, count, _sharesSystemMemory); + } + if (vramOverride != 0) { + workingSet = vramOverride; + } + // 1 GiB was the previous fixed value; keep it when Vulkan reports nothing usable. + _workingSetSize = workingSet != 0 ? workingSet : (1ull << 30); +} + - (instancetype)initWithDevice: (std::shared_ptr)device { self = [super init]; @@ -123,6 +219,7 @@ void MTLRemoveDeviceObserver(id observer) { _device = device; _threadExitCondition = [NSCondition new]; _threadIsRunning = NO; + [self loadHostDeviceInfo]; _pollingThread = [[NSThread alloc] initWithTarget: self selector: @selector(pollingLoop) object: nil]; [_pollingThread start]; } @@ -133,6 +230,7 @@ void MTLRemoveDeviceObserver(id observer) { { [_pollingThread release]; [_threadExitCondition release]; + [_name release]; [super dealloc]; } @@ -153,12 +251,60 @@ void MTLRemoveDeviceObserver(id observer) { [_threadExitCondition unlock]; } +- (MTLReadWriteTextureTier)readWriteTextureSupport +{ + /* darling-lol: ANGLE queries this during GPU init; it was missing + * (unrecognized selector -> terminate). Report Tier2; TierNone made + * ANGLE refuse context creation outright (eglCreateContext failure -> + * GPU exit(1) -> browser FATAL). If Indium lacks something ANGLE then + * uses, that will surface as the next concrete gap instead. */ + return MTLReadWriteTextureTier2; +} + +/* darling-lol M57: feature queries ANGLE's Metal backend needs. */ +- (BOOL)isDepth24Stencil8PixelFormatSupported { return _device->supportsDepth24Stencil8PixelFormat(); } +- (BOOL)hasUnifiedMemory +{ + // Stays YES by default: with NO, ANGLE and games take the Managed/Private paths + // (synchronizeResource:, Private uploads) that this Metal/Indium does not fully implement. + // DARLING_METAL_HOST_UNIFIED_MEMORY=1 reports the host GPU type instead + // (YES for integrated and CPU devices, NO for discrete ones). + static const bool reportHost = [] { + const char* value = getenv("DARLING_METAL_HOST_UNIFIED_MEMORY"); + return value && value[0] == '1' && value[1] == '\0'; + }(); + return reportHost ? _sharesSystemMemory : YES; +} +- (NSUInteger)maxBufferLength { return 256u * 1024u * 1024u; } +- (NSUInteger)recommendedMaxWorkingSetSize +{ + return _workingSetSize > NSUIntegerMax ? NSUIntegerMax : static_cast(_workingSetSize); +} +- (BOOL)isLowPower { return NO; } +- (BOOL)isRemovable { return NO; } +- (BOOL)isHeadless { return NO; } +- (NSString*)name { return _name; } +- (uint64_t)registryID { return MTLDeviceInfo::kDefaultRegistryID; } +- (BOOL)supportsFamily: (NSUInteger)family { return YES; } +- (BOOL)supportsFeatureSet: (NSUInteger)featureSet { return YES; } +- (BOOL)supportsTextureSampleCount: (NSUInteger)sampleCount { return sampleCount <= 4; } +- (id)newCommandQueueWithMaxCommandBufferCount: (NSUInteger)maxCommandBufferCount +{ + return [self newCommandQueue]; +} + - (id)newComputePipelineStateWithDescriptor: (MTLComputePipelineDescriptor*)descriptor options: (MTLPipelineOption)options reflection: (MTLAutoreleasedComputePipelineReflection*)reflection error: (NSError**)error { - auto pso = _device->newComputePipelineState(descriptor.asIndiumDescriptor, static_cast(options), nullptr); + std::decay_tnewComputePipelineState(descriptor.asIndiumDescriptor, static_cast(options), nullptr))> pso; + try { pso = _device->newComputePipelineState(descriptor.asIndiumDescriptor, static_cast(options), nullptr); } + catch (...) { + MTLDebugPipelineFailure("computePipelineFailure"); + if (error) *error = MTLErrorFromCurrentException("newComputePipelineStateWithDescriptor"); + return nil; + } if (!pso) { if (error) { // TODO: better error and/or match what the official Metal method does @@ -166,6 +312,7 @@ void MTLRemoveDeviceObserver(id observer) { } return nil; } + if (reflection) *reflection = [[[MTLComputePipelineReflection alloc] initWithFunction:((MTLFunctionInternal*)descriptor.computeFunction).function] autorelease]; return [[MTLComputePipelineStateInternal alloc] initWithState: pso device: self label: descriptor.label]; } @@ -183,7 +330,13 @@ void MTLRemoveDeviceObserver(id observer) { reflection: (MTLAutoreleasedComputePipelineReflection*)reflection error: (NSError**)error { - auto pso = _device->newComputePipelineState(((MTLFunctionInternal*)computeFunction).function, static_cast(options), nullptr); + std::decay_tnewComputePipelineState(((MTLFunctionInternal*)computeFunction).function, static_cast(options), nullptr))> pso; + try { pso = _device->newComputePipelineState(((MTLFunctionInternal*)computeFunction).function, static_cast(options), nullptr); } + catch (...) { + MTLDebugPipelineFailure("computePipelineFailure"); + if (error) *error = MTLErrorFromCurrentException("newComputePipelineStateWithFunction"); + return nil; + } if (!pso) { if (error) { // TODO: better error and/or match what the official Metal method does @@ -191,13 +344,44 @@ void MTLRemoveDeviceObserver(id observer) { } return nil; } + if (reflection) *reflection = [[[MTLComputePipelineReflection alloc] initWithFunction:((MTLFunctionInternal*)computeFunction).function] autorelease]; return [[MTLComputePipelineStateInternal alloc] initWithState: pso device: self label: nil]; } +- (id)newRenderPipelineStateWithDescriptor:(MTLRenderPipelineDescriptor*)descriptor options:(MTLPipelineOption)options reflection:(MTLAutoreleasedRenderPipelineReflection*)reflection error:(NSError**)error { + if (reflection) *reflection = nil; + if (options & ~(MTLPipelineOptionArgumentInfo | MTLPipelineOptionBufferTypeInfo)) { + if (error) *error = [NSError errorWithDomain:NSPOSIXErrorDomain code:ENOTSUP userInfo:nil]; + return nil; + } + id state = [self newRenderPipelineStateWithDescriptor:descriptor error:error]; + if (state && reflection) *reflection = [[[MTLRenderPipelineReflection alloc] initWithVertexFunction:((MTLFunctionInternal*)descriptor.vertexFunction).function fragmentFunction:((MTLFunctionInternal*)descriptor.fragmentFunction).function vertexDescriptor:descriptor.vertexDescriptor] autorelease]; + if (state && reflection && *reflection && MTLDrawTracking()) { + // Diagnostic: remember what the game is told about this pipeline's resources (it binds by these names/indices) + std::string text; + char piece[200]; + for (MTLArgument* argument in [*reflection fragmentArguments]) { + snprintf(piece, sizeof(piece), " %lu:%lu:%s", (unsigned long) argument.type, (unsigned long) argument.index, [argument.name UTF8String]); + text += piece; + } + std::lock_guard guard(MTLDebugReflectionLock()); + MTLDebugReflectionMap()[(void*) state] = text; + } + return state; +} + - (id)newRenderPipelineStateWithDescriptor: (MTLRenderPipelineDescriptor*)descriptor error: (NSError**)error { - auto pso = _device->newRenderPipelineState([descriptor asIndiumDescriptor]); + MTLTraceCount("pipelineRequest"); + std::decay_tnewRenderPipelineState([descriptor asIndiumDescriptor]))> pso; + try { pso = _device->newRenderPipelineState([descriptor asIndiumDescriptor]); } + catch (...) { + MTLDebugPipelineFailure("renderPipelineFailure"); + MTLTraceCount("pipelineException"); + if (error) *error = MTLErrorFromCurrentException("newRenderPipelineStateWithDescriptor"); + return nil; + } if (!pso) { if (error) { // TODO: better error and/or match what the official Metal method does @@ -205,7 +389,118 @@ void MTLRemoveDeviceObserver(id observer) { } return nil; } - return [[MTLRenderPipelineStateInternal alloc] initWithState: pso device: self label: descriptor.label]; + id created = [[MTLRenderPipelineStateInternal alloc] initWithState: pso device: self label: descriptor.label]; + if (MTLDrawTracking()) { + // Diagnostic: the reflection the game would get for this pipeline (it binds resources by these names/indices) + MTLRenderPipelineReflection* debugReflection = [[[MTLRenderPipelineReflection alloc] initWithVertexFunction:((MTLFunctionInternal*)descriptor.vertexFunction).function fragmentFunction:((MTLFunctionInternal*)descriptor.fragmentFunction).function vertexDescriptor:descriptor.vertexDescriptor] autorelease]; + std::string text = "vs=" + std::string([descriptor.vertexFunction.name UTF8String] ?: "(nil)") + + " fs=" + std::string([descriptor.fragmentFunction.name UTF8String] ?: "(nil)"); + char piece[200]; + for (MTLArgument* argument in [debugReflection fragmentArguments]) { + snprintf(piece, sizeof(piece), " %lu:%lu:%s", (unsigned long) argument.type, (unsigned long) argument.index, [argument.name UTF8String]); + text += piece; + } + std::lock_guard guard(MTLDebugReflectionLock()); + MTLDebugReflectionMap()[(void*) created] = text; + } + static unsigned long logged = 0; + if (MTLDebugEnabled() && ++logged <= 300) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + MTLRenderPipelineColorAttachmentDescriptor* color = [descriptor.colorAttachments objectAtIndexedSubscript: 0]; + fprintf(log, "pso %p indium=%p vs=%s fs=%s color0 fmt=%lu blend=%d src=%lu/%lu dst=%lu/%lu op=%lu/%lu mask=%lu depthFmt=%lu samples=%lu rasterDiscard=%d\n", created, (void*) pso.get(), + [descriptor.vertexFunction.name UTF8String], [descriptor.fragmentFunction.name UTF8String], + (unsigned long) color.pixelFormat, (int) color.blendingEnabled, + (unsigned long) color.sourceRGBBlendFactor, (unsigned long) color.sourceAlphaBlendFactor, + (unsigned long) color.destinationRGBBlendFactor, (unsigned long) color.destinationAlphaBlendFactor, + (unsigned long) color.rgbBlendOperation, (unsigned long) color.alphaBlendOperation, (unsigned long) color.writeMask, + (unsigned long) descriptor.depthAttachmentPixelFormat, (unsigned long) descriptor.sampleCount, (int) descriptor.rasterizationEnabled); + MTLVertexDescriptor* vertex = descriptor.vertexDescriptor; + for (NSUInteger i = 0; vertex && i < 31; ++i) { + MTLVertexAttributeDescriptor* attr = vertex.attributes[i]; + if (attr.format == MTLVertexFormatInvalid) continue; + MTLVertexBufferLayoutDescriptor* layout = vertex.layouts[attr.bufferIndex]; + fprintf(log, " attr[%lu] fmt=%lu off=%lu buf=%lu stride=%lu step=%lu rate=%lu\n", (unsigned long) i, (unsigned long) attr.format, + (unsigned long) attr.offset, (unsigned long) attr.bufferIndex, (unsigned long) layout.stride, (unsigned long) layout.stepFunction, (unsigned long) layout.stepRate); + } + if (!vertex) fprintf(log, " no vertex descriptor\n"); + fclose(log); + } + } + return created; +} + +- (id)newDepthStencilStateWithDescriptor:(MTLDepthStencilDescriptor*)descriptor +{ + if (!descriptor) return nil; + std::shared_ptr state; + try { state = _device->newDepthStencilState([descriptor asIndiumDescriptor]); } + catch (...) { return nil; } + if (!state) return nil; + return [[MTLDepthStencilStateInternal alloc] initWithState:state device:self label:descriptor.label]; +} + +- (id)newSamplerStateWithDescriptor:(MTLSamplerDescriptor*)descriptor +{ + if (!descriptor) return nil; + Indium::SamplerDescriptor desc; + desc.minFilter = static_cast(descriptor.minFilter); + desc.magFilter = static_cast(descriptor.magFilter); + desc.mipFilter = static_cast(descriptor.mipFilter); + desc.maxAnisotropy = descriptor.maxAnisotropy; + desc.sAddressMode = static_cast(descriptor.sAddressMode); + desc.tAddressMode = static_cast(descriptor.tAddressMode); + desc.rAddressMode = static_cast(descriptor.rAddressMode); + desc.borderColor = static_cast(descriptor.borderColor); + desc.normalizedCoordinates = descriptor.normalizedCoordinates; + desc.lodMinClamp = descriptor.lodMinClamp; + desc.lodMaxClamp = descriptor.lodMaxClamp; + desc.supportArgumentBuffers = descriptor.supportArgumentBuffers; + desc.compareFunction = static_cast(descriptor.compareFunction); + std::shared_ptr sampler; + try { sampler = _device->newSamplerState(desc); } + catch (...) { return nil; } + if (!sampler) return nil; + return [[MTLSamplerStateInternal alloc] initWithSampler:sampler device:self label:descriptor.label]; +} + +- (id)newTextureWithDescriptor:(MTLTextureDescriptor*)descriptor +{ + MTLTraceCount("newTexture"); + if (!descriptor || !descriptor.width || !descriptor.height || !descriptor.depth || + !descriptor.mipmapLevelCount || !descriptor.sampleCount || !descriptor.arrayLength) return nil; + Indium::TextureDescriptor desc; + desc.textureType = static_cast(descriptor.textureType); + desc.pixelFormat = static_cast(descriptor.pixelFormat); + desc.width = descriptor.width; + desc.height = descriptor.height; + desc.depth = descriptor.depth; + desc.mipmapLevelCount = descriptor.mipmapLevelCount; + desc.sampleCount = descriptor.sampleCount; + desc.arrayLength = descriptor.arrayLength; + desc.resourceOptions = static_cast(descriptor.resourceOptions); + desc.allowGPUOptimizedContents = descriptor.allowGPUOptimizedContents; + desc.usage = static_cast(descriptor.usage); + desc.swizzle = MTLTextureSwizzleChannelsToIndium(descriptor.swizzle); + std::shared_ptr texture; + std::string failure; + try { texture = _device->newTexture(desc); } + catch (const std::exception& e) { failure = e.what(); } + catch (...) { failure = "unknown exception"; } + if (MTLDebugEnabled() || MTLTimingEnabled()) { + static unsigned long renderTargets = 0, failures = 0; + bool renderTarget = (descriptor.usage & MTLTextureUsageRenderTarget) != 0; + if (!texture ? ++failures <= 60 : (renderTarget && ++renderTargets <= 80)) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "%s %lux%lux%lu type=%lu fmt=%lu mips=%lu len=%lu samples=%lu usage=%lu storage=%lu %s\n", texture ? "renderTargetTexture" : "TEXTURE-FAILED", + (unsigned long) descriptor.width, (unsigned long) descriptor.height, (unsigned long) descriptor.depth, (unsigned long) descriptor.textureType, + (unsigned long) descriptor.pixelFormat, (unsigned long) descriptor.mipmapLevelCount, (unsigned long) descriptor.arrayLength, (unsigned long) descriptor.sampleCount, + (unsigned long) descriptor.usage, (unsigned long) descriptor.storageMode, failure.c_str()); + fclose(log); + } + } + } + if (!texture) return nil; + return [[MTLTextureInternal alloc] initWithTexture:texture device:self resourceOptions:descriptor.resourceOptions]; } - (id)newCommandQueue @@ -275,8 +570,21 @@ void MTLRemoveDeviceObserver(id observer) { - (id)newLibraryWithData: (dispatch_data_t)data error: (NSError**)error { - NSData* nsdata = (NSData*)data; - auto lib = _device->newLibrary(nsdata.bytes, nsdata.length); + const void* bytes = nullptr; + size_t length = 0; + dispatch_data_t mapped = dispatch_data_create_map(data, &bytes, &length); + if (!mapped) { + if (error) *error = [NSError errorWithDomain:NSPOSIXErrorDomain code:ENOMEM userInfo:nil]; + return nil; + } + std::shared_ptr lib; + try { lib = _device->newLibrary(bytes, length); } + catch (...) { + dispatch_release(mapped); + if (error) *error = MTLErrorFromCurrentException("newLibraryWithData"); + return nil; + } + dispatch_release(mapped); if (!lib) { return nil; } diff --git a/src/Metal/MTLLibrary.mm b/src/Metal/MTLLibrary.mm index 0feb46e..e0c0e93 100644 --- a/src/Metal/MTLLibrary.mm +++ b/src/Metal/MTLLibrary.mm @@ -1,6 +1,7 @@ // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 +#import #import #import #import @@ -27,10 +28,12 @@ self = [super init]; if (self != nil) { _function = function; + _functionType = static_cast(function->functionType()); + _device = [device retain]; // cache the name - _name = [NSString stringWithUTF8String: _function->name().c_str()]; + _name = [[NSString alloc] initWithUTF8String: _function->name().c_str()]; } return self; } @@ -121,7 +124,12 @@ MTL_UNSUPPORTED_CLASS - (id)newFunctionWithName: (NSString*)functionName { - auto func = _library->newFunction(functionName.UTF8String); + std::shared_ptr func; + try { func = _library->newFunction(functionName.UTF8String); } + catch (...) { + MTLErrorFromCurrentException("newFunctionWithName"); + return nil; + } if (!func) { return nil; } diff --git a/src/Metal/MTLRenderCommandEncoder.mm b/src/Metal/MTLRenderCommandEncoder.mm index e931b90..b2fe0a5 100644 --- a/src/Metal/MTLRenderCommandEncoder.mm +++ b/src/Metal/MTLRenderCommandEncoder.mm @@ -1,3 +1,9 @@ +#import +#import +#import +#import +#import +#import // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 @@ -9,8 +15,408 @@ #import #import +#include +#include +#include #include +#include +#if DARLING_METAL_ENABLED +extern std::string MTLDebugReflectionFor(void* pso); +#endif + +#if DARLING_METAL_ENABLED +// Diagnostic (MTLDebugEnabled only): the state a draw is issued with, logged for the first few hundred draws. +struct MTLDebugDrawState { + MTLViewport viewport = {}; + MTLScissorRect scissor = {}; + long cull = -1, winding = -1, fill = -1; + void* pipeline = nullptr; + void* depthStencil = nullptr; + id vertexBuffers[31] = {}; NSUInteger vertexOffsets[31] = {}; + id fragmentBuffers[31] = {}; NSUInteger fragmentOffsets[31] = {}; + unsigned char vertexBytes[31][64] = {}; NSUInteger vertexBytesLength[31] = {}; + unsigned char fragmentBytes[31][64] = {}; NSUInteger fragmentBytesLength[31] = {}; + id fragmentTextures[31] = {}; + id vertexTextures[31] = {}; +}; +static thread_local MTLDebugDrawState debugDrawState; + +// Experiments, each enabled by a file in the guest /tmp present when the process starts (with metal-debug-trace) +#define MTL_DEBUG_FLAG(name) ([]{ static const bool on = MTLDebugEnabled() && access("/tmp/" name, F_OK) == 0; return on; }()) + + +static void MTLDebugDumpFloats(FILE* log, const char* what, int index, const void* bytes, size_t available) +{ + fprintf(log, " %s[%d]:", what, index); + const float* floats = (const float*) bytes; + for (size_t i = 0; i < 16 && (i + 1) * 4 <= available; ++i) { + fprintf(log, " %g", floats[i]); + } + fprintf(log, "\n"); +} + +static void MTLDebugDumpBindings(unsigned long drawn) +{ + const MTLDebugDrawState& state = debugDrawState; + FILE* log = fopen("/tmp/metal-trace.log", "a"); + if (!log) { + return; + } + fprintf(log, "bindings at draw #%lu\n", drawn); + for (int i = 0; i < 31; ++i) { + if (state.vertexBuffers[i]) { + id buffer = state.vertexBuffers[i]; + const char* contents = (const char*) buffer.contents; + fprintf(log, " vbuf[%d] len=%lu off=%lu contents=%s\n", i, (unsigned long) buffer.length, (unsigned long) state.vertexOffsets[i], contents ? "cpu" : "none"); + if (contents && state.vertexOffsets[i] < buffer.length) { + MTLDebugDumpFloats(log, "vdata", i, contents + state.vertexOffsets[i], buffer.length - state.vertexOffsets[i]); + } + } + if (state.vertexBytesLength[i]) { + MTLDebugDumpFloats(log, "vbytes", i, state.vertexBytes[i], std::min(64, state.vertexBytesLength[i])); + } + if (state.fragmentBuffers[i]) { + id buffer = state.fragmentBuffers[i]; + const char* contents = (const char*) buffer.contents; + fprintf(log, " fbuf[%d] len=%lu off=%lu contents=%s\n", i, (unsigned long) buffer.length, (unsigned long) state.fragmentOffsets[i], contents ? "cpu" : "none"); + if (contents && state.fragmentOffsets[i] < buffer.length) { + MTLDebugDumpFloats(log, "fdata", i, contents + state.fragmentOffsets[i], buffer.length - state.fragmentOffsets[i]); + } + } + if (state.fragmentBytesLength[i]) { + MTLDebugDumpFloats(log, "fbytes", i, state.fragmentBytes[i], std::min(64, state.fragmentBytesLength[i])); + } + if (state.fragmentTextures[i]) { + id texture = state.fragmentTextures[i]; + unsigned long width = texture.width, height = texture.height, format = texture.pixelFormat; + unsigned long lit = 0, alpha = 0, total = 0; + bool read = MTLDebugEnabled() && (format == 80 || format == 81 || format == 70 || format == 71) && width * height <= 4096 * 4096; + if (read) { + id queue = [[texture device] newCommandQueue]; + id readback = [[texture device] newBufferWithLength: width * height * 4 options: MTLResourceStorageModeShared]; + id commandBuffer = [queue commandBuffer]; + id blit = [commandBuffer blitCommandEncoder]; + [blit copyFromTexture: texture sourceSlice: 0 sourceLevel: 0 sourceOrigin: MTLOriginMake(0, 0, 0) + sourceSize: MTLSizeMake(width, height, 1) toBuffer: readback destinationOffset: 0 + destinationBytesPerRow: width * 4 destinationBytesPerImage: width * height * 4]; + [blit endEncoding]; + [commandBuffer commit]; + [commandBuffer waitUntilCompleted]; + const uint8_t* bytes = (const uint8_t*) readback.contents; + for (unsigned long k = 0; k < width * height * 4; k += 4) { + ++total; + if (bytes[k] | bytes[k + 1] | bytes[k + 2]) ++lit; + if (bytes[k + 3]) ++alpha; + } + [readback release]; + [queue release]; + } + fprintf(log, " ftex[%d] ptr=%p %lux%lu type=%lu mips=%lu len=%lu samples=%lu fmt=%lu%s litRGB=%lu alphaNonZero=%lu of %lu\n", i, (void*) texture, width, height, (unsigned long) texture.textureType, (unsigned long) texture.mipmapLevelCount, (unsigned long) texture.arrayLength, (unsigned long) texture.sampleCount, format, read ? "" : " (not read)", lit, alpha, total); + } + } + fclose(log); +} + +static void MTLDebugLogDraw(const char* kind, unsigned long count, unsigned long instances, unsigned long indexType, long baseVertex, id indexBuffer, NSUInteger indexOffset) +{ + if (!MTLDebugEnabled()) { + // Light mode: only while /tmp/metal-snapshot exists, log up to 300 draws with their bindings + static unsigned long lightLogged = 0; + static unsigned long checks = 0; + static bool armed = false; + static std::string captureTag; + // re-read a few times a second: this runs for every draw, and a guest fopen is several host syscalls + if (checks++ % 4096 == 0) { + armed = false; + if (FILE* file = fopen("/tmp/metal-snapshot", "r")) { + char tag[48] = {}; + if (fgets(tag, sizeof(tag), file)) { + if (captureTag != tag) { captureTag = tag; lightLogged = 0; } + armed = true; + } + fclose(file); + } + } + if (!armed || lightLogged >= 300) { + return; + } + ++lightLogged; + MTLDebugDumpBindings(2000 + lightLogged); + { + // Save the raw blocks of a few block-compressed textures (decoded offline) to check the stored image + static std::vector dumped; + for (int i = 0; i < 31; ++i) { + id texture = debugDrawState.fragmentTextures[i]; + unsigned long format = texture ? (unsigned long) texture.pixelFormat : 0; + bool bc = format == 130 || format == 131 || format == 134 || format == 135; + if (!bc || texture.width < 128 || dumped.size() >= 8 || std::find(dumped.begin(), dumped.end(), (void*) texture) != dumped.end()) continue; + dumped.push_back((void*) texture); + NSUInteger level = std::min(1, texture.mipmapLevelCount - 1); + NSUInteger width = std::max(1, texture.width >> level), height = std::max(1, texture.height >> level); + NSUInteger blocksX = (width + 3) / 4, blocksY = (height + 3) / 4, blockBytes = (format == 130 || format == 131) ? 8 : 16; + NSUInteger rowBytes = blocksX * blockBytes, total = rowBytes * blocksY; + id queue = [[texture device] newCommandQueue]; + id readback = [[texture device] newBufferWithLength: total options: MTLResourceStorageModeShared]; + id commandBuffer = [queue commandBuffer]; + id blit = [commandBuffer blitCommandEncoder]; + [blit copyFromTexture: texture sourceSlice: 0 sourceLevel: level sourceOrigin: MTLOriginMake(0, 0, 0) sourceSize: MTLSizeMake(width, height, 1) + toBuffer: readback destinationOffset: 0 destinationBytesPerRow: rowBytes destinationBytesPerImage: total]; + [blit endEncoding]; + [commandBuffer commit]; + [commandBuffer waitUntilCompleted]; + char path[128]; + snprintf(path, sizeof(path), "/tmp/tex-%02zu-%lux%lu-f%lu-level%lu.bc", dumped.size(), (unsigned long) width, (unsigned long) height, format, (unsigned long) level); + if (FILE* out = fopen(path, "wb")) { fwrite(readback.contents, 1, total, out); fclose(out); } + [readback release]; + [queue release]; + } + } + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "draw #%lu %s count=%lu inst=%lu baseVtx=%ld pso=%p ds=%p cull=%ld wind=%ld\n", 2000 + lightLogged, kind, count, instances, baseVertex, debugDrawState.pipeline, debugDrawState.depthStencil, debugDrawState.cull, debugDrawState.winding); + fprintf(log, "reflect pso=%p frag:%s\n", debugDrawState.pipeline, MTLDebugReflectionFor(debugDrawState.pipeline).c_str()); + fclose(log); + } + return; + } + static unsigned long drawn = 0; + static unsigned long bigDrawn = 0; + bool big = count >= 600 || instances > 1; + if (big && ++bigDrawn <= 120) { + MTLDebugDumpBindings(1000 + bigDrawn); + } else if (++drawn > 400) { + return; + } + const MTLDebugDrawState& state = debugDrawState; + if (drawn == 1 || drawn == 7 || drawn == 13 || drawn == 40) { + MTLDebugDumpBindings(drawn); + if (indexBuffer && indexBuffer.contents) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + const uint16_t* indices = (const uint16_t*)((const char*) indexBuffer.contents + indexOffset); + { + const unsigned char* base = (const unsigned char*) indexBuffer.contents; + NSUInteger from = indexOffset >= 256 ? indexOffset - 256 : 0; + fprintf(log, " before-index bytes from %lu:", (unsigned long) from); + for (NSUInteger k = from; k < indexOffset + 16; k += 4) { + float f; memcpy(&f, base + k, 4); + unsigned u; memcpy(&u, base + k, 4); + fprintf(log, " %g/%08x", f, u); + } + fprintf(log, "\n"); + } + fprintf(log, " indices(type %lu, off %lu, len %lu):", indexType, (unsigned long) indexOffset, (unsigned long) indexBuffer.length); + for (int k = 0; k < 12; ++k) fprintf(log, " %u", indices[k]); + fprintf(log, "\n"); + fclose(log); + } + } + } + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "draw #%lu %s count=%lu inst=%lu idxType=%lu baseVtx=%ld pso=%p ds=%p vp=(%.0f %.0f %.0f %.0f %.2f %.2f) sc=(%lu %lu %lu %lu) cull=%ld wind=%ld fill=%ld\n", + drawn, kind, count, instances, indexType, baseVertex, state.pipeline, state.depthStencil, + state.viewport.originX, state.viewport.originY, state.viewport.width, state.viewport.height, state.viewport.znear, state.viewport.zfar, + (unsigned long) state.scissor.x, (unsigned long) state.scissor.y, (unsigned long) state.scissor.width, (unsigned long) state.scissor.height, + state.cull, state.winding, state.fill); + fclose(log); + } +} +#endif + + +#if DARLING_METAL_ENABLED +// Experiment: /tmp/metal-draw-limit holds N; every render pass then executes only its first N draws. +static thread_local unsigned long passDrawIndex = 0; +static bool MTLDrawLimited(unsigned long count) +{ + static std::atomic calls { 0 }; + static std::atomic limit { -1 }; + static std::atomic minimum { 0 }; + static std::atomic maximum { -1 }; + static const bool possible = access("/tmp/metal-debug-draws", F_OK) == 0 || MTLDebugEnabled(); + if (!possible) { + return false; + } + if (calls.fetch_add(1) % 200 == 0) { + long value = -1; + if (FILE* file = fopen("/tmp/metal-draw-limit", "r")) { + if (fscanf(file, "%ld", &value) != 1) value = -1; + fclose(file); + } + limit = value; + long low = 0, high = -1; + if (FILE* file = fopen("/tmp/metal-draw-min", "r")) { if (fscanf(file, "%ld", &low) != 1) low = 0; fclose(file); } + if (FILE* file = fopen("/tmp/metal-draw-max", "r")) { if (fscanf(file, "%ld", &high) != 1) high = -1; fclose(file); } + minimum = low; + maximum = high; + } + // skip draws outside [min, max] index counts (experiment: which draws build which part of the picture) + if (count < (unsigned long) (long) minimum || (maximum >= 0 && count > (unsigned long) (long) maximum)) { + return true; + } + long n = limit; + return n >= 0 && ++passDrawIndex > (unsigned long) n; +} +#endif + +#if DARLING_METAL_ENABLED +// Metal's render encoders start with every slot unbound, and a bind only replaces the value its slot +// holds, so re-binding the identical value at the same slot cannot change what later draws see: the +// call can be dropped before any of its work (the resource's shared_ptr copy and its atomic refcount, +// the Indium op, and the first touch of both objects). These records are what the setters compare +// against. +// +// A record holds the *Indium* object's address (`shared_ptr::get()`, see MTLBindKey), never the +// Objective-C wrapper's address. The wrapper is not a stable identity: the game may release it while +// the Indium encoder still keeps the resource it wrapped alive, the allocator can then hand the same +// address to a wrapper around a different resource (transient per-draw buffers do exactly that), and a +// comparison on the wrapper's address would drop that new resource's bind -- every later draw would +// keep using the old one (the ABA problem). The Indium encoder holds each resource it was given until +// the encoder ends -- `FunctionResources::textures/buffers/samplers` store the shared_ptr the setter +// received (command-encoder.private.hpp), and the pipeline is held by `_privatePSO`/`_pipelineOwners` +// (render-command-encoder.cpp), on both the deferred and the direct encoder path -- so an Indium +// address cannot be reused while it is recorded. The depth-stencil state is the one kind Indium does +// not keep (its setter reads the descriptor and issues the Vulkan commands), so its record retains +// one itself (`_depthStencilOwner`). +// +// `valid` marks the slots a value was recorded for: a slot that was never bound is always forwarded, +// and setVertexBytes/setFragmentBytes (inline data in the same slot) just clears the mark. +struct MTLBindObjectSlots { + const void* keys[31] = {}; + uint32_t valid = 0; +}; + +struct MTLBindBufferSlots { + const void* keys[31] = {}; + NSUInteger offsets[31] = {}; + uint32_t valid = 0; +}; + +// The key a bind is recorded and compared under: the Indium object's address. Reading it out of the +// accessor's const shared_ptr reference is one load of the wrapper's payload -- no copy, no atomic. +template +static inline const void* MTLBindKey(const std::shared_ptr& object) +{ + return object.get(); +} + +// true: the slot already holds this value, so the caller can drop the call; false: forward it, and the +// value is recorded so the next identical call can be dropped. +static bool MTLBindObjectSlot(MTLBindObjectSlots& slots, NSUInteger index, const void* key) +{ + if (index >= 31) { + return false; + } + const uint32_t bit = 1u << index; + if ((slots.valid & bit) && slots.keys[index] == key) { + return true; + } + slots.keys[index] = key; + slots.valid |= bit; + return false; +} + +static bool MTLBindBufferSlot(MTLBindBufferSlots& slots, NSUInteger index, const void* key, NSUInteger offset) +{ + if (index >= 31) { + return false; + } + const uint32_t bit = 1u << index; + if ((slots.valid & bit) && slots.keys[index] == key && slots.offsets[index] == offset) { + return true; + } + slots.keys[index] = key; + slots.offsets[index] = offset; + slots.valid |= bit; + return false; +} + +// The offset belongs to the buffer's slot state, so an unchanged one is dropped the same way. +static bool MTLBindBufferOffset(MTLBindBufferSlots& slots, NSUInteger index, NSUInteger offset) +{ + if (index >= 31 || !(slots.valid & (1u << index))) { + return false; + } + if (slots.offsets[index] == offset) { + return true; + } + slots.offsets[index] = offset; + return false; +} + +static void MTLBindForgetBufferSlot(MTLBindBufferSlots& slots, NSUInteger index) +{ + if (index < 31) { + slots.valid &= ~(1u << index); + } +} + +// endEncoding: nothing may be bound after it, and the Indium encoder releases the resources it held, +// so no record -- and no address a record kept meaningful -- survives the pass. +static void MTLBindForgetAll(MTLBindObjectSlots& slots) +{ + slots = MTLBindObjectSlots {}; +} + +static void MTLBindForgetAll(MTLBindBufferSlots& slots) +{ + slots = MTLBindBufferSlots {}; +} + +// Diagnostic (guest flag /tmp/metal-debug-binds, or the full trace flag): count the dropped re-binds. +// The setters test MTLBindSkipsEnabled() inline first, so a dropped call pays one load and a branch +// rather than the counter's call frame; MTLBindSkipCount then appends the totals to the trace log on +// the same 3 s interval as the other counters. +static bool MTLBindSkipsEnabled() +{ + static const bool enabled = MTLDebugEnabled() || access("/tmp/metal-debug-binds", F_OK) == 0; + return enabled; +} + +static void MTLBindSkipCount(const char* what) +{ + static const bool enabled = MTLDebugEnabled() || access("/tmp/metal-debug-binds", F_OK) == 0; + if (!enabled) { + return; + } + static std::mutex lock; + static std::map counts; + static time_t last = 0; + std::lock_guard guard(lock); + ++counts[what]; + time_t now = time(nullptr); + if (now - last >= 3) { + last = now; + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "bind-skips@%ld:", (long) now); + for (auto& count: counts) { + fprintf(log, " %s=%lu", count.first.c_str(), count.second); + } + fprintf(log, "\n"); + fclose(log); + } + } +} +#endif +#if DARLING_METAL_ENABLED +// Live experiments (guest flag files, re-read every few hundred state changes): /tmp/metal-flag-nocull, -flipwind, -nodepth +static bool MTLLiveFlag(int which) +{ + static std::atomic calls { 0 }; + static std::atomic state { 0 }; + static const bool possible = access("/tmp/metal-debug-draws", F_OK) == 0 || MTLDebugEnabled(); + if (!possible) { + return false; + } + if (calls.fetch_add(1) % 100 == 0) { + unsigned value = 0; + if (access("/tmp/metal-flag-nocull", F_OK) == 0) value |= 1; + if (access("/tmp/metal-flag-flipwind", F_OK) == 0) value |= 2; + if (access("/tmp/metal-flag-nodepth", F_OK) == 0) value |= 4; + state = value; + } + return (state & (1u << which)) != 0; +} +#endif @implementation MTLRenderPassAttachmentDescriptor #if DARLING_METAL_ENABLED @@ -96,11 +502,11 @@ MTL_UNSUPPORTED_CLASS - (Indium::RenderPassAttachmentDescriptor)asIndiumDescriptor { return Indium::RenderPassAttachmentDescriptor { - ((MTLTextureInternal*)_texture).texture, + MTLTextureGetIndium(_texture), _level, _slice, _depthPlane, - ((MTLTextureInternal*)_resolveTexture).texture, + MTLTextureGetIndium(_resolveTexture), _resolveLevel, _resolveSlice, _resolveDepthPlane, @@ -367,7 +773,11 @@ MTL_UNSUPPORTED_CLASS std::vector result; for (const auto& [key, value]: tmp2) { - result.push_back(value); + // A slot that was only looked at (or never given a texture) is not an attachment; depth-only passes rely on this + if (value.texture) { + result.resize(key + 1); + result[key] = value; + } } return result; @@ -404,6 +814,10 @@ MTL_UNSUPPORTED_CLASS - (void)_doInit { _colorAttachments = [MTLRenderPassColorAttachmentDescriptorArray new]; + _depthAttachment = [MTLRenderPassDepthAttachmentDescriptor new]; + _stencilAttachment = [MTLRenderPassStencilAttachmentDescriptor new]; + _renderTargetArrayLength = 1; + _defaultRasterSampleCount = 1; } - (void)_doDealloc @@ -502,9 +916,9 @@ MTL_UNSUPPORTED_CLASS { return Indium::RenderPassDescriptor { [_colorAttachments asContiguousIndiumDescriptors], - (_depthAttachment == nil) ? std::nullopt : std::make_optional([_depthAttachment asIndiumDescriptor]), - (_stencilAttachment == nil) ? std::nullopt : std::make_optional([_stencilAttachment asIndiumDescriptor]), - ((MTLBufferInternal*)_visibilityResultBuffer).buffer, + (_depthAttachment.texture == nil) ? std::nullopt : std::make_optional([_depthAttachment asIndiumDescriptor]), + (_stencilAttachment.texture == nil) ? std::nullopt : std::make_optional([_stencilAttachment asIndiumDescriptor]), + MTLBufferGetIndium(_visibilityResultBuffer), _renderTargetArrayLength, _imageblockSampleLength, _threadgroupMemoryLength, @@ -527,6 +941,22 @@ MTL_UNSUPPORTED_CLASS #if DARLING_METAL_ENABLED +{ + // The bindings this encoder has been given, so an identical re-bind can be dropped (see the + // tracking helpers above). Every encoder object is new and starts with every slot unbound, and + // nothing outside this class changes an encoder's bindings, so the records are per-encoder state. + // `_depthStencilOwner` is the strong reference the depth-stencil record needs (see the helpers). + MTLBindObjectSlots _vertexTextures; + MTLBindObjectSlots _fragmentTextures; + MTLBindObjectSlots _vertexSamplers; + MTLBindObjectSlots _fragmentSamplers; + MTLBindObjectSlots _pipelineState; + MTLBindObjectSlots _depthStencilState; + MTLBindBufferSlots _vertexBuffers; + MTLBindBufferSlots _fragmentBuffers; + std::shared_ptr _depthStencilOwner; +} + @synthesize device = _device; @synthesize encoder = _encoder; @synthesize label = _label; @@ -538,6 +968,7 @@ MTL_UNSUPPORTED_CLASS if (self != nil) { _encoder = encoder; _device = [device retain]; + passDrawIndex = 0; } return self; } @@ -553,8 +984,60 @@ MTL_UNSUPPORTED_CLASS // overridden methods // +- (void)setVertexTexture:(id)texture atIndex:(NSUInteger)index { + const std::shared_ptr& indiumTexture = MTLTextureGetIndium(texture); + if (MTLBindObjectSlot(_vertexTextures, index, MTLBindKey(indiumTexture))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipVertexTexture"); } return; } + _encoder->setVertexTexture(indiumTexture, index); +} +- (void)setFragmentTexture:(id)texture atIndex:(NSUInteger)index { + const std::shared_ptr& indiumTexture = MTLTextureGetIndium(texture); + if (MTLBindObjectSlot(_fragmentTextures, index, MTLBindKey(indiumTexture))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipFragmentTexture"); } return; } + if (MTLDrawTracking() && index < 31) debugDrawState.fragmentTextures[index] = texture; + _encoder->setFragmentTexture(indiumTexture, index); +} +- (void)setVertexSamplerState:(id)sampler atIndex:(NSUInteger)index { + const std::shared_ptr& indiumSampler = MTLSamplerGetIndium(sampler); + if (MTLBindObjectSlot(_vertexSamplers, index, MTLBindKey(indiumSampler))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipVertexSamplerState"); } return; } + _encoder->setVertexSamplerState(indiumSampler, index); +} +- (void)setFragmentSamplerState:(id)sampler atIndex:(NSUInteger)index { + const std::shared_ptr& indiumSampler = MTLSamplerGetIndium(sampler); + if (MTLBindObjectSlot(_fragmentSamplers, index, MTLBindKey(indiumSampler))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipFragmentSamplerState"); } return; } + _encoder->setFragmentSamplerState(indiumSampler, index); +} +- (void)setDepthBias:(float)depthBias slopeScale:(float)slopeScale clamp:(float)clamp { + _encoder->setDepthBias(depthBias, slopeScale, clamp); +} +- (void)setDepthStencilState:(id)state { + const std::shared_ptr& indiumState = MTLDepthStencilGetIndium(state); + // Indium's setter does not keep the state (it reads the descriptor and issues the Vulkan commands), + // so this record holds the reference that keeps the recorded address from being reused instead. + if (MTLBindObjectSlot(_depthStencilState, 0, MTLBindKey(indiumState))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipDepthStencilState"); } return; } + _depthStencilOwner = indiumState; + if (MTLDrawTracking()) debugDrawState.depthStencil = state; + if (MTLLiveFlag(2)) return; + _encoder->setDepthStencilState(indiumState); +} +- (void)setStencilReferenceValue:(uint32_t)value { + _encoder->setStencilReferenceValue(value); +} +- (void)memoryBarrierWithScope:(MTLBarrierScope)scope afterStages:(MTLRenderStages)after beforeStages:(MTLRenderStages)before { + _encoder->memoryBarrier(scope, static_cast(after), static_cast(before)); +} + - (void)endEncoding { + // Nothing may be bound after this, and the Indium encoder releases the resources it held, so the + // records must not survive the pass (a later bind must never be compared against them again). + MTLBindForgetAll(_vertexTextures); + MTLBindForgetAll(_fragmentTextures); + MTLBindForgetAll(_vertexSamplers); + MTLBindForgetAll(_fragmentSamplers); + MTLBindForgetAll(_pipelineState); + MTLBindForgetAll(_depthStencilState); + MTLBindForgetAll(_vertexBuffers); + MTLBindForgetAll(_fragmentBuffers); + _depthStencilOwner.reset(); _encoder->endEncoding(); } @@ -564,26 +1047,37 @@ MTL_UNSUPPORTED_CLASS - (void)setRenderPipelineState: (id)pipelineState { - _encoder->setRenderPipelineState(((MTLRenderPipelineStateInternal*)pipelineState).state); + const std::shared_ptr& indiumPipelineState = MTLRenderPipelineGetIndium(pipelineState); + // dropped re-binds are not counted as "setPipeline" in the diagnostics, they are not calls + if (MTLBindObjectSlot(_pipelineState, 0, MTLBindKey(indiumPipelineState))) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipPipelineState"); } return; } + if (MTLDrawTracking()) debugDrawState.pipeline = pipelineState; + MTLTraceCount(pipelineState ? "setPipeline" : "setPipelineNil"); + _encoder->setRenderPipelineState(indiumPipelineState); } - (void)setTriangleFillMode: (MTLTriangleFillMode)fillMode { + if (MTLDrawTracking()) debugDrawState.fill = (long) fillMode; _encoder->setTriangleFillMode(static_cast(fillMode)); } - (void)setFrontFacingWinding: (MTLWinding)frontFacingWinding { + if (MTLDrawTracking()) debugDrawState.winding = (long) frontFacingWinding; + if (MTLLiveFlag(1)) frontFacingWinding = (frontFacingWinding == MTLWindingClockwise) ? MTLWindingCounterClockwise : MTLWindingClockwise; _encoder->setFrontFacingWinding(static_cast(frontFacingWinding)); } - (void)setCullMode: (MTLCullMode)cullMode { + if (MTLDrawTracking()) debugDrawState.cull = (long) cullMode; + if (MTLLiveFlag(0)) cullMode = MTLCullModeNone; _encoder->setCullMode(static_cast(cullMode)); } - (void)setViewport: (MTLViewport)viewport { + if (MTLDrawTracking()) debugDrawState.viewport = viewport; _encoder->setViewport(MTLViewportToIndium(viewport)); } @@ -599,6 +1093,8 @@ MTL_UNSUPPORTED_CLASS - (void)setScissorRect: (MTLScissorRect)rect { + if (MTLDrawTracking()) debugDrawState.scissor = rect; + if (MTL_DEBUG_FLAG("metal-debug-noscissor")) return; _encoder->setScissorRect(MTLScissorRectToIndium(rect)); } @@ -616,7 +1112,27 @@ MTL_UNSUPPORTED_CLASS offset: (NSUInteger)offset atIndex: (NSUInteger)index { - _encoder->setVertexBuffer(((MTLBufferInternal*)buffer).buffer, offset, index); + const std::shared_ptr& indiumBuffer = MTLBufferGetIndium(buffer); + if (MTLBindBufferSlot(_vertexBuffers, index, MTLBindKey(indiumBuffer), offset)) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipVertexBuffer"); } return; } + if (MTLDebugEnabled()) { + static unsigned long calls = 0; + static unsigned long unusual = 0; + bool other = index != 3 && index != 4 && index != 10; + if (other && ++unusual <= 200) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "setVertexBuffer-other idx=%lu buf=%p off=%lu\n", (unsigned long) index, buffer, (unsigned long) offset); + fclose(log); + } + } + if (++calls <= 120) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "setVertexBuffer #%lu idx=%lu buf=%p off=%lu len=%lu encoder=%p\n", calls, (unsigned long) index, buffer, (unsigned long) offset, (unsigned long) buffer.length, self); + fclose(log); + } + } + } + if (MTLDrawTracking() && index < 31) { debugDrawState.vertexBuffers[index] = buffer; debugDrawState.vertexOffsets[index] = offset; debugDrawState.vertexBytesLength[index] = 0; } + _encoder->setVertexBuffer(indiumBuffer, offset, index); } - (void)setVertexBuffers: (const id*)buffers @@ -626,7 +1142,17 @@ MTL_UNSUPPORTED_CLASS std::vector> tmp; std::vector tmp2(offsets, offsets + range.length); for (size_t i = 0; i < range.length; ++i) { - tmp.push_back(((MTLBufferInternal*)buffers[i]).buffer); + tmp.push_back(MTLBufferGetIndium(buffers[i])); + } + if (MTLDrawTracking()) { + for (NSUInteger i = 0; i < range.length && range.location + i < 31; ++i) { + debugDrawState.vertexBuffers[range.location + i] = buffers[i]; + debugDrawState.vertexOffsets[range.location + i] = offsets[i]; + debugDrawState.vertexBytesLength[range.location + i] = 0; + } + } + for (NSUInteger i = 0; i < range.length; ++i) { + MTLBindBufferSlot(_vertexBuffers, range.location + i, MTLBindKey(tmp[i]), offsets[i]); } _encoder->setVertexBuffers(tmp, tmp2, NSRangeToIndium(range)); } @@ -634,6 +1160,17 @@ MTL_UNSUPPORTED_CLASS - (void)setVertexBufferOffset: (NSUInteger)offset atIndex: (NSUInteger)index { + if (MTLBindBufferOffset(_vertexBuffers, index, offset)) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipVertexBufferOffset"); } return; } + if (MTLDebugEnabled()) { + static unsigned long calls = 0; + if (++calls <= 120) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "setVertexBufferOffset #%lu idx=%lu off=%lu encoder=%p\n", calls, (unsigned long) index, (unsigned long) offset, self); + fclose(log); + } + } + } + if (MTLDrawTracking() && index < 31) debugDrawState.vertexOffsets[index] = offset; _encoder->setVertexBufferOffset(offset, index); } @@ -641,6 +1178,8 @@ MTL_UNSUPPORTED_CLASS length: (NSUInteger)length atIndex: (NSUInteger)index { + MTLBindForgetBufferSlot(_vertexBuffers, index); + if (MTLDrawTracking() && index < 31) { debugDrawState.vertexBuffers[index] = nil; debugDrawState.vertexBytesLength[index] = length; memcpy(debugDrawState.vertexBytes[index], bytes, std::min(64, length)); } _encoder->setVertexBytes(bytes, length, index); } @@ -648,7 +1187,10 @@ MTL_UNSUPPORTED_CLASS offset: (NSUInteger)offset atIndex: (NSUInteger)index { - _encoder->setFragmentBuffer(((MTLBufferInternal*)buffer).buffer, offset, index); + const std::shared_ptr& indiumBuffer = MTLBufferGetIndium(buffer); + if (MTLBindBufferSlot(_fragmentBuffers, index, MTLBindKey(indiumBuffer), offset)) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipFragmentBuffer"); } return; } + if (MTLDrawTracking() && index < 31) { debugDrawState.fragmentBuffers[index] = buffer; debugDrawState.fragmentOffsets[index] = offset; debugDrawState.fragmentBytesLength[index] = 0; } + _encoder->setFragmentBuffer(indiumBuffer, offset, index); } - (void)setFragmentBuffers: (const id*)buffers @@ -658,7 +1200,17 @@ MTL_UNSUPPORTED_CLASS std::vector> tmp; std::vector tmp2(offsets, offsets + range.length); for (size_t i = 0; i < range.length; ++i) { - tmp.push_back(((MTLBufferInternal*)buffers[i]).buffer); + tmp.push_back(MTLBufferGetIndium(buffers[i])); + } + if (MTLDrawTracking()) { + for (NSUInteger i = 0; i < range.length && range.location + i < 31; ++i) { + debugDrawState.fragmentBuffers[range.location + i] = buffers[i]; + debugDrawState.fragmentOffsets[range.location + i] = offsets[i]; + debugDrawState.fragmentBytesLength[range.location + i] = 0; + } + } + for (NSUInteger i = 0; i < range.length; ++i) { + MTLBindBufferSlot(_fragmentBuffers, range.location + i, MTLBindKey(tmp[i]), offsets[i]); } _encoder->setFragmentBuffers(tmp, tmp2, NSRangeToIndium(range)); } @@ -666,6 +1218,8 @@ MTL_UNSUPPORTED_CLASS - (void)setFragmentBufferOffset: (NSUInteger)offset atIndex: (NSUInteger)index { + if (MTLBindBufferOffset(_fragmentBuffers, index, offset)) { if (MTLBindSkipsEnabled()) { MTLBindSkipCount("bindSkipFragmentBufferOffset"); } return; } + if (MTLDrawTracking() && index < 31) debugDrawState.fragmentOffsets[index] = offset; _encoder->setFragmentBufferOffset(offset, index); } @@ -673,6 +1227,8 @@ MTL_UNSUPPORTED_CLASS length: (NSUInteger)length atIndex: (NSUInteger)index { + MTLBindForgetBufferSlot(_fragmentBuffers, index); + if (MTLDrawTracking() && index < 31) { debugDrawState.fragmentBuffers[index] = nil; debugDrawState.fragmentBytesLength[index] = length; memcpy(debugDrawState.fragmentBytes[index], bytes, std::min(64, length)); } _encoder->setFragmentBytes(bytes, length, index); } @@ -682,6 +1238,9 @@ MTL_UNSUPPORTED_CLASS instanceCount: (NSUInteger)instanceCount baseInstance: (NSUInteger)baseInstance { + if (MTLDrawLimited(vertexCount)) return; + MTLTraceCount("draw"); + if (MTLDrawTracking()) MTLDebugLogDraw("direct", vertexCount, 0, 0, 0, nil, 0); _encoder->drawPrimitives(static_cast(primitiveType), vertexStart, vertexCount, instanceCount, baseInstance); } @@ -690,6 +1249,9 @@ MTL_UNSUPPORTED_CLASS vertexCount: (NSUInteger)vertexCount instanceCount: (NSUInteger)instanceCount { + if (MTLDrawLimited(vertexCount)) return; + MTLTraceCount("draw"); + if (MTLDrawTracking()) MTLDebugLogDraw("direct", vertexCount, 0, 0, 0, nil, 0); _encoder->drawPrimitives(static_cast(primitiveType), vertexStart, vertexCount, instanceCount); } @@ -697,6 +1259,9 @@ MTL_UNSUPPORTED_CLASS vertexStart: (NSUInteger)vertexStart vertexCount: (NSUInteger)vertexCount { + if (MTLDrawLimited(vertexCount)) return; + MTLTraceCount("draw"); + if (MTLDrawTracking()) MTLDebugLogDraw("direct", vertexCount, 0, 0, 0, nil, 0); _encoder->drawPrimitives(static_cast(primitiveType), vertexStart, vertexCount); } @@ -709,7 +1274,10 @@ MTL_UNSUPPORTED_CLASS baseVertex: (NSInteger)baseVertex baseInstance: (NSUInteger)baseInstance { - _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), ((MTLBufferInternal*)indexBuffer).buffer, indexBufferOffset, instanceCount, baseVertex, baseInstance); + if (MTLDrawLimited(indexCount)) return; + MTLTraceCount("drawIndexed"); + if (MTLDrawTracking()) MTLDebugLogDraw("indexed", indexCount, instanceCount, (unsigned long) indexType, (long) baseVertex, indexBuffer, indexBufferOffset); + _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), MTLBufferGetIndium(indexBuffer), indexBufferOffset, instanceCount, baseVertex, baseInstance); } - (void)drawIndexedPrimitives: (MTLPrimitiveType)primitiveType @@ -719,7 +1287,10 @@ MTL_UNSUPPORTED_CLASS indexBufferOffset: (NSUInteger)indexBufferOffset instanceCount: (NSUInteger)instanceCount { - _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), ((MTLBufferInternal*)indexBuffer).buffer, indexBufferOffset, instanceCount); + if (MTLDrawLimited(indexCount)) return; + MTLTraceCount("drawIndexed"); + if (MTLDrawTracking()) MTLDebugLogDraw("indexed", indexCount, instanceCount, (unsigned long) indexType, 0, indexBuffer, indexBufferOffset); + _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), MTLBufferGetIndium(indexBuffer), indexBufferOffset, instanceCount); } - (void)drawIndexedPrimitives: (MTLPrimitiveType)primitiveType @@ -728,7 +1299,10 @@ MTL_UNSUPPORTED_CLASS indexBuffer: (id)indexBuffer indexBufferOffset: (NSUInteger)indexBufferOffset { - _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), ((MTLBufferInternal*)indexBuffer).buffer, indexBufferOffset); + if (MTLDrawLimited(indexCount)) return; + MTLTraceCount("drawIndexed"); + if (MTLDrawTracking()) MTLDebugLogDraw("indexed", indexCount, 0, (unsigned long) indexType, 0, indexBuffer, indexBufferOffset); + _encoder->drawIndexedPrimitives(static_cast(primitiveType), indexCount, static_cast(indexType), MTLBufferGetIndium(indexBuffer), indexBufferOffset); } #else diff --git a/src/Metal/MTLRenderPipeline.mm b/src/Metal/MTLRenderPipeline.mm index 8aa709b..140d0da 100644 --- a/src/Metal/MTLRenderPipeline.mm +++ b/src/Metal/MTLRenderPipeline.mm @@ -6,6 +6,7 @@ #import #import #import +#import #include @@ -204,7 +205,7 @@ MTL_UNSUPPORTED_CLASS ((MTLFunctionInternal*)_fragmentFunction).function, _maxVertexCallStackDepth, _maxFragmentCallStackDepth, - std::nullopt, // TODO + _vertexDescriptor ? std::make_optional([_vertexDescriptor asIndiumDescriptor]) : std::nullopt, [_vertexBuffers asIndiumDescriptors], [_fragmentBuffers asIndiumDescriptors], [_colorAttachments asContiguousIndiumDescriptors], @@ -317,7 +318,10 @@ MTL_UNSUPPORTED_CLASS std::vector result; for (const auto& [key, value]: tmp2) { - result.push_back(value); + if (value.pixelFormat != Indium::PixelFormat::Invalid) { + result.resize(key + 1); + result[key] = value; + } } return result; @@ -420,6 +424,10 @@ MTL_UNSUPPORTED_CLASS #if DARLING_METAL_ENABLED @synthesize state = _state; +const std::shared_ptr& MTLRenderPipelineGetIndium(id object) { + static const std::shared_ptr none; + return object ? ((MTLRenderPipelineStateInternal*) object)->_state : none; +} @synthesize device = _device; @synthesize label = _label; diff --git a/src/Metal/MTLSamplerDescriptor.mm b/src/Metal/MTLSamplerDescriptor.mm index 242d285..f82d083 100644 --- a/src/Metal/MTLSamplerDescriptor.mm +++ b/src/Metal/MTLSamplerDescriptor.mm @@ -1,24 +1,69 @@ -// SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 - +#import #import -#import -#import - +#include @implementation MTLSamplerDescriptor - +@synthesize minFilter = _minFilter; +@synthesize magFilter = _magFilter; +@synthesize mipFilter = _mipFilter; +@synthesize maxAnisotropy = _maxAnisotropy; +@synthesize sAddressMode = _sAddressMode; +@synthesize tAddressMode = _tAddressMode; +@synthesize rAddressMode = _rAddressMode; +@synthesize borderColor = _borderColor; +@synthesize normalizedCoordinates = _normalizedCoordinates; +@synthesize lodMinClamp = _lodMinClamp; +@synthesize lodMaxClamp = _lodMaxClamp; +@synthesize supportArgumentBuffers = _supportArgumentBuffers; +@synthesize compareFunction = _compareFunction; +@synthesize label = _label; +- (instancetype)init { + self = [super init]; + if (self) { + _maxAnisotropy = 1; + _normalizedCoordinates = YES; + _lodMaxClamp = FLT_MAX; + } + return self; +} +- (void)dealloc { [_label release]; [super dealloc]; } +- (instancetype)copyWithZone:(NSZone*)zone { + MTLSamplerDescriptor* copy = [[[self class] allocWithZone:zone] init]; + copy.minFilter = _minFilter; + copy.magFilter = _magFilter; + copy.mipFilter = _mipFilter; + copy.maxAnisotropy = _maxAnisotropy; + copy.sAddressMode = _sAddressMode; + copy.tAddressMode = _tAddressMode; + copy.rAddressMode = _rAddressMode; + copy.borderColor = _borderColor; + copy.normalizedCoordinates = _normalizedCoordinates; + copy.lodMinClamp = _lodMinClamp; + copy.lodMaxClamp = _lodMaxClamp; + copy.supportArgumentBuffers = _supportArgumentBuffers; + copy.compareFunction = _compareFunction; + copy.label = _label; + return copy; +} +@end #if DARLING_METAL_ENABLED - -- (NSMethodSignature *)methodSignatureForSelector:(SEL)aSelector -{ - return [NSMethodSignature signatureWithObjCTypes: "v@:"]; +@implementation MTLSamplerStateInternal +@synthesize sampler = _sampler; +const std::shared_ptr& MTLSamplerGetIndium(id object) { + static const std::shared_ptr none; + return object ? ((MTLSamplerStateInternal*) object)->_sampler : none; } - -- (void)forwardInvocation:(NSInvocation *)anInvocation -{ - NSLog(@"Stub called: %@ in %@", NSStringFromSelector([anInvocation selector]), [self class]); +@synthesize device = _device; +@synthesize label = _label; +- (instancetype)initWithSampler:(std::shared_ptr)sampler device:(id)device label:(NSString*)label { + self = [super init]; + if (self) { + _sampler = sampler; + _device = [device retain]; + _label = [label copy]; + } + return self; } - -#endif - +- (void)dealloc { [_device release]; [_label release]; [super dealloc]; } @end +#endif diff --git a/src/Metal/MTLStencilDescriptor.mm b/src/Metal/MTLStencilDescriptor.mm index 66a9056..f28235c 100644 --- a/src/Metal/MTLStencilDescriptor.mm +++ b/src/Metal/MTLStencilDescriptor.mm @@ -1,24 +1,35 @@ // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 -#import -#import -#import - +#import @implementation MTLStencilDescriptor - -#if DARLING_METAL_ENABLED - -- (NSMethodSignature *)methodSignatureForSelector:(SEL)aSelector -{ - return [NSMethodSignature signatureWithObjCTypes: "v@:"]; +@synthesize stencilCompareFunction = _stencilCompareFunction; +@synthesize stencilFailureOperation = _stencilFailureOperation; +@synthesize depthFailureOperation = _depthFailureOperation; +@synthesize depthStencilPassOperation = _depthStencilPassOperation; +@synthesize readMask = _readMask; +@synthesize writeMask = _writeMask; +- (instancetype)init { + self = [super init]; + if (self) { _stencilCompareFunction = MTLCompareFunctionAlways; _readMask = UINT32_MAX; _writeMask = UINT32_MAX; } + return self; } - -- (void)forwardInvocation:(NSInvocation *)anInvocation -{ - NSLog(@"Stub called: %@ in %@", NSStringFromSelector([anInvocation selector]), [self class]); +- (id)copyWithZone:(NSZone*)zone { + MTLStencilDescriptor* copy = [[[self class] allocWithZone:zone] init]; + copy.stencilCompareFunction = _stencilCompareFunction; + copy.stencilFailureOperation = _stencilFailureOperation; + copy.depthFailureOperation = _depthFailureOperation; + copy.depthStencilPassOperation = _depthStencilPassOperation; + copy.readMask = _readMask; + copy.writeMask = _writeMask; + return copy; +} +#if DARLING_METAL_ENABLED +- (Indium::StencilDescriptor)asIndiumDescriptor { + return {static_cast(_stencilCompareFunction), + static_cast(_stencilFailureOperation), + static_cast(_depthFailureOperation), + static_cast(_depthStencilPassOperation), _readMask, _writeMask}; } - #endif - @end diff --git a/src/Metal/MTLTexture.mm b/src/Metal/MTLTexture.mm index 6c71fb9..186b8c0 100644 --- a/src/Metal/MTLTexture.mm +++ b/src/Metal/MTLTexture.mm @@ -1,3 +1,5 @@ +#import +#include // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 @@ -7,11 +9,38 @@ #import #import + +#if DARLING_METAL_ENABLED +// Diagnostic: CPU texture uploads (replaceRegion) block on the GPU queue, so count, time and sample them. +static void MTLDebugReplaceRegion(id texture, MTLRegion region, NSUInteger level, const void* bytes, double seconds) +{ + MTLTraceCount("replaceRegion"); + MTLTraceTiming("replaceRegion", seconds); + static unsigned long logged = 0; + if ((MTLDebugEnabled() || MTLTimingEnabled()) && ++logged <= 60) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + const unsigned char* b = (const unsigned char*) bytes; + unsigned long nonZero = 0; + // Every pixel format takes at least one byte per pixel, so the first `width` bytes lie inside the first row. + size_t probe = std::min(4096, (size_t) region.size.width); + for (size_t i = 0; i < probe; ++i) if (b[i]) ++nonZero; + fprintf(log, "replaceRegion tex=%p %lux%lu fmt=%lu level=%lu region=(%lu,%lu %lux%lu) firstBytesNonZero=%lu/%zu took=%.1fms\n", texture, + (unsigned long) texture.width, (unsigned long) texture.height, (unsigned long) texture.pixelFormat, (unsigned long) level, + (unsigned long) region.origin.x, (unsigned long) region.origin.y, (unsigned long) region.size.width, (unsigned long) region.size.height, nonZero, probe, seconds * 1000); + fclose(log); + } + } +} +#endif @implementation MTLTextureInternal #if DARLING_METAL_ENABLED @synthesize texture = _texture; +const std::shared_ptr& MTLTextureGetIndium(id object) { + static const std::shared_ptr none; + return object ? ((MTLTextureInternal*) object)->_texture : none; +} @synthesize device = _device; @synthesize resourceOptions = _resourceOptions; @synthesize label = _label; @@ -164,7 +193,9 @@ withBytes: (const void*)pixelBytes bytesPerRow: (NSUInteger)bytesPerRow { + double start = (MTLDebugEnabled() || MTLTimingEnabled()) ? MTLTraceNow() : 0; _texture->replaceRegion(MTLRegionToIndium(region), level, pixelBytes, bytesPerRow); + if (start != 0) MTLDebugReplaceRegion(self, region, level, pixelBytes, MTLTraceNow() - start); } - (void)replaceRegion: (MTLRegion)region @@ -174,7 +205,9 @@ bytesPerRow: (NSUInteger)bytesPerRow bytesPerImage: (NSUInteger)bytesPerImage { + double start = (MTLDebugEnabled() || MTLTimingEnabled()) ? MTLTraceNow() : 0; _texture->replaceRegion(MTLRegionToIndium(region), level, slice, pixelBytes, bytesPerRow, bytesPerImage); + if (start != 0) MTLDebugReplaceRegion(self, region, level, pixelBytes, MTLTraceNow() - start); } - (void)getBytes: (void*)pixelBytes diff --git a/src/Metal/MTLTextureDescriptor.mm b/src/Metal/MTLTextureDescriptor.mm index 37cccb3..79b2350 100644 --- a/src/Metal/MTLTextureDescriptor.mm +++ b/src/Metal/MTLTextureDescriptor.mm @@ -1,24 +1,64 @@ -// SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 - #import -#import -#import - @implementation MTLTextureDescriptor - -#if DARLING_METAL_ENABLED - -- (NSMethodSignature *)methodSignatureForSelector:(SEL)aSelector -{ - return [NSMethodSignature signatureWithObjCTypes: "v@:"]; +@synthesize textureType = _textureType; +@synthesize pixelFormat = _pixelFormat; +@synthesize width = _width; +@synthesize height = _height; +@synthesize depth = _depth; +@synthesize mipmapLevelCount = _mipmapLevelCount; +@synthesize sampleCount = _sampleCount; +@synthesize arrayLength = _arrayLength; +@synthesize resourceOptions = _resourceOptions; +@synthesize allowGPUOptimizedContents = _allowGPUOptimizedContents; +@synthesize usage = _usage; +@synthesize swizzle = _swizzle; +- (instancetype)init { + self = [super init]; + if (self) { + _textureType = MTLTextureType2D; + _pixelFormat = MTLPixelFormatRGBA8Unorm; + _width = _height = _depth = _mipmapLevelCount = _sampleCount = _arrayLength = 1; + _resourceOptions = MTLResourceStorageModeManaged; + _allowGPUOptimizedContents = YES; + _swizzle = MTLTextureSwizzleChannelsDefault; + } + return self; } - -- (void)forwardInvocation:(NSInvocation *)anInvocation -{ - NSLog(@"Stub called: %@ in %@", NSStringFromSelector([anInvocation selector]), [self class]); +- (instancetype)copyWithZone:(NSZone*)zone { + MTLTextureDescriptor* copy = [[[self class] allocWithZone:zone] init]; + copy.textureType = _textureType; + copy.pixelFormat = _pixelFormat; + copy.width = _width; + copy.height = _height; + copy.depth = _depth; + copy.mipmapLevelCount = _mipmapLevelCount; + copy.sampleCount = _sampleCount; + copy.arrayLength = _arrayLength; + copy.resourceOptions = _resourceOptions; + copy.allowGPUOptimizedContents = _allowGPUOptimizedContents; + copy.usage = _usage; + copy.swizzle = _swizzle; + return copy; +} +- (MTLCPUCacheMode)cpuCacheMode { return static_cast((_resourceOptions >> 0) & 0xf); } +- (void)setCpuCacheMode:(MTLCPUCacheMode)value { _resourceOptions = static_cast((_resourceOptions & ~(0xful << 0)) | (static_cast(value) << 0)); } +- (MTLStorageMode)storageMode { return static_cast((_resourceOptions >> 4) & 0xf); } +- (void)setStorageMode:(MTLStorageMode)value { _resourceOptions = static_cast((_resourceOptions & ~(0xful << 4)) | (static_cast(value) << 4)); } +- (MTLHazardTrackingMode)hazardTrackingMode { return static_cast((_resourceOptions >> 8) & 0xf); } +- (void)setHazardTrackingMode:(MTLHazardTrackingMode)value { _resourceOptions = static_cast((_resourceOptions & ~(0xful << 8)) | (static_cast(value) << 8)); } ++ (instancetype)texture2DDescriptorWithPixelFormat:(MTLPixelFormat)format width:(NSUInteger)width height:(NSUInteger)height mipmapped:(BOOL)mipmapped { + MTLTextureDescriptor* desc = [[[self alloc] init] autorelease]; + desc.pixelFormat = format; desc.width = width; desc.height = height; + if (mipmapped) { + NSUInteger size = width > height ? width : height; + while (size > 1) { ++desc->_mipmapLevelCount; size >>= 1; } + } + return desc; +} ++ (instancetype)textureCubeDescriptorWithPixelFormat:(MTLPixelFormat)format size:(NSUInteger)size mipmapped:(BOOL)mipmapped { + MTLTextureDescriptor* desc = [self texture2DDescriptorWithPixelFormat:format width:size height:size mipmapped:mipmapped]; + desc.textureType = MTLTextureTypeCube; + return desc; } - -#endif - @end diff --git a/src/Metal/MTLVertexDescriptor.mm b/src/Metal/MTLVertexDescriptor.mm index cde968d..0a016c6 100644 --- a/src/Metal/MTLVertexDescriptor.mm +++ b/src/Metal/MTLVertexDescriptor.mm @@ -1,24 +1,435 @@ // SPDX-FileCopyrightText: 2026 Darling Developers // SPDX-License-Identifier: MPL-2.0 -#import -#import -#import - -@implementation MTLVertexDescriptor - -#if DARLING_METAL_ENABLED - -- (NSMethodSignature *)methodSignatureForSelector:(SEL)aSelector +#import +#import +#import +// The bridge owns these concrete descriptor implementations. Keep their +// scalar storage accessible within this translation unit so reset/conversion +// can batch it. Subclasses continue through their public property methods. +// Every item is created by an array for one slot, and the descriptor resets a slot by writing +// the defaults into the item that holds it (in place, so an item a caller kept reads them back). +// The array therefore hands each item the word and bit of the descriptor mask that reset reads +// (`_dirtyWord`/`_dirtyBit`, see MTLVertexDescriptor below): a setter marks its own slot, and +// reset only has to visit the slots that may differ from the defaults instead of every item in +// the table. An item that was not handed a slot (a copy, or an array nobody owns) marks nothing, +// which is what the "always" mask in the descriptor is for. +@interface MTLVertexBufferLayoutDescriptor () { +@public + NSUInteger _stride; + MTLVertexStepFunction _stepFunction; + NSUInteger _stepRate; + uint32_t* _dirtyWord; // the descriptor's layout mask, or NULL when the item has no slot + uint32_t _dirtyBit; // 1u << slot, meaningful while _dirtyWord is not NULL +} +@end +@interface MTLVertexAttributeDescriptor () { +@public + MTLVertexFormat _format; + NSUInteger _offset; + NSUInteger _bufferIndex; + uint32_t* _dirtyWord; // the descriptor's attribute mask, or NULL when the item has no slot + uint32_t _dirtyBit; // 1u << slot, meaningful while _dirtyWord is not NULL +} +@end +// The masks that make -reset cheap, declared here so the two arrays (same file) can set them. +// (darling-lol:) A descriptor's reset used to write the defaults into every item the arrays had +// ever handed out - 16-31 attributes plus 16-31 layouts for the game's descriptors, each item a +// separate allocation last touched a frame ago, i.e. a cache miss per item even though only the +// few slots the game refilled in the previous pass could differ from the defaults. The game +// profile put 1.5% of the main thread there (build/perf-research/metal-vdesc/REPORT.md). +// `_attrDirty`/`_layoutDirty` name the slots whose values may differ from the defaults and must be +// written by the next reset; `_attrPinned`/`_layoutPinned` name the slots whose item is not +// tracked (a subclass instance, an item installed with setObject:, or a copy), which are written +// every time. Over-marking is always safe - the reset writes the defaults either way - so the +// only invariant that has to hold is that a slot whose value can differ from the defaults is +// marked. Mutations mark through the item (see MTLVertexDescriptorItemTouched), which is why +// every fallback pins the slot instead of tracking it. +@interface MTLVertexDescriptor () { +@public + MTLVertexBufferLayoutDescriptorArray* _layouts; + MTLVertexAttributeDescriptorArray* _attributes; + uint32_t _attrDirty; // attributes: slots to write at the next reset + uint32_t _attrPinned; // attributes: slots whose item cannot be tracked + uint32_t _layoutDirty; // layouts: slots to write at the next reset + uint32_t _layoutPinned; // layouts: slots whose item cannot be tracked +} +@end +// "Is this item exactly the bridge class?" - the class that declares the plain ivars both the +// reset and the marking write. A hook rather than the inline expression it replaces, so the host +// model (tools/riot-metal-bind-model) can compile the helpers that call it. +static inline bool MTLVertexBufferLayoutItemIsExactClass(const MTLVertexBufferLayoutDescriptor* item) { - return [NSMethodSignature signatureWithObjCTypes: "v@:"]; + return object_getClass((id) item) == [MTLVertexBufferLayoutDescriptor class]; } - -- (void)forwardInvocation:(NSInvocation *)anInvocation +static inline bool MTLVertexAttributeItemIsExactClass(const MTLVertexAttributeDescriptor* item) { - NSLog(@"Stub called: %@ in %@", NSStringFromSelector([anInvocation selector]), [self class]); + return object_getClass((id) item) == [MTLVertexAttributeDescriptor class]; +} +// Marking an untouched item is harmless (reset would write the same default value), so the +// marking is allowed to over-approximate; it must never miss a mutation. See the model in +// tools/riot-metal-bind-model/ for the mutation paths this covers. +static inline void MTLVertexDescriptorItemTouched(uint32_t* word, uint32_t bit) +{ + if (word) *word |= bit; +} +// The decision a -setObject:atIndexedSubscript: makes for one slot: 1 when the slot's item is one +// this descriptor tracks at exactly this slot (the caller marks the dirty bit), 2 when the slot +// has to be pinned because the item is not trackable (a copy with no slot, an item of another +// descriptor, or the same item at another slot). Only the concrete classes declare the two fields, +// so a subclass instance is never passed here (see the caller). +static inline int MTLVertexSlotTrackingFor(uint32_t* itemWord, uint32_t itemBit, uint32_t* ownerWord, uint32_t bit) +{ + if (!ownerWord) return 2; + if (!itemWord) return 2; + if (itemWord != ownerWord) return 2; + return itemBit == bit ? 1 : 2; +} +static bool MTLVertexDescriptorResetCounters() +{ + static const bool enabled = MTLDebugEnabled() || access("/tmp/metal-debug-binds", F_OK) == 0; + return enabled; +} +static void MTLVertexDescriptorResetCount(unsigned long written, unsigned long skipped) +{ + static std::mutex lock; + static unsigned long calls = 0, writtenTotal = 0, skippedTotal = 0; + static time_t last = 0; + std::lock_guard guard(lock); + ++calls; writtenTotal += written; skippedTotal += skipped; + time_t now = time(nullptr); + if (now - last >= 3) { + last = now; + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "vdesc-reset@%ld: calls=%lu slotsWritten=%lu slotsAtDefaults=%lu\n", + (long) now, calls, writtenTotal, skippedTotal); + fclose(log); + } + } +} +@implementation MTLVertexBufferLayoutDescriptor +@synthesize stride = _stride; +@synthesize stepFunction = _stepFunction; +@synthesize stepRate = _stepRate; +// The setters are hand-written so they can mark their slot (over-marking never changes what a +// reset writes); the getters stay synthesized. +- (void)setStride:(NSUInteger)stride { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _stride = stride; } +- (void)setStepFunction:(MTLVertexStepFunction)stepFunction { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _stepFunction = stepFunction; } +- (void)setStepRate:(NSUInteger)stepRate { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _stepRate = stepRate; } +- (instancetype)init { self = [super init]; if (self) { _stepFunction = MTLVertexStepFunctionPerVertex; _stepRate = 1; } return self; } +- (id)copyWithZone:(NSZone*)zone { MTLVertexBufferLayoutDescriptor* copy = [[[self class] allocWithZone:zone] init]; copy.stride = _stride; copy.stepFunction = _stepFunction; copy.stepRate = _stepRate; return copy; } +@end +// The array storage is visible within this file (same pattern as the item classes above) so that -[MTLVertexDescriptor reset] +// can run the in-place reset without a message send per array: the game resets a vertex descriptor for every draw. Both arrays +// are only ever created by -init/-reset/-copyWithZone: below, and the two properties are readonly, so the storage a descriptor +// holds is always the concrete class declared here. +@interface MTLVertexBufferLayoutDescriptorArray () { +@public + MTLVertexBufferLayoutDescriptor* _items[31]; + uint32_t _present; // bit i: _items[i] exists + uint32_t _concrete; // bit i: _items[i] is exactly MTLVertexBufferLayoutDescriptor (ivars can be written directly) + MTLVertexDescriptor* _owner; // the descriptor whose mask the items mark (not retained; nil after it dies) +} +- (MTLVertexBufferLayoutDescriptor*)existingObjectAtIndex:(NSUInteger)index; +- (void)resetAll; +@end +@interface MTLVertexAttributeDescriptorArray () { +@public + MTLVertexAttributeDescriptor* _items[31]; + uint32_t _present; // bit i: _items[i] exists + uint32_t _concrete; // bit i: _items[i] is exactly MTLVertexAttributeDescriptor (ivars can be written directly) + MTLVertexDescriptor* _owner; // the descriptor whose mask the items mark (not retained; nil after it dies) +} +- (MTLVertexAttributeDescriptor*)existingObjectAtIndex:(NSUInteger)index; +- (void)resetAll; +@end +// A present-but-non-concrete entry is a caller-installed subclass instance (see +// -setObject:atIndexedSubscript:): it keeps its own property setters. Kept out of +// line so the common all-concrete loop does not carry the selector references and +// the message-send stub in its preheader. +__attribute__((noinline)) static void MTLVertexBufferLayoutSetDefaults(MTLVertexBufferLayoutDescriptor* item) +{ + item.stride = 0; item.stepFunction = MTLVertexStepFunctionPerVertex; item.stepRate = 1; +} +__attribute__((noinline)) static void MTLVertexAttributeSetDefaults(MTLVertexAttributeDescriptor* item) +{ + item.format = MTLVertexFormatInvalid; item.offset = 0; item.bufferIndex = 0; +} +// Reset the slots whose mask bit is set, in ascending order like the old full-table loop: an +// object a caller still holds must read back the defaults, and absent (nil) slots stay absent. +// Exact-class entries are written through their ivars; an entry that is not exactly the bridge +// class (a subclass an array was handed, or a graft) keeps its own property setters. +static void MTLVertexBufferLayoutArrayResetSlots(MTLVertexBufferLayoutDescriptorArray* array, uint32_t slots) +{ + const uint32_t exotic = slots & ~array->_concrete; + for (uint32_t bits = slots; bits; bits &= bits - 1) { + const int i = __builtin_ctz(bits); + MTLVertexBufferLayoutDescriptor* item = array->_items[i]; + if (!item) continue; + if (exotic & (1u << i)) { + MTLVertexBufferLayoutSetDefaults(item); + } else { + item->_stride = 0; item->_stepFunction = MTLVertexStepFunctionPerVertex; item->_stepRate = 1; + } + } +} +static void MTLVertexAttributeArrayResetSlots(MTLVertexAttributeDescriptorArray* array, uint32_t slots) +{ + const uint32_t exotic = slots & ~array->_concrete; + for (uint32_t bits = slots; bits; bits &= bits - 1) { + const int i = __builtin_ctz(bits); + MTLVertexAttributeDescriptor* item = array->_items[i]; + if (!item) continue; + if (exotic & (1u << i)) { + MTLVertexAttributeSetDefaults(item); + } else { + item->_format = MTLVertexFormatInvalid; item->_offset = 0; item->_bufferIndex = 0; + } + } +} +// An item that leaves its slot (a -setObject: replaces it, or installs nil) may still be held by +// the caller. If no slot of this array holds it any more, its mask pointer must be cleared: the +// descriptor's -dealloc only revokes the items its arrays still hold, so a stale pointer would +// write into the descriptor's storage after it is gone. An item that is still in a slot must keep +// its pointer - that slot is tracked through it. +static void MTLVertexBufferLayoutArrayRetireItem(MTLVertexBufferLayoutDescriptorArray* array, MTLVertexBufferLayoutDescriptor* item) +{ + if (!item || !array->_owner) return; + if (!MTLVertexBufferLayoutItemIsExactClass(item)) return; + if (item->_dirtyWord != &array->_owner->_layoutDirty) return; + for (int i = 0; i < 31; ++i) if (array->_items[i] == item) return; + item->_dirtyWord = NULL; item->_dirtyBit = 0; +} +static void MTLVertexAttributeArrayRetireItem(MTLVertexAttributeDescriptorArray* array, MTLVertexAttributeDescriptor* item) +{ + if (!item || !array->_owner) return; + if (!MTLVertexAttributeItemIsExactClass(item)) return; + if (item->_dirtyWord != &array->_owner->_attrDirty) return; + for (int i = 0; i < 31; ++i) if (array->_items[i] == item) return; + item->_dirtyWord = NULL; item->_dirtyBit = 0; +} +// The arrays are released with the descriptor, but a caller may hold an array (or an item) past +// it: the array must then not reach the dead descriptor again, and an item that still points at +// the descriptor's mask must not mark it. Both are cleared here, while the items are alive (the +// descriptor keeps the arrays, and they keep the items). +static void MTLVertexBufferLayoutArrayRevokeItems(MTLVertexBufferLayoutDescriptorArray* array, uint32_t* word) +{ + if (!array) return; + array->_owner = nil; + for (uint32_t bits = array->_present; bits; bits &= bits - 1) { + const int i = __builtin_ctz(bits); + MTLVertexBufferLayoutDescriptor* item = array->_items[i]; + if (item && MTLVertexBufferLayoutItemIsExactClass(item) && item->_dirtyWord == word) { + item->_dirtyWord = NULL; item->_dirtyBit = 0; + } + } +} +static void MTLVertexAttributeArrayRevokeItems(MTLVertexAttributeDescriptorArray* array, uint32_t* word) +{ + if (!array) return; + array->_owner = nil; + for (uint32_t bits = array->_present; bits; bits &= bits - 1) { + const int i = __builtin_ctz(bits); + MTLVertexAttributeDescriptor* item = array->_items[i]; + if (item && MTLVertexAttributeItemIsExactClass(item) && item->_dirtyWord == word) { + item->_dirtyWord = NULL; item->_dirtyBit = 0; + } + } +} +// The arrays used to be NSMutableDictionaries keyed by boxed NSNumbers, and converting a descriptor created all 31 entries; +// under a game that builds vertex descriptors constantly that was ~30% of the render thread. Fixed storage, entries created lazily. +@implementation MTLVertexBufferLayoutDescriptorArray +- (void)dealloc { for (int i = 0; i < 31; ++i) [_items[i] release]; [super dealloc]; } +- (MTLVertexBufferLayoutDescriptor*)objectAtIndexedSubscript:(NSUInteger)index { + if (index >= 31) [NSException raise:NSRangeException format:@"Vertex descriptor index %lu exceeds 30", (unsigned long)index]; + if (!_items[index]) { + _items[index] = [MTLVertexBufferLayoutDescriptor new]; _present |= 1u << index; _concrete |= 1u << index; + // a fresh item holds the defaults, so nothing is marked until it is written; the slot is + // what a later setter marks + if (_owner) { _items[index]->_dirtyWord = &_owner->_layoutDirty; _items[index]->_dirtyBit = 1u << index; } + } + return _items[index]; +} +- (MTLVertexBufferLayoutDescriptor*)existingObjectAtIndex:(NSUInteger)index { return index < 31 ? _items[index] : nil; } +// reset in place: the game calls -[MTLVertexDescriptor reset] for every draw and then refills it, so keep the objects +- (void)resetAll { + MTLVertexBufferLayoutArrayResetSlots(self, _present); + if (_owner) _owner->_layoutDirty &= ~_present; +} +- (void)setObject:(MTLVertexBufferLayoutDescriptor*)descriptor atIndexedSubscript:(NSUInteger)index { + if (index >= 31) [NSException raise:NSRangeException format:@"Vertex descriptor index %lu exceeds 30", (unsigned long)index]; + MTLVertexBufferLayoutDescriptor* old = _items[index]; + _items[index] = [descriptor retain]; + MTLVertexBufferLayoutArrayRetireItem(self, old); + [old release]; + const uint32_t bit = 1u << index; + if (descriptor) _present |= bit; else _present &= ~bit; + if (descriptor && object_getClass(descriptor) == [MTLVertexBufferLayoutDescriptor class]) _concrete |= bit; else _concrete &= ~bit; + // The next reset must write this slot unless its item is one this descriptor tracks at exactly + // this slot: a subclass with its own setters, an item that belongs to another descriptor or + // another slot, or an item with no slot at all is marked "always" instead, which keeps the + // reset exact (over-marking never changes what a reset writes). + if (_owner) { + if (!descriptor) { + _owner->_layoutDirty &= ~bit; + _owner->_layoutPinned &= ~bit; + } else { + const int tracking = MTLVertexBufferLayoutItemIsExactClass(descriptor) + ? MTLVertexSlotTrackingFor(descriptor->_dirtyWord, descriptor->_dirtyBit, &_owner->_layoutDirty, bit) + : 2; + if (tracking == 1) _owner->_layoutDirty |= bit; else _owner->_layoutPinned |= bit; + } + } +} +- (id)copyWithZone:(NSZone*)zone { + MTLVertexBufferLayoutDescriptorArray* copy = [[[self class] allocWithZone:zone] init]; + for (int i = 0; i < 31; ++i) if (_items[i]) copy->_items[i] = [_items[i] copy]; + copy->_present = _present; + copy->_concrete = _concrete; // -copy of a concrete item returns the same class + // the copied items are new objects and are not tracked; the descriptor that adopts this array + // binds them and marks their slots (see MTLVertexDescriptor -copyWithZone:) + return copy; +} +@end +@implementation MTLVertexAttributeDescriptor +@synthesize format = _format; +@synthesize offset = _offset; +@synthesize bufferIndex = _bufferIndex; +// Hand-written for the same reason as MTLVertexBufferLayoutDescriptor's: the setter marks the +// descriptor slot this item fills. +- (void)setFormat:(MTLVertexFormat)format { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _format = format; } +- (void)setOffset:(NSUInteger)offset { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _offset = offset; } +- (void)setBufferIndex:(NSUInteger)bufferIndex { MTLVertexDescriptorItemTouched(_dirtyWord, _dirtyBit); _bufferIndex = bufferIndex; } +- (id)copyWithZone:(NSZone*)zone { MTLVertexAttributeDescriptor* copy = [[[self class] allocWithZone:zone] init]; copy.format = _format; copy.offset = _offset; copy.bufferIndex = _bufferIndex; return copy; } +@end +@implementation MTLVertexAttributeDescriptorArray +- (void)dealloc { for (int i = 0; i < 31; ++i) [_items[i] release]; [super dealloc]; } +- (MTLVertexAttributeDescriptor*)objectAtIndexedSubscript:(NSUInteger)index { + if (index >= 31) [NSException raise:NSRangeException format:@"Vertex descriptor index %lu exceeds 30", (unsigned long)index]; + if (!_items[index]) { + _items[index] = [MTLVertexAttributeDescriptor new]; _present |= 1u << index; _concrete |= 1u << index; + // a fresh item holds the defaults; the slot is what a later setter marks + if (_owner) { _items[index]->_dirtyWord = &_owner->_attrDirty; _items[index]->_dirtyBit = 1u << index; } + } + return _items[index]; +} +- (MTLVertexAttributeDescriptor*)existingObjectAtIndex:(NSUInteger)index { return index < 31 ? _items[index] : nil; } +- (void)resetAll { + MTLVertexAttributeArrayResetSlots(self, _present); + if (_owner) _owner->_attrDirty &= ~_present; +} +- (void)setObject:(MTLVertexAttributeDescriptor*)descriptor atIndexedSubscript:(NSUInteger)index { + if (index >= 31) [NSException raise:NSRangeException format:@"Vertex descriptor index %lu exceeds 30", (unsigned long)index]; + MTLVertexAttributeDescriptor* old = _items[index]; + _items[index] = [descriptor retain]; + MTLVertexAttributeArrayRetireItem(self, old); + [old release]; + const uint32_t bit = 1u << index; + if (descriptor) _present |= bit; else _present &= ~bit; + if (descriptor && object_getClass(descriptor) == [MTLVertexAttributeDescriptor class]) _concrete |= bit; else _concrete &= ~bit; + // see MTLVertexBufferLayoutDescriptorArray -setObject:atIndexedSubscript: for the rule + if (_owner) { + if (!descriptor) { + _owner->_attrDirty &= ~bit; + _owner->_attrPinned &= ~bit; + } else { + const int tracking = MTLVertexAttributeItemIsExactClass(descriptor) + ? MTLVertexSlotTrackingFor(descriptor->_dirtyWord, descriptor->_dirtyBit, &_owner->_attrDirty, bit) + : 2; + if (tracking == 1) _owner->_attrDirty |= bit; else _owner->_attrPinned |= bit; + } + } +} +- (id)copyWithZone:(NSZone*)zone { + MTLVertexAttributeDescriptorArray* copy = [[[self class] allocWithZone:zone] init]; + for (int i = 0; i < 31; ++i) if (_items[i]) copy->_items[i] = [_items[i] copy]; + copy->_present = _present; + copy->_concrete = _concrete; // -copy of a concrete item returns the same class + return copy; +} +@end +// The reset work itself, kept in a static so the host model can run the shipping code (see +// tools/riot-metal-bind-model/extract-vertex-reset.py): the mask words name the slots whose values +// may differ from the defaults, and the helpers write exactly those, in ascending slot order. +static void MTLVertexDescriptorResetMarks(MTLVertexDescriptor* descriptor) +{ + const uint32_t layouts = descriptor->_layoutDirty | descriptor->_layoutPinned; + const uint32_t attributes = descriptor->_attrDirty | descriptor->_attrPinned; + descriptor->_layoutDirty = descriptor->_layoutPinned; + descriptor->_attrDirty = descriptor->_attrPinned; + MTLVertexBufferLayoutArrayResetSlots(descriptor->_layouts, layouts); + MTLVertexAttributeArrayResetSlots(descriptor->_attributes, attributes); + if (MTLVertexDescriptorResetCounters()) { + MTLVertexDescriptorResetCount(__builtin_popcount(layouts) + __builtin_popcount(attributes), + __builtin_popcount(descriptor->_layouts->_present & ~layouts) + __builtin_popcount(descriptor->_attributes->_present & ~attributes)); + } +} +// Guest diagnosis, same shape as the bind counters in MTLRenderCommandEncoder.mm: with +// /tmp/metal-debug-binds (or the trace flag) present every reset reports how many slots it had to +// write and how many it left at the defaults, so the next match measures what the marks save. +// The flag is tested inline at the call site, so an ordinary match pays one load and a branch. +@implementation MTLVertexDescriptor +@synthesize layouts = _layouts; +@synthesize attributes = _attributes; ++ (instancetype)vertexDescriptor { return [[[self alloc] init] autorelease]; } +- (instancetype)init { self = [super init]; if (self) [self reset]; return self; } +- (void)dealloc { + // an item (or an array) a caller kept must not reach this descriptor's storage after this + MTLVertexBufferLayoutArrayRevokeItems(_layouts, &_layoutDirty); + MTLVertexAttributeArrayRevokeItems(_attributes, &_attrDirty); + [_layouts release]; [_attributes release]; [super dealloc]; +} +- (void)reset { + if (_layouts && _attributes) { + // Only the slots whose values may differ from the defaults are written, in the same + // ascending order as before; the rest were left at the defaults by the previous reset (or + // hold a fresh item), so writing them would change nothing. The dirty marks are cleared as + // the slots are written; the pinned marks stay (their items are written every time). + MTLVertexDescriptorResetMarks(self); + return; + } + [_layouts release]; [_attributes release]; + _layouts = [MTLVertexBufferLayoutDescriptorArray new]; + _attributes = [MTLVertexAttributeDescriptorArray new]; + ((MTLVertexBufferLayoutDescriptorArray*)_layouts)->_owner = self; + ((MTLVertexAttributeDescriptorArray*)_attributes)->_owner = self; + _attrDirty = 0; _attrPinned = 0; _layoutDirty = 0; _layoutPinned = 0; +} +- (id)copyWithZone:(NSZone*)zone { + MTLVertexDescriptor* copy = [[[self class] allocWithZone:zone] init]; + [copy->_layouts release]; [copy->_attributes release]; + copy->_layouts = [_layouts copy]; copy->_attributes = [_attributes copy]; + // the copied items are new objects nothing tracks; bind them to the copy and let its resets + // write every slot they occupy (a copy is rare, and this keeps it exact without tracking it) + MTLVertexBufferLayoutDescriptorArray* layouts = copy->_layouts; + MTLVertexAttributeDescriptorArray* attributes = copy->_attributes; + layouts->_owner = copy; + attributes->_owner = copy; + copy->_layoutPinned = layouts->_present; + copy->_attrPinned = attributes->_present; + return copy; +} +#if DARLING_METAL_ENABLED +- (Indium::VertexDescriptor)asIndiumDescriptor { + static Class attributeClass = [MTLVertexAttributeDescriptor class]; + static Class layoutClass = [MTLVertexBufferLayoutDescriptor class]; + Indium::VertexDescriptor result; + for (NSUInteger i = 0; i < 31; ++i) { + MTLVertexAttributeDescriptor* attr = [_attributes existingObjectAtIndex: i]; + if (!attr) continue; + const bool concrete = object_getClass(attr) == attributeClass; + const auto format = concrete ? attr->_format : attr.format; + if (format == MTLVertexFormatInvalid) continue; + const auto offset = concrete ? attr->_offset : attr.offset; + const auto index = concrete ? attr->_bufferIndex : attr.bufferIndex; + result.attributes[i] = {static_cast(format), offset, index}; + MTLVertexBufferLayoutDescriptor* layout = _layouts[index]; + if (object_getClass(layout) == layoutClass) + result.layouts[index] = {layout->_stride, static_cast(layout->_stepFunction), layout->_stepRate}; + else + result.layouts[index] = {layout.stride, static_cast(layout.stepFunction), layout.stepRate}; + } + return result; } - #endif - @end diff --git a/include/Metal/MTLArgument.h b/include/Metal/MTLArgument.h new file mode 100644 --- /dev/null +++ b/include/Metal/MTLArgument.h @@ -0,0 +1,29 @@ +#ifndef _METAL_MTLARGUMENT_H_ +#define _METAL_MTLARGUMENT_H_ +#import +typedef NS_ENUM(NSUInteger, MTLArgumentType) { MTLArgumentTypeBuffer=0, MTLArgumentTypeThreadgroupMemory=1, MTLArgumentTypeTexture=2, MTLArgumentTypeSampler=3 }; +@interface MTLStructMember : NSObject +@property(readonly) NSString* name; +@property(readonly) NSUInteger offset; +@end +@interface MTLStructType : NSObject +@property(readonly) NSArray* members; +- (MTLStructMember*)memberByName:(NSString*)name; +@end +@interface MTLArgument : NSObject +@property(readonly) NSString* name; +@property(readonly) MTLArgumentType type; +@property(readonly) NSUInteger index; +@property(readonly, getter=isActive) BOOL active; +@property(readonly) NSUInteger bufferAlignment; +@property(readonly) NSUInteger bufferDataSize; +@property(readonly) MTLStructType* bufferStructType; +@end +@interface MTLComputePipelineReflection : NSObject +@property(readonly) NSArray* arguments; +@end +@interface MTLRenderPipelineReflection : NSObject +@property(readonly) NSArray* vertexArguments; +@property(readonly) NSArray* fragmentArguments; +@end +#endif diff --git a/include/Metal/MTLBlitCommandEncoder.h b/include/Metal/MTLBlitCommandEncoder.h new file mode 100644 --- /dev/null +++ b/include/Metal/MTLBlitCommandEncoder.h @@ -0,0 +1,28 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#ifndef _METAL_MTLBLITCOMMANDENCODER_H_ +#define _METAL_MTLBLITCOMMANDENCODER_H_ +#import +#import +@protocol MTLBuffer; +@protocol MTLTexture; +typedef NS_OPTIONS(NSUInteger, MTLBlitOption) { + MTLBlitOptionNone = 0, + MTLBlitOptionDepthFromDepthStencil = 1, + MTLBlitOptionStencilFromDepthStencil = 2, + MTLBlitOptionRowLinearPVRTC = 4, +}; +@protocol MTLBlitCommandEncoder +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset size:(NSUInteger)size; +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset sourceBytesPerRow:(NSUInteger)sourceBytesPerRow sourceBytesPerImage:(NSUInteger)sourceBytesPerImage sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin; +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset destinationBytesPerRow:(NSUInteger)destinationBytesPerRow destinationBytesPerImage:(NSUInteger)destinationBytesPerImage; +- (void)copyFromTexture:(id)source toTexture:(id)destination; +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel sliceCount:(NSUInteger)sliceCount levelCount:(NSUInteger)levelCount; +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin; +- (void)fillBuffer:(id)buffer range:(NSRange)range value:(uint8_t)value; +- (void)generateMipmapsForTexture:(id)texture; +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset sourceBytesPerRow:(NSUInteger)sourceBytesPerRow sourceBytesPerImage:(NSUInteger)sourceBytesPerImage sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin options:(MTLBlitOption)options; +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset destinationBytesPerRow:(NSUInteger)destinationBytesPerRow destinationBytesPerImage:(NSUInteger)destinationBytesPerImage options:(MTLBlitOption)options; +@end +#endif diff --git a/include/Metal/MTLExceptionBridge.h b/include/Metal/MTLExceptionBridge.h new file mode 100644 --- /dev/null +++ b/include/Metal/MTLExceptionBridge.h @@ -0,0 +1,23 @@ +#pragma once +#import +#include +#include +#include + +// Metal reports failures with nil and an NSError, while the Indium backend reports them with C++ +// exceptions, which the callers of the Metal API do not expect. Call from a catch block: builds the +// NSError from the active exception and records it in the guest's /tmp/metal-errors.log. +static inline NSError* MTLErrorFromCurrentException(const char* where) +{ + std::string message = "unknown error"; + try { throw; } + catch (const std::exception& e) { message = e.what(); } + catch (...) {} + + if (FILE* log = fopen("/tmp/metal-errors.log", "a")) { + fprintf(log, "%s: %s\n", where, message.c_str()); + fclose(log); + } + return [NSError errorWithDomain: @"MTLLibraryErrorDomain" code: 3 + userInfo: @{NSLocalizedDescriptionKey: [NSString stringWithUTF8String: message.c_str()]}]; +} diff --git a/include/Metal/MTLTrace.h b/include/Metal/MTLTrace.h new file mode 100644 --- /dev/null +++ b/include/Metal/MTLTrace.h @@ -0,0 +1,106 @@ +#pragma once +#include +#include +#include +#include +#include +#include +#include + +// All the Metal diagnostics (counters, render-pass descriptions, the magenta overrides, drawable readbacks) are +// off unless /tmp/metal-debug-trace exists when the process starts. +inline bool MTLDebugEnabled() +{ + static const bool enabled = access("/tmp/metal-debug-trace", F_OK) == 0; + return enabled; +} + +// Total draw calls seen so far, used to attribute draws to render passes. +inline std::atomic& MTLTraceDrawTotal() +{ + static std::atomic total { 0 }; + return total; +} + +// Diagnostic event counters, appended to the guest's /tmp/metal-trace.log every few seconds. +inline void MTLTraceCount(const char* name) +{ + static const bool countDraws = MTLDebugEnabled() || access("/tmp/metal-debug-timing", F_OK) == 0 || access("/tmp/metal-debug-snapshotdraws", F_OK) == 0; + if (countDraws && name[0] == 'd' && name[1] == 'r' && name[2] == 'a' && name[3] == 'w') { + ++MTLTraceDrawTotal(); + } + if (!MTLDebugEnabled()) { + return; + } + + static std::mutex lock; + static std::map counts; + static time_t last = 0; + + std::lock_guard guard(lock); + ++counts[name]; + + time_t now = time(nullptr); + if (now - last >= 3) { + last = now; + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "counts@%ld:", (long) now); + for (auto& count: counts) { + fprintf(log, " %s=%lu", count.first.c_str(), count.second); + } + fprintf(log, "\n"); + fclose(log); + } + } +} + +// Cheap aggregated timings (flag file /tmp/metal-debug-timing in the guest when the process starts), printed every 5 s. +inline void MTLTraceTiming(const char* what, double seconds) +{ + static const bool enabled = access("/tmp/metal-debug-timing", F_OK) == 0; + if (!enabled) { + return; + } + struct Stat { double sum = 0, max = 0; unsigned long count = 0; }; + static std::mutex lock; + static std::map stats; + static time_t last = 0; + std::lock_guard guard(lock); + Stat& stat = stats[what]; + stat.sum += seconds; + stat.count++; + if (seconds > stat.max) stat.max = seconds; + time_t now = time(nullptr); + if (now - last >= 5) { + last = now; + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "mtl-timing@%ld:", (long) now); + for (auto& entry: stats) { + fprintf(log, " %s[n=%lu avg=%.1fms max=%.1fms]", entry.first.c_str(), entry.second.count, entry.second.sum / entry.second.count * 1000, entry.second.max * 1000); + } + fprintf(log, "\n"); + fclose(log); + } + stats.clear(); + } +} + +inline bool MTLTimingEnabled() +{ + static const bool enabled = access("/tmp/metal-debug-timing", F_OK) == 0; + return enabled; +} + +inline double MTLTraceNow() +{ + struct timespec ts; + clock_gettime(CLOCK_MONOTONIC, &ts); + return ts.tv_sec + ts.tv_nsec / 1e9; +} + +// Draw-state tracking (cheap bookkeeping for the draw log): on with the trace flag, or with /tmp/metal-debug-draws. +inline bool MTLDrawTracking() +{ + static const bool enabled = MTLDebugEnabled() || access("/tmp/metal-debug-draws", F_OK) == 0; + return enabled; +} diff --git a/private-include/Metal/MTLArgumentInternal.h b/private-include/Metal/MTLArgumentInternal.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLArgumentInternal.h @@ -0,0 +1,15 @@ +#ifndef _METAL_MTLARGUMENTINTERNAL_H_ +#define _METAL_MTLARGUMENTINTERNAL_H_ +#import +#import +#if DARLING_METAL_ENABLED +#include +@interface MTLComputePipelineReflection (Internal) +- (instancetype)initWithFunction:(std::shared_ptr)function; +@end +@interface MTLRenderPipelineReflection (Internal) +- (instancetype)initWithVertexFunction:(std::shared_ptr)vertex fragmentFunction:(std::shared_ptr)fragment; +- (instancetype)initWithVertexFunction:(std::shared_ptr)vertex fragmentFunction:(std::shared_ptr)fragment vertexDescriptor:(MTLVertexDescriptor*)vertexDescriptor; +@end +#endif +#endif diff --git a/private-include/Metal/MTLBlitCommandEncoderInternal.h b/private-include/Metal/MTLBlitCommandEncoderInternal.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLBlitCommandEncoderInternal.h @@ -0,0 +1,15 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#ifndef _METAL_MTLBLITCOMMANDENCODERINTERNAL_H_ +#define _METAL_MTLBLITCOMMANDENCODERINTERNAL_H_ +#import +#if DARLING_METAL_ENABLED +#include +#endif +@interface MTLBlitCommandEncoderInternal : NSObject +#if DARLING_METAL_ENABLED +- (instancetype)initWithEncoder:(std::shared_ptr)encoder device:(id)device; +#endif +@end +#endif diff --git a/private-include/Metal/MTLDepthStencilInternal.h b/private-include/Metal/MTLDepthStencilInternal.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLDepthStencilInternal.h @@ -0,0 +1,27 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#ifndef _METAL_MTLDEPTHSTENCILINTERNAL_H_ +#define _METAL_MTLDEPTHSTENCILINTERNAL_H_ +#import +#if DARLING_METAL_ENABLED +#include +@interface MTLStencilDescriptor (Internal) +- (Indium::StencilDescriptor)asIndiumDescriptor; +@end +@interface MTLDepthStencilDescriptor (Internal) +- (Indium::DepthStencilDescriptor)asIndiumDescriptor; +@end +#endif +@interface MTLDepthStencilStateInternal : NSObject +#if DARLING_METAL_ENABLED +@property(nonatomic, readonly) std::shared_ptr state; +- (instancetype)initWithState:(std::shared_ptr)state device:(id)device label:(NSString*)label; +#endif +@end +#if DARLING_METAL_ENABLED +// Direct access for the encoders: the property getter is a message send that returns the shared_ptr by value. +const std::shared_ptr& MTLDepthStencilGetIndium(id object); +#endif + +#endif diff --git a/private-include/Metal/MTLDeviceInfo.h b/private-include/Metal/MTLDeviceInfo.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLDeviceInfo.h @@ -0,0 +1,166 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#ifndef _METAL_MTLDEVICEINFO_H_ +#define _METAL_MTLDEVICEINFO_H_ + +// Host-testable helpers for the values MTLDevice reports about the GPU. +// Plain C++ only (no Objective-C, Vulkan or Indium) so tools/test-metal-device-info.cpp +// can compile this header on the host. + +#include +#include +#include + +namespace MTLDeviceInfo { + +// Used when neither the cleaned nor the raw Vulkan name has any characters. +static constexpr const char* kFallbackName = "Darling Indium Device"; +// A real Mac never reports registry ID 0, and apps use it as a key for the device. +static constexpr uint64_t kDefaultRegistryID = 0x100000439ULL; + +static inline bool isSpace(char c) { + return c == ' ' || c == '\t' || c == '\n' || c == '\r' || c == '\f' || c == '\v'; +} + +static inline std::string collapseWhitespace(const std::string& text) { + std::string out; + bool pendingSpace = false; + for (char c: text) { + if (isSpace(c)) { + pendingSpace = !out.empty(); + continue; + } + if (pendingSpace) { + out.push_back(' '); + pendingSpace = false; + } + out.push_back(c); + } + return out; +} + +static inline char lower(char c) { + return (c >= 'A' && c <= 'Z') ? static_cast(c - 'A' + 'a') : c; +} + +// Removes "(R)" and "(TM)" in any letter case (Intel's Vulkan names use both). +static inline std::string removeTrademarks(const std::string& text) { + std::string out; + for (size_t i = 0; i < text.size();) { + if (text[i] == '(') { + if (i + 2 < text.size() && lower(text[i + 1]) == 'r' && text[i + 2] == ')') { + i += 3; + continue; + } + if (i + 3 < text.size() && lower(text[i + 1]) == 't' && lower(text[i + 2]) == 'm' && text[i + 3] == ')') { + i += 4; + continue; + } + } + out.push_back(text[i++]); + } + return out; +} + +// Drops one trailing parenthetical group (Mesa/driver annotation such as +// " (RADV NAVI21)", " (CML GT2)" or " (LLVM 22.1.0, 256 bits)"), but never when +// the group is the whole name. Expects whitespace already collapsed and trimmed. +static inline std::string removeTrailingGroup(const std::string& text) { + if (text.empty() || text.back() != ')') { + return text; + } + int depth = 0; + for (size_t i = text.size(); i-- > 0;) { + if (text[i] == ')') { + ++depth; + } else if (text[i] == '(' && --depth == 0) { + std::string prefix = collapseWhitespace(text.substr(0, i)); + return prefix.empty() ? text : prefix; + } + } + return text; // unbalanced: leave it alone +} + +// Converts a VkPhysicalDeviceProperties::deviceName into macOS Metal style, +// e.g. "AMD Radeon RX 6800 XT (RADV NAVI21)" -> "AMD Radeon RX 6800 XT" and +// "Intel(R) UHD Graphics 630 (CML GT2)" -> "Intel UHD Graphics 630". +// This only cleans the real name; it never substitutes another GPU. +static inline std::string metalNameFromVulkanName(const char* vulkanName) { + std::string raw = vulkanName ? vulkanName : ""; + std::string cleaned = removeTrailingGroup(collapseWhitespace(removeTrademarks(raw))); + if (!cleaned.empty()) { + return cleaned; + } + std::string rawTrimmed = collapseWhitespace(raw); + return rawTrimmed.empty() ? std::string(kFallbackName) : rawTrimmed; +} + +// DARLING_METAL_GPU_NAME (used verbatim when non-empty) wins over the Vulkan name. +// With DARLING_METAL_MAC_GPU_NAME=1: the GPU from the host's vendor that the MacBookPro16,1 +// iokitd reports shipped with, by Vulkan vendor ID. nullptr for other vendors (NVIDIA), which +// keep their own name so apps never take another vendor's driver paths. +static inline const char* macGpuName(uint32_t vendorID) { + switch (vendorID) { + case 0x1002: return "AMD Radeon Pro 5500M"; + case 0x8086: return "Intel UHD Graphics 630"; + default: return nullptr; + } +} + +// DARLING_METAL_GPU_NAME wins, then the Mac GPU name (when enabled and known), then the +// cleaned Vulkan name. +static inline std::string resolveName(const char* overrideName, const char* vulkanName, const char* macName = nullptr) { + if (overrideName && overrideName[0] != '\0') { + return overrideName; + } + if (macName && macName[0] != '\0') { + return macName; + } + return metalNameFromVulkanName(vulkanName); +} + +// Parses DARLING_METAL_VRAM_MB: a positive decimal integer of mebibytes, digits only. +// Returns 0 (meaning "ignore") for anything else, including overflow. +static inline uint64_t parseMegabytes(const char* text) { + if (!text || text[0] == '\0') { + return 0; + } + uint64_t megabytes = 0; + for (const char* p = text; *p; ++p) { + if (*p < '0' || *p > '9') { + return 0; + } + uint64_t digit = static_cast(*p - '0'); + if (megabytes > (UINT64_MAX / (1024ull * 1024ull) - digit) / 10) { + return 0; + } + megabytes = megabytes * 10 + digit; + } + return megabytes * 1024ull * 1024ull; +} + +struct Heap { + uint64_t size; + bool deviceLocal; +}; + +// Working-set size from the Vulkan memory heaps; 0 when nothing usable is reported. +// Discrete GPUs report their device-local (VRAM) heaps. GPUs sharing system RAM +// (integrated/CPU device types) report every heap: their device-local heap may be +// only the firmware carve-out (RADV APUs) while the GPU can also use the system heap, +// which matches macOS reporting a large share of system RAM for unified-memory GPUs. +static inline uint64_t workingSetBytes(const Heap* heaps, size_t count, bool sharesSystemMemory) { + uint64_t total = 0; + for (size_t i = 0; i < count; ++i) { + if (!sharesSystemMemory && !heaps[i].deviceLocal) { + continue; + } + total = heaps[i].size > UINT64_MAX - total ? UINT64_MAX : total + heaps[i].size; + } + return total; +} + +} // namespace MTLDeviceInfo + +#endif // _METAL_MTLDEVICEINFO_H_ diff --git a/private-include/Metal/MTLSamplerStateInternal.h b/private-include/Metal/MTLSamplerStateInternal.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLSamplerStateInternal.h @@ -0,0 +1,13 @@ +// SPDX-License-Identifier: MPL-2.0 +#pragma once +#import +#if DARLING_METAL_ENABLED +#include +@interface MTLSamplerStateInternal : NSObject +@property(nonatomic, readonly) std::shared_ptr sampler; +- (instancetype)initWithSampler:(std::shared_ptr)sampler device:(id)device label:(NSString*)label; +@end +// Direct access for the encoders: the property getter is a message send that returns the shared_ptr by value. +const std::shared_ptr& MTLSamplerGetIndium(id object); + +#endif diff --git a/private-include/Metal/MTLVertexDescriptorInternal.h b/private-include/Metal/MTLVertexDescriptorInternal.h new file mode 100644 --- /dev/null +++ b/private-include/Metal/MTLVertexDescriptorInternal.h @@ -0,0 +1,13 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#ifndef _METAL_MTLVERTEXDESCRIPTORINTERNAL_H_ +#define _METAL_MTLVERTEXDESCRIPTORINTERNAL_H_ +#import +#if DARLING_METAL_ENABLED +#include +@interface MTLVertexDescriptor (Internal) +- (Indium::VertexDescriptor)asIndiumDescriptor; +@end +#endif +#endif diff --git a/src/Metal/MTLArgument.mm b/src/Metal/MTLArgument.mm new file mode 100644 --- /dev/null +++ b/src/Metal/MTLArgument.mm @@ -0,0 +1,78 @@ +#import +@implementation MTLStructMember +@synthesize name=_name, offset=_offset; +- (instancetype)initWithName:(NSString*)name offset:(NSUInteger)offset { self=[super init]; if(self){_name=[name copy];_offset=offset;}return self; } +- (void)dealloc { [_name release];[super dealloc]; } +@end +@implementation MTLStructType +@synthesize members=_members; +- (instancetype)initWithMembers:(NSArray*)members { self=[super init];if(self)_members=[members copy];return self; } +- (MTLStructMember*)memberByName:(NSString*)name { for(MTLStructMember* member in _members) if([member.name isEqualToString:name])return member; return nil; } +- (void)dealloc { [_members release];[super dealloc]; } +@end +@implementation MTLArgument +@synthesize name=_name, type=_type, index=_index, bufferAlignment=_bufferAlignment, bufferDataSize=_bufferDataSize, bufferStructType=_bufferStructType; +- (BOOL)isActive { return YES; } +#if DARLING_METAL_ENABLED +- (instancetype)initWithInfo:(const Iridium::ArgumentInfo&)info { + self=[super init];if(self){ + _name=[[NSString alloc] initWithUTF8String:info.name.c_str()];_index=info.index; + _type=info.type==Iridium::BindingType::Buffer ? MTLArgumentTypeBuffer : (info.type==Iridium::BindingType::Texture ? MTLArgumentTypeTexture : MTLArgumentTypeSampler); + _bufferAlignment=info.alignment;_bufferDataSize=info.size; + if(!info.members.empty()){ + NSMutableArray* members=[NSMutableArray new]; + for(const auto& member:info.members){ MTLStructMember* object=[[MTLStructMember alloc] initWithName:[NSString stringWithUTF8String:member.name.c_str()] offset:member.offset];[members addObject:object];[object release]; } + _bufferStructType=[[MTLStructType alloc] initWithMembers:members];[members release]; + } + }return self; +} +#endif +- (void)dealloc { [_name release];[_bufferStructType release];[super dealloc]; } +@end +#if DARLING_METAL_ENABLED +static NSArray* argumentsForFunction(std::shared_ptr function) { + NSMutableArray* result=[NSMutableArray array]; + if(function)for(const auto& info:function->argumentInfo()){ + MTLArgument* argument=[[MTLArgument alloc] initWithInfo:info];[result addObject:argument];[argument release]; + } + return result; +} +#endif +@implementation MTLComputePipelineReflection +@synthesize arguments=_arguments; +#if DARLING_METAL_ENABLED +- (instancetype)initWithFunction:(std::shared_ptr)function { self=[super init];if(self)_arguments=[argumentsForFunction(function) copy];return self; } +#endif +- (void)dealloc { [_arguments release];[super dealloc]; } +@end +@implementation MTLRenderPipelineReflection +@synthesize vertexArguments=_vertexArguments, fragmentArguments=_fragmentArguments; +#if DARLING_METAL_ENABLED +- (instancetype)initWithVertexFunction:(std::shared_ptr)vertex fragmentFunction:(std::shared_ptr)fragment { + return [self initWithVertexFunction:vertex fragmentFunction:fragment vertexDescriptor:nil]; +} +// Metal lists the vertex buffers that feed [[stage_in]] among the vertex arguments, as buffers named "vertexBuffer." +// (the game derives which vertex streams to bind from them). +- (instancetype)initWithVertexFunction:(std::shared_ptr)vertex fragmentFunction:(std::shared_ptr)fragment vertexDescriptor:(MTLVertexDescriptor*)vertexDescriptor { + self=[super init];if(self){ + NSMutableArray* vertexArguments=[argumentsForFunction(vertex) mutableCopy]; + NSMutableIndexSet* seen=[NSMutableIndexSet indexSet]; + for(NSUInteger attribute=0;vertexDescriptor&&attribute<31;++attribute){ + MTLVertexAttributeDescriptor* description=vertexDescriptor.attributes[attribute]; + if(description.format==MTLVertexFormatInvalid||[seen containsIndex:description.bufferIndex])continue; + [seen addIndex:description.bufferIndex]; + Iridium::ArgumentInfo info; + info.type=Iridium::BindingType::Buffer; + info.index=description.bufferIndex; + info.name="vertexBuffer."+std::to_string(description.bufferIndex); + info.size=vertexDescriptor.layouts[description.bufferIndex].stride; + info.alignment=4; + MTLArgument* argument=[[MTLArgument alloc] initWithInfo:info];[vertexArguments addObject:argument];[argument release]; + } + _vertexArguments=vertexArguments; + _fragmentArguments=[argumentsForFunction(fragment) copy]; + }return self; +} +#endif +- (void)dealloc { [_vertexArguments release];[_fragmentArguments release];[super dealloc]; } +@end diff --git a/src/Metal/MTLBlitCommandEncoder.mm b/src/Metal/MTLBlitCommandEncoder.mm new file mode 100644 --- /dev/null +++ b/src/Metal/MTLBlitCommandEncoder.mm @@ -0,0 +1,101 @@ +// SPDX-FileCopyrightText: 2026 Darling Developers +// SPDX-License-Identifier: MPL-2.0 + +#import +#include +#import +#import +#import +#import +#import + +#if DARLING_METAL_ENABLED +// Diagnostic (/tmp/metal-debug-timing): are the texture uploads carrying data? Sample the start of each source. +static void MTLDebugUpload(id source, NSUInteger offset, id destination, NSUInteger level, MTLSize size) +{ + if (!MTLTimingEnabled()) { + return; + } + const uint8_t* bytes = (const uint8_t*) source.contents; + if (!bytes) { + MTLTraceTiming("upload_sourceNotCPUVisible", 0); + return; + } + NSUInteger available = source.length > offset ? source.length - offset : 0; + NSUInteger probe = std::min(available, 4096); + NSUInteger nonZero = 0; + for (NSUInteger i = 0; i < probe; ++i) if (bytes[offset + i]) ++nonZero; + char name[64]; + snprintf(name, sizeof(name), "upload_%s_f%lu", nonZero ? "data" : "ZERO", (unsigned long) destination.pixelFormat); + MTLTraceTiming(name, 0); + static unsigned long logged = 0; + if (!nonZero && destination.width >= 256 && ++logged <= 60) { + if (FILE* log = fopen("/tmp/metal-trace.log", "a")) { + fprintf(log, "zero-upload tex=%lux%lu fmt=%lu mips=%lu level=%lu size=%lux%lu srcOffset=%lu\n", (unsigned long) destination.width, (unsigned long) destination.height, (unsigned long) destination.pixelFormat, + (unsigned long) destination.mipmapLevelCount, (unsigned long) level, (unsigned long) size.width, (unsigned long) size.height, (unsigned long) offset); + fclose(log); + } + } +} +#endif +@implementation MTLBlitCommandEncoderInternal +#if DARLING_METAL_ENABLED +{ + std::shared_ptr _encoder; +} +@synthesize device = _device; +@synthesize label = _label; +- (instancetype)initWithEncoder:(std::shared_ptr)encoder device:(id)device { + self = [super init]; + if (self) { _encoder = encoder; _device = [device retain]; } + return self; +} +- (void)dealloc { + [_device release]; [_label release]; [super dealloc]; +} +- (void)endEncoding { _encoder->endEncoding(); } +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset size:(NSUInteger)size { + MTLTraceCount("blitBufferToBuffer"); + _encoder->copy(MTLBufferGetIndium(source), sourceOffset, MTLBufferGetIndium(destination), destinationOffset, size); +} +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset sourceBytesPerRow:(NSUInteger)sourceBytesPerRow sourceBytesPerImage:(NSUInteger)sourceBytesPerImage sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin { + MTLTraceCount("blitBufferToTexture"); + MTLDebugUpload(source, sourceOffset, destination, destinationLevel, sourceSize); + _encoder->copy(MTLBufferGetIndium(source), sourceOffset, sourceBytesPerRow, sourceBytesPerImage, MTLSizeToIndium(sourceSize), MTLTextureGetIndium(destination), destinationSlice, destinationLevel, (Indium::Origin {destinationOrigin.x, destinationOrigin.y, destinationOrigin.z})); +} +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset destinationBytesPerRow:(NSUInteger)destinationBytesPerRow destinationBytesPerImage:(NSUInteger)destinationBytesPerImage { + MTLTraceCount("blitTextureToBuffer"); + _encoder->copy(MTLTextureGetIndium(source), sourceSlice, sourceLevel, (Indium::Origin {sourceOrigin.x, sourceOrigin.y, sourceOrigin.z}), MTLSizeToIndium(sourceSize), MTLBufferGetIndium(destination), destinationOffset, destinationBytesPerRow, destinationBytesPerImage); +} +- (void)copyFromTexture:(id)source toTexture:(id)destination { + MTLTraceCount("blitTextureToTexture"); + _encoder->copy(MTLTextureGetIndium(source), MTLTextureGetIndium(destination)); +} +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel sliceCount:(NSUInteger)sliceCount levelCount:(NSUInteger)levelCount { + MTLTraceCount("blitTextureToTexture"); + _encoder->copy(MTLTextureGetIndium(source), sourceSlice, sourceLevel, MTLTextureGetIndium(destination), destinationSlice, destinationLevel, sliceCount, levelCount); +} +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin { + MTLTraceCount("blitTextureToTexture"); + _encoder->copy(MTLTextureGetIndium(source), sourceSlice, sourceLevel, (Indium::Origin {sourceOrigin.x, sourceOrigin.y, sourceOrigin.z}), MTLSizeToIndium(sourceSize), MTLTextureGetIndium(destination), destinationSlice, destinationLevel, (Indium::Origin {destinationOrigin.x, destinationOrigin.y, destinationOrigin.z})); +} +- (void)fillBuffer:(id)buffer range:(NSRange)range value:(uint8_t)value { + _encoder->fillBuffer(MTLBufferGetIndium(buffer), NSRangeToIndium(range), value); +} +- (void)generateMipmapsForTexture:(id)texture { + MTLTraceCount("generateMipmaps"); + _encoder->generateMipmapsForTexture(MTLTextureGetIndium(texture)); +} +- (void)copyFromBuffer:(id)source sourceOffset:(NSUInteger)sourceOffset sourceBytesPerRow:(NSUInteger)sourceBytesPerRow sourceBytesPerImage:(NSUInteger)sourceBytesPerImage sourceSize:(MTLSize)sourceSize toTexture:(id)destination destinationSlice:(NSUInteger)destinationSlice destinationLevel:(NSUInteger)destinationLevel destinationOrigin:(MTLOrigin)destinationOrigin options:(MTLBlitOption)options { + MTLTraceCount("blitBufferToTexture"); + _encoder->copy(MTLBufferGetIndium(source), sourceOffset, sourceBytesPerRow, sourceBytesPerImage, MTLSizeToIndium(sourceSize), MTLTextureGetIndium(destination), destinationSlice, destinationLevel, (Indium::Origin {destinationOrigin.x, destinationOrigin.y, destinationOrigin.z}), static_cast(options)); +} +- (void)copyFromTexture:(id)source sourceSlice:(NSUInteger)sourceSlice sourceLevel:(NSUInteger)sourceLevel sourceOrigin:(MTLOrigin)sourceOrigin sourceSize:(MTLSize)sourceSize toBuffer:(id)destination destinationOffset:(NSUInteger)destinationOffset destinationBytesPerRow:(NSUInteger)destinationBytesPerRow destinationBytesPerImage:(NSUInteger)destinationBytesPerImage options:(MTLBlitOption)options { + MTLTraceCount("blitTextureToBuffer"); + _encoder->copy(MTLTextureGetIndium(source), sourceSlice, sourceLevel, (Indium::Origin {sourceOrigin.x, sourceOrigin.y, sourceOrigin.z}), MTLSizeToIndium(sourceSize), MTLBufferGetIndium(destination), destinationOffset, destinationBytesPerRow, destinationBytesPerImage, static_cast(options)); +} +#else +@dynamic device, label; +MTL_UNSUPPORTED_CLASS +#endif +@end