Render advanced 3D graphics and perform data-parallel computations using graphics processors using Metal.

Posts under Metal tag

200 Posts

Post

Replies

Boosts

Views

Activity

coreai-build crashes for A-series Neural Engine targets since Xcode 27.2
Since Xcode 27.2, ahead-of-time compiling a Core AI model for an iPhone chip crashes as soon as any part of the model is placed on the Neural Engine: xcrun coreai-build compile Model.aimodel --platform iOS --min-deployment-version 27.0 \ --architecture h17p --preferred-compute neural-engine --output outcoreai-build exits with Segmentation fault: 11 (status 139), prints nothing and writes no output. The crash report points into the Neural Engine compiler: ANEHostCompiler ZinIrCompilerParameters::PropogateGlobalOptionsToPerProcedureOptions(...) ANEHostCompiler ANECompilerOffline::CompileAllTargetProcedures()What we found: It affects every A-series architecture (h16p, h17p, h18p, h19p). M-series targets (h13g through h17g) compile fine. It doesn't depend on the model. A model with a single convolution crashes too. It happens with Xcode 27.2 and 27.3 beta 1. Xcode 27.1 compiles the same models. Models that 27.2 places entirely on the GPU do compile. The Xcode 27.1 toolchain doesn't run on macOS 27.2. It fails at launch with dyld: Symbol not found: … AIModelAsset.delegatesFolderURL(artifactName:). So on macOS 27.2, we currently can't compile Core AI models for an iPhone's Neural Engine at all. The only working setup we have is Xcode 27.1 on macOS 27.0. Reported as FB25501691, with a minimal model and crash reports. Has anyone else run into this or found a workaround?
0
0
20
1h
Is Metal 4.1 `interleave` available on M4 Pro? Pipeline creation fails with "unlowered function call"
The Metal Shading Language Specification 4.1, section 6.4, lists interleave and deinterleave as "All OS: Metal 4.1 and later", with no GPU family restriction. The Metal Feature Set Tables list the M4 series as Apple9 and have no row for interleave. On an M4 Pro, a kernel that calls interleave compiles, but creating its compute pipeline fails. Which GPUs and macOS versions support interleave and deinterleave, and is the M4 family expected to? import Metal let src = """ #include <metal_stdlib> kernel void k(device const metal::uchar2 *a [[buffer(0)]], device metal::ushort *b [[buffer(1)]], uint i [[thread_position_in_grid]]) { b[i] = metal::interleave(a[i]); } """ let device = MTLCreateSystemDefaultDevice()! let options = MTLCompileOptions() options.languageVersion = .version4_1 let function = try! device.makeLibrary(source: src, options: options).makeFunction(name: "k")! do { _ = try device.makeComputePipelineState(function: function) print(device.name, "| pipeline created") } catch { print(device.name, "| pipeline failed:", error.localizedDescription) }The library compiles, but creating the pipeline fails. ushort2/uint, uint2/ulong and deinterleave fail the same way. Apple M4 Pro | pipeline failed: Encountered unlowered function call to air.interleave.i16. Verify device or GPU architecture supports this featureIf I replace the kernel line b[i] = metal::interleave(a[i]); with b[i] = metal::ushort(a[i].x) | (metal::ushort(a[i].y) << 8);, it prints Apple M4 Pro | pipeline created. Configuration Apple M4 Pro, macOS 27.2.1 (26B131), Xcode 27.1 (27A9275), Swift 6.4, metal 32023.921. Software Update shows nothing newer.
1
0
101
3h
xctrace gpu-counter-value: what do sample-index and timestamp mean?
I export GPU counters with xcrun xctrace export --xpath '/trace-toc/run[@number="1"]/data/table[@schema="gpu-counter-value"]' (Metal System Trace + Metal GPU Counters, compute workload) and need to confirm two columns. On an M5 Max (macOS 27.0.1, Xcode 26.5, counter profile "Performance Limiters", 88 counters) the rows come in three groups with different periods: 74 counters every 20.2 µs, 6 texture counters every 30.9 µs, 8 bandwidth/MMU/LLC counters every 50.0 µs. Each group has its own sample-index sequence; the indices of the groups drift apart by thousands within seconds. On an M1 Pro (macOS 26.7.1, Xcode 26.4.1, 31 counters) every period appears twice, with ring-buffer-index 0 and 1, the second copy timestamped ~35 µs later. Is timestamp the sampling time of the row, and does value describe the interval since the previous sample of that counter or an instantaneous reading? Is sample-index a per-group sequence number, or a key meant to join all counters of one sampling period? How should counters from groups with different periods be combined for one time window? On M1 Pro, are ring-buffer-index 0 and 1 two readings of the same period (use one) or two independent samples? Thanks in advance! Best, Lin
0
0
85
1d
Metal 4: MTL4Compiler crashes when many threads capture binaries into one MTL4PipelineDataSetSerializer
Hey all, hoping someone on the Metal team can take a look at this one. It's a crash inside Metal 4's binary capture, and as far as I can tell there's no way to fix it from the app side. Filed as FB25478288 (partial repro is link is there too). Why I think it matters The setup that crashes is basically the recommended Metal 4 workflow: one MTL4Compiler with a MTL4PipelineDataSetSerializer (CaptureBinaries) attached, called from maximumConcurrentCompilationTaskCount threads, to harvest pipelines for an offline archive. Lots of pipelines share a vertex function, which I'd guess is true for most real renderers. The crash is 100% inside Metal.framework, with no app code on the crashing thread. It's easy to hit: 3 out of 3 harvest runs crashed for us, one of them 2 s in. The serializer is attached to the compiler implicitly, so the app can't put a lock around whatever it does internally. MTL4Compiler is Sendable, and the docs say it handles multithreading by default, so I'd expect capture to be thread-safe too. Environment MacBook Pro M2 Max, macOS 27.2, Xcode 27.2, Metal toolchain 32023.922.4 Also reported on an M3 Pro ~33k render pipelines over ~16k vertex functions, up to ~1,000 pipelines in a row sharing one vertex function MTL4RenderPipelineDescriptor with MTL4LibraryFunctionDescriptor stages (sometimes wrapped in MTL4SpecializedFunctionDescriptor), all colour attachments Unspecialized The serializer gets written out with serializeAsArchiveAndFlushToURL: every so often, then we start a fresh compiler + serializer Crash EXC_BAD_ACCESS (SIGSEGV), "possible pointer authentication failure": objc_msgSend MultiLevelAirntBinaryCache::getElementAtDestination(MTLBinaryKey*, ...) -[MTLCompiler compileFunctionRequestInternal:frameworkLinking:linkDataSize:reflectionOnly:compilerTask:completionHandler:] ... AGX::Compiler::compileProgram<AGX::VertexProgramKey, ...>It looks like a use-after-free: two compiles capturing the same vertex program at the same time, and an entry in the destination archive gets freed while the other one is still using it. It only happens with capture on, and it goes away if we never let two capture compiles that share a stage function run at once. Workaround we ship (not great) A lock per stage function around each capture compile, plus sending runs of pipelines that share a vertex shader to the same thread. That stopped the crashes, but we had to put a timeout on the lock, because the compiler occasionally never returns, so in theory it can still race. Second, related crash If serializeAsArchiveAndFlushToURL: runs while other threads are still compiling into that serializer, we get a different crash (null deref at 0x30) in dispatch_data_create_map ← -[_MTLBinaryArchive newArchiverIdWithBinaryKey:entry:srcArchiverId:functionType:]. Our standalone repro hits this one reliably. The getElementAtDestination crash only shows up with our real workload so far. Repro + details https://github.com/Aeonitis/metal4-capture-race It's a small Objective-C repro (Metal + Foundation only) with a bunch of modes, the results so far, and a comparison with how Vulkan (VkPipelineCache) and D3D12 spell out the thread-safety rules for this same pattern. Questions Is capturing from multiple threads into one MTL4PipelineDataSetSerializer meant to be supported? If not, could the docs say so? Is one compiler + one serializer per thread a safe workaround, or is there shared state between compilers in the same process that would still race? Is calling serializeAsArchiveAndFlushToURL: while compiles are still in flight on that serializer allowed? Is this already fixed in a seed
0
0
306
3d
Request for PROLED display to support Adaptive Sync (VRR)
I was hoping with an OLED display that the new MBP might support true Adaptive Sync aka VRR. Alas, it only gets ProMotion, with the same huge gap between supported refresh rates at the top end (60hz and 120hz). Hooking up an HDMI VRR display such as any LG TV results in a display that works much better with variable framerates than the built-in display, since you can deliver frames at any interval between the min and max. VRR greatly improves games and already works great in macOS with an external display, so I humbly request that Apple consider another option next to ProMotion, for Adaptive Sync/VRR.
0
0
206
4d
Xcode 27.0 Metal Toolchain download stuck at "Beginning asset download..."
Hi everyone, I downloaded Xcode 27.0 today because I need the Metal Toolchain for Unreal Engine 5.6 on my Mac. Xcode itself is installed and working, but Unreal is having issues because the Metal Toolchain doesn't seem to be installed. When I run: xcrun metal --version I get: error: error: cannot execute tool 'metal' due to missing Metal Toolchain; use: xcodebuild -downloadComponent MetalToolchain I tried: sudo xcodebuild -downloadComponent MetalToolchain but it just stays on: Beginning asset download... If I cancel it, I get: xcodebuild: error: Download was cancelled. (Asset download for com.apple.MobileAsset.MetalToolchain ...) I've also run: xcodebuild -runFirstLaunch and installed the latest Command Line Tools for Xcode 27.0 through Software Update. xcrun --find metal does find the executable, but the actual Metal Toolchain appears to be missing. I'm mainly trying to get this working so I can use Unreal Engine 5.6. Has anyone had this issue with Xcode 27, or know the correct way to install the Metal Toolchain when xcodebuild -downloadComponent gets stuck at "Beginning asset download"? Thanks!
1
1
501
6d
D3DMetal: present hangs permanently when the game window becomes occluded
Environment macOS 27.0.1 (26A434), Apple M4 Pro Apple Game Porting Toolkit 4.0b2 (D3DMetal.framework) via CrossOver 27 D3D12 game, borderless fullscreen window (2540x1650) Symptom When the game window becomes occluded — Mission Control, a system OSD (volume/brightness), or another window being brought to the front — the game never renders another frame. It does not crash: GPU utilization drops to 0%, the process stays alive with near-zero CPU, and it never recovers. Killing and restarting the game is the only way out. D3DMetal log at the moment it happens: [D3DMetal:FATAL:%llX][%s:%d] Failed to get drawable for HWND. Run with D3DM_NO_WINDOW=1 env. Question When CAMetalLayer/nextDrawable has no drawable (window not visible), is there a supported way for D3DMetal to skip frames instead of blocking forever? Is there an environment variable or configuration that makes D3DMetal treat "no drawable" as a recoverable condition? Does D3DM_NO_WINDOW=1 apply here, or is it only for headless/offscreen use? I understand apps should handle a nil drawable and continue, but this path is inside D3DMetal. Any guidance would help.
3
0
1.2k
6d
iOS app running on macOS 27.0.1 terminated on suspension due to a Metal archiveUsage.db file lock (0xDEAD10CC)
Hello, We are investigating an issue where our iOS app, running as a “Designed for iPad” app on an Apple Silicon Mac, is terminated when it is suspended in the background. Environment OS: macOS 27.0.1 Hardware: Apple Silicon Mac App type: An iOS/iPadOS app running on Mac, rather than a separately developed macOS app The app uses Core Image’s CIContext to generate some image effects Steps to reproduce Launch the app on Mac and start playing a podcast. Move the app to the background. Start playback on another device. Our app’s cross-device playback coordination pauses playback on the Mac. After a short period, the Mac app is terminated. We have also observed similar termination after visiting a screen that uses Core Image image effects. Actual result The debugger reports: Terminated due to signal 9 RunningBoard logs indicate that the app was terminated with code 0xDEAD10CC because it was suspended while holding a disallowed shared file lock. The file identified in the log is: /var/folders/.../C/app.podcast.cosmos/app.podcast.cosmos/com.apple.metal/archiveUsage.db/lock.mdb The relevant log message includes: suspended while holding a shared file lock Investigation so far Using breakpoints on open and openat, we captured a call stack corresponding to: App creates a CIContext → Core Image CI::MetalContext → Metal recordBinaryArchiveUsage → archiveUsage.db/lock.mdb is opened We also tried: CIContext(options: [.useSoftwareRenderer: true]) However, in our test environment, we still observed Metal initialization and access to the relevant cache files. This option did not avoid the issue. For the gradient-mask generation code we had identified, we replaced the Core Image implementation with a CPU implementation using CGContext and CGGradient, avoiding CIContext creation on that path. After repeatedly testing the same reproduction steps, we could no longer reproduce the termination in that scenario. This suggests a connection to the Metal cache file lock accessed during Core Image initialization. However, we have not established the exact lifetime or internal owner of the lock, and we have not confirmed a thread deadlock. Questions Is this a known issue involving Core Image/Metal and suspension of iOS apps running on macOS 27.0.1? Why would a system-managed Metal cache file lock cause the app to be terminated with 0xDEAD10CC when suspended? Is there a recommended context configuration, resource cleanup procedure, or lifecycle handling approach that allows us to continue using Core Image safely in this environment?
1
0
76
1w
Xcode 27 - Device Hub - backboardd crashes when displaying multiple apps under iPadOS 27
I'm getting persistent backboardd crashes in the iPadOS 27 simulator when my app and another app (Safari, Photos or Calendar, etc.) has windows visible on the screen at the same time, including slide over, etc. Steps to reproduce: Launch my app in an iPadOS 27 simulator. Open Safari or Photos in another window. Arrange both app windows so they're visible. Within a second backboardd crashes with EXC_CRASH (SIGABRT) in a com.apple.coreanimation.display.primary thread. The simulator log reports… -[MTLTextureDescriptorInternal validateWithDevice:]: error 'Texture Descriptor Validation MTLTextureDescriptor has invalid pixelFormat (0).' Xcode subsequently reports connection-interrupted messages and ends the app's debug session with: “Debug session ended with code 9: Terminated due to signal 9.” This works correctly in iPadOS 26.x simulators and on a physical device running iPadOS 27. I can reproduce this with both Xcode 27 and 27.1. Resetting the simulator and testing different simulated iPad models does not resolve it. Is anyone else seeing anything like this? It's presumably something I'm doing wrong, so I've double checked my UIScene handling and all seems well (it's been in there for a while) so I'm somewhat at a loss as to where to look next. This is on a Mac Mini M4 Pro running macOS 27.0.1 (26A434).
1
0
264
1w
Metal Compositor Service & Persona (VisionOS)
Hello, I'm currently trying to make a collaborative app. But it just works only on Reality View, when I tried to use Compositor Layer like below, the personas disappeared. ImmersiveSpace(id: "ImmersiveSpace-Metal") { CompositorLayer(configuration: MetalLayerConfiguration()) { layerRenderer in SpatialRenderer_InitAndRun(layerRenderer) } } Is there any potential solution too see Personas in Metal view? Thanks in advance!
3
0
2k
3w
Photogrammetry with Object masks hangs and terminates with masks of objects
[PhotogrammetrySample]) with objectMask set and traps on my ios26.5.2, see the attached screenshot on feedback FB24379913 , it gets to the function and hangs . Even the folder reconstruction with lazy sequence as recommended from your video sample also doesn;t complete as it can't find alignment and displays the CoreOC.PhotogrammetrySession.Error error 6 means alignment failed. What can be done or do you guys expose any functions that can be used to check or trace or handle these internally The ObjectMasks are actually segmentation masks from an segmentation algorithm I will appreciate a timely response and willing to provide more clarity and informations, thank you so much for your understanding
3
0
1.5k
3w
Clarifying Fence Behavior in Metal 4
I have some clarifying questions on how fences work in Metal 4. If two encoders A, B update the same fence and encoder C waits on that fence, is the work encoded by C guaranteed to execute after both A and B have completed, or after either A or B have completed? The following quotes in the waitForFence() documentation seem contradictory. "Encodes a command that instructs the GPU to pause before starting one or more stages of the pass until a pass updates a fence." "When encoding a pass that reuses a fence, wait for other passes to update the fence before repurposing that fence..." Are fences unsignaled when they are waited on? Specifically, can two (or more) encoders wait on a fence that is only updated in a single prior encoder? If an encoder updates a fence, no subsequent encoders in the command queue wait on the fence, and the command queue is committed, is that fence still signaled when used in a subsequent command buffer?
1
0
2.1k
Oct ’26
On-screen RealityView starves CADisplayLink to 30 Hz on ProMotion (Mac Catalyst)
FB24536235 On Mac Catalyst under macOS 27, a plain CADisplayLink asking for CAFrameRateRange(minimum: 60, maximum: 60, preferred: 60) gets serviced at 30 Hz for as long as a RealityView is on screen in the same window. The link does nothing per tick but count, so there's nothing of mine to blame it on. RealityKit's own statistics overlay reads 60.41 fps in the same frame. Click a segmented control that removes the RealityView and the same link goes straight back to 60. Nothing else changes. That's the whole reproducer, and I've attached it to the radar. It only happens while the display panel is in ProMotion mode. Set the built-in to a fixed 60 Hz and it's correct again. With an external 60 Hz display attached the roles swap: the built-in is fine and the external drops to somewhere between 18 and 30, and setting the built-in to 60 Hz fixes that one too without touching the external's own settings. A raw MTKView presenting continuously at 60, at 120, and on a 120 Hz link presenting every second callback are all fine, so it isn't continuous presentation and it isn't the requested rate. It's RealityKit specifically. Worth knowing if you're testing: RealityView on Catalyst is an ARView underneath, so both paths give you the same answer. This is VERY rough for anything that puts RealityKit next to a UI. In an editor that's the sidebar, the inspector, gizmos, drag handles, every display-link-driven or UIKit animation in the window running at half rate around a viewport that stays smooth. Likely Related to FB24091347, which is the same defect seen as SwiftUI scroll judder. If you can reproduce either, please file a duplicate. Attached two screenshots; first with promotion enabled, second with promotion off. PLEASE fix this, it drives me crazy and there seems to be no workaround. On release day of macOS 27 our app will likely be blamed for it by users and my hands are tied. Thank you!
2
1
1.4k
Oct ’26
Contribution route for native Metal point-cloud operators used with PyTorch MPS
I maintain mps-pointops, an open-source library implementing farthest-point sampling (FPS), k-nearest-neighbor search, and radius / Ball Query operations with custom Metal kernels for PyTorch MPS tensors. The intended use is 3D point-cloud pipelines whose existing extensions depend on CUDA. I would like to prepare a contribution proposal at the appropriate layer. Apple’s PyTorch/MPS page links to the PyTorch MPS backend contribution guide and directs backend issues to PyTorch’s GitHub tracker with the module: mps label. The current implementation has numerical-contract tests for selection order, ties, radius boundaries, and padding, plus benchmark records from physical M1 and M5 Pro Macs. The repository contains the kernels, Python interfaces, tests, and measurement conditions: https://github.com/gamzerA/mps-pointops My questions are: For specialized operators such as FPS and radius search, is the recommended contribution structure to keep the operator/API in a domain library (for example, a point-cloud or graph library), while proposing only reusable backend primitives to PyTorch MPS? Is there a public Apple-maintained contribution or technical-review channel relevant to this work, beyond the PyTorch route linked from Apple’s page? For an external library that dispatches custom Metal work alongside PyTorch MPS operations, which documented extension/interoperability path should be the long-term target for command-stream ordering and buffer lifetime? I would like to minimize reliance on backend internals when preparing an upstream proposal. Pointers to the appropriate public interface, maintainers, or contribution requirements would help me scope a small, reviewable proposal. I understand that inclusion in an upstream project is subject to its maintainers’ review. Apple documentation referenced: https://developer.apple.com/metal/pytorch/
0
0
208
Oct ’26
Is there a way to directly go from VideoToolbox to Metal for 10-bit/BT.2020 YCbCr HEVC?
tl;dr how can I get raw YUV in a Metal fragment shader from a VideoToolbox 10-bit/BT.2020 HEVC stream without any extra/secret format conversions? With VideoToolbox and 10-bit HEVC, I've found that it defaults to CVPixelBuffers w/ formats kCVPixelFormatType_Lossless_420YpCbCr10PackedBiPlanarFullRange or kCVPixelFormatType_Lossy_420YpCbCr10PackedBiPlanarFullRange. To mitigate this, I have the following snippet of code to my application: // We need our pixels unpacked for 10-bit so that the Metal textures actually work var pixelFormat:OSType? = nil let bpc = getBpcForVideoFormat(videoFormat!) let isFullRange = getIsFullRangeForVideoFormat(videoFormat!) // TODO: figure out how to check for 422/444, CVImageBufferChromaLocationBottomField? if bpc == 10 { pixelFormat = isFullRange ? kCVPixelFormatType_420YpCbCr10BiPlanarFullRange : kCVPixelFormatType_420YpCbCr10BiPlanarVideoRange } let videoDecoderSpecification:[NSString: AnyObject] = [kVTVideoDecoderSpecification_EnableHardwareAcceleratedVideoDecoder:kCFBooleanTrue] var destinationImageBufferAttributes:[NSString: AnyObject] = [kCVPixelBufferMetalCompatibilityKey: true as NSNumber, kCVPixelBufferPoolMinimumBufferCountKey: 3 as NSNumber] if pixelFormat != nil { destinationImageBufferAttributes[kCVPixelBufferPixelFormatTypeKey] = pixelFormat! as NSNumber } var decompressionSession:VTDecompressionSession? = nil err = VTDecompressionSessionCreate(allocator: nil, formatDescription: videoFormat!, decoderSpecification: videoDecoderSpecification as CFDictionary, imageBufferAttributes: destinationImageBufferAttributes as CFDictionary, outputCallback: nil, decompressionSessionOut: &decompressionSession) In short, I need kCVPixelFormatType_420YpCbCr10BiPlanar so that I have a straightforward MTLPixelFormat.r16Unorm/MTLPixelFormat.rg16Unorm texture binding for Y/CbCr. Metal, seemingly, has no direct pixel format for 420YpCbCr10PackedBiPlanar. I'd also rather not use any color conversion in VideoToolbox, in order to save on processing (and to ensure that the color transforms/transfer characteristics match between streamer/client, since I also have a custom transfer characteristic to mitigate blocking in dark scenes). However, I noticed that in visionOS 2, the CVPixelBuffer I receive is no longer a compressed render target (likely a bug), which caused GPU texture read bandwidth to skyrocket from 2GiB/s to 30GiB/s. More importantly, this implies that VideoToolbox may in fact be doing an extra color conversion step, wasting memory bandwidth. Does Metal actually have no way to handle 420YpCbCr10PackedBiPlanar? Are there any examples for reading 10-bit HDR HEVC buffers directly with Metal?
3
0
2.2k
Sep ’26
macOS 27 beta: ProMotion refresh cadence is unstable, causing constant scroll judder
FB24091347 On macOS 27.0 beta (26A5388g), MacBook Pro M4 Pro, the built-in ProMotion display never settles on a stable refresh cadence. Scrolling in SwiftUI judders constantly. The same app binary was smooth on macOS 26, and is smooth on a 120 Hz ProMotion iPad. I captured two 60-second Instruments traces — same app, same scene, same scrolling, no external display — changing only the display's refresh-rate setting. On ProMotion the vsync interval standard deviation is 4.093 ms across six different cadences, mostly flip-flopping between 120 Hz and 60 Hz. Forced to a fixed 60 Hz it drops to 0.391 ms with a single cadence. The app presented an identical 59 fps median in both runs — frame production is perfectly steady, the display just holds each frame for an unpredictable length of time. That's what makes this nasty: it's invisible to every frame-rate metric, so it looks like the app got slow when nothing about the app changed. I spent most of a day profiling my own code before realising the app was never the problem. Workaround: force the built-in display to 60 Hz. Worth noting, because it complicates the picture: attaching a 60 Hz Studio Display makes the built-in smooth, but the Studio itself then judders — despite its own vsync cadence measuring perfectly stable. So refresh rate alone isn't the whole story, and there may be a second mechanism. The clean, reproducible, single-variable result is the ProMotion vs forced-60 Hz comparison on the built-in panel. If you can reproduce this on an M-series MacBook Pro on 27 beta, please file a duplicate referencing FB24091347.
6
2
2.7k
Sep ’26
Metal FP32 arithmetic rounding-mode and denormal controls for deterministic shaders
Metal already provides a strong precise floating-point contract. With fast math disabled (mathMode = .safe and mathFloatingPointFunctions = .precise, or -fno-fast-math), the Metal Shading Language specification requires correctly rounded FP32 add, subtract, multiply, reciprocal, divide, sqrt, rsqrt, and fma. I am looking for clarification and, if necessary, API support for the two remaining pieces needed for portable bit-exact numerical shaders: Arithmetic rounding mode MSL §8.2 says either round-to-nearest-ties-to-even or round-toward-zero may be supported for floating-point operations. I cannot find a way to select or query the arithmetic rounding mode. The newer MTLCompileOptions.floatingPointConversionRoundingMode appears to apply only to narrowing float-to-float conversions, not arithmetic operations. Do all currently supported Apple GPU families use round-to-nearest-ties-to-even for precise FP32 add/subtract/multiply/divide/sqrt? If so, could that be made a documented guarantee? Otherwise, could Metal expose an arithmetic rounding-mode compile option and a corresponding MTLDevice capability query? Denormal behavior MSL §8.1 and §8.5 permit denormalized FP32 operands and results to be flushed to zero, including with fast math disabled. I cannot find a control or capability query for preserving denormal inputs and results. Do current Apple GPU families support denormal-preserving FP32 arithmetic? Could Metal expose a preserve/flush mode and a MTLDevice query? A convenient end state would be a queryable strict FP32 configuration combining: safe math; precise FP32 functions; contraction disabled when separate rounding points are required; round-to-nearest-ties-to-even arithmetic; preserved FP32 denormal inputs and results; defined signed-zero, infinity, and NaN behavior. My use case is deterministic GPU numerical simulation. A small compute-shader probe can identify effective rounding and denormal behavior on one GPU/OS/compiler combination, but it cannot provide the portable or future-proof contract needed by applications and higher-level APIs such as WebGPU. Related cross-API work: SPIR-V/Vulkan: https://github.com/KhronosGroup/SPIRV-Registry/issues/448 HLSL/DXIL/D3D12: https://github.com/microsoft/hlsl-specs/issues/926 WebGPU/WGSL umbrella issue: https://github.com/gpuweb/gpuweb/issues/2259 Relevant Metal documentation: Metal Shading Language Specification, §§1.6.3 and 8.1–8.5: https://developer.apple.com/metal/Metal-Shading-Language-Specification.pdf MTLCompileOptions.mathMode: https://developer.apple.com/documentation/metal/mtlcompileoptions/mathmode MTLCompileOptions.floatingPointConversionRoundingMode: https://developer.apple.com/documentation/metal/mtlcompileoptions/floatingpointconversionroundingmode
2
0
3.2k
Sep ’26
Data corruption when using MTLBlitCommandEncoder.copy from buffer to texture
When we use MTLBlitCommandEncoder.copy to copy from buffer to textures with specific size, we find that the texture content is corruppt. The specific rules we found are: On Apple GPUs (reproduced on Apple M5, macOS 26.3 and iPhone 13 / A15), for ASTC textures uploaded to private storage via the blit encoder, the sampler / texel-fetch mis-addresses the tail of a mip level when ALL of the following hold: that level's texel width is an exact multiple of the 16 KB page width (32 blocks) — 128 texels for ASTC 4x4, 192 for 6x6, 256 for 8x8; the texture width is not divisible by 2^level (a partial block column exists in the chain); that level is taller than one page (>32 block rows). The level's first 128 texel rows (first page row, 4x4) are always read correctly; everything beyond is mis-addressed. Reads landing on unmapped/invalid memory decode as opaque magenta (1,0,1); reads aliasing valid memory show wrong image content. This produced visible purple artifacts in a shipping game (impostor tree atlas, 514x1024 ASTC 4x4, mip2 bottom half). The minimum reproduce code is: ASTCMipTailBugRepro.swift The output is: 514x1024 blit(private) mip2: CORRUPTED (bottom of mip2 mis-addressed) 512x1024 blit(private) mip2: INTACT 514x1024 replaceRegion mip2: INTACT
1
0
1.6k
Sep ’26
coreai-build crashes for A-series Neural Engine targets since Xcode 27.2
Since Xcode 27.2, ahead-of-time compiling a Core AI model for an iPhone chip crashes as soon as any part of the model is placed on the Neural Engine: xcrun coreai-build compile Model.aimodel --platform iOS --min-deployment-version 27.0 \ --architecture h17p --preferred-compute neural-engine --output outcoreai-build exits with Segmentation fault: 11 (status 139), prints nothing and writes no output. The crash report points into the Neural Engine compiler: ANEHostCompiler ZinIrCompilerParameters::PropogateGlobalOptionsToPerProcedureOptions(...) ANEHostCompiler ANECompilerOffline::CompileAllTargetProcedures()What we found: It affects every A-series architecture (h16p, h17p, h18p, h19p). M-series targets (h13g through h17g) compile fine. It doesn't depend on the model. A model with a single convolution crashes too. It happens with Xcode 27.2 and 27.3 beta 1. Xcode 27.1 compiles the same models. Models that 27.2 places entirely on the GPU do compile. The Xcode 27.1 toolchain doesn't run on macOS 27.2. It fails at launch with dyld: Symbol not found: … AIModelAsset.delegatesFolderURL(artifactName:). So on macOS 27.2, we currently can't compile Core AI models for an iPhone's Neural Engine at all. The only working setup we have is Xcode 27.1 on macOS 27.0. Reported as FB25501691, with a minimal model and crash reports. Has anyone else run into this or found a workaround?
Replies
0
Boosts
0
Views
20
Activity
1h
Is Metal 4.1 `interleave` available on M4 Pro? Pipeline creation fails with "unlowered function call"
The Metal Shading Language Specification 4.1, section 6.4, lists interleave and deinterleave as "All OS: Metal 4.1 and later", with no GPU family restriction. The Metal Feature Set Tables list the M4 series as Apple9 and have no row for interleave. On an M4 Pro, a kernel that calls interleave compiles, but creating its compute pipeline fails. Which GPUs and macOS versions support interleave and deinterleave, and is the M4 family expected to? import Metal let src = """ #include <metal_stdlib> kernel void k(device const metal::uchar2 *a [[buffer(0)]], device metal::ushort *b [[buffer(1)]], uint i [[thread_position_in_grid]]) { b[i] = metal::interleave(a[i]); } """ let device = MTLCreateSystemDefaultDevice()! let options = MTLCompileOptions() options.languageVersion = .version4_1 let function = try! device.makeLibrary(source: src, options: options).makeFunction(name: "k")! do { _ = try device.makeComputePipelineState(function: function) print(device.name, "| pipeline created") } catch { print(device.name, "| pipeline failed:", error.localizedDescription) }The library compiles, but creating the pipeline fails. ushort2/uint, uint2/ulong and deinterleave fail the same way. Apple M4 Pro | pipeline failed: Encountered unlowered function call to air.interleave.i16. Verify device or GPU architecture supports this featureIf I replace the kernel line b[i] = metal::interleave(a[i]); with b[i] = metal::ushort(a[i].x) | (metal::ushort(a[i].y) << 8);, it prints Apple M4 Pro | pipeline created. Configuration Apple M4 Pro, macOS 27.2.1 (26B131), Xcode 27.1 (27A9275), Swift 6.4, metal 32023.921. Software Update shows nothing newer.
Replies
1
Boosts
0
Views
101
Activity
3h
xctrace gpu-counter-value: what do sample-index and timestamp mean?
I export GPU counters with xcrun xctrace export --xpath '/trace-toc/run[@number="1"]/data/table[@schema="gpu-counter-value"]' (Metal System Trace + Metal GPU Counters, compute workload) and need to confirm two columns. On an M5 Max (macOS 27.0.1, Xcode 26.5, counter profile "Performance Limiters", 88 counters) the rows come in three groups with different periods: 74 counters every 20.2 µs, 6 texture counters every 30.9 µs, 8 bandwidth/MMU/LLC counters every 50.0 µs. Each group has its own sample-index sequence; the indices of the groups drift apart by thousands within seconds. On an M1 Pro (macOS 26.7.1, Xcode 26.4.1, 31 counters) every period appears twice, with ring-buffer-index 0 and 1, the second copy timestamped ~35 µs later. Is timestamp the sampling time of the row, and does value describe the interval since the previous sample of that counter or an instantaneous reading? Is sample-index a per-group sequence number, or a key meant to join all counters of one sampling period? How should counters from groups with different periods be combined for one time window? On M1 Pro, are ring-buffer-index 0 and 1 two readings of the same period (use one) or two independent samples? Thanks in advance! Best, Lin
Replies
0
Boosts
0
Views
85
Activity
1d
Metal 4: MTL4Compiler crashes when many threads capture binaries into one MTL4PipelineDataSetSerializer
Hey all, hoping someone on the Metal team can take a look at this one. It's a crash inside Metal 4's binary capture, and as far as I can tell there's no way to fix it from the app side. Filed as FB25478288 (partial repro is link is there too). Why I think it matters The setup that crashes is basically the recommended Metal 4 workflow: one MTL4Compiler with a MTL4PipelineDataSetSerializer (CaptureBinaries) attached, called from maximumConcurrentCompilationTaskCount threads, to harvest pipelines for an offline archive. Lots of pipelines share a vertex function, which I'd guess is true for most real renderers. The crash is 100% inside Metal.framework, with no app code on the crashing thread. It's easy to hit: 3 out of 3 harvest runs crashed for us, one of them 2 s in. The serializer is attached to the compiler implicitly, so the app can't put a lock around whatever it does internally. MTL4Compiler is Sendable, and the docs say it handles multithreading by default, so I'd expect capture to be thread-safe too. Environment MacBook Pro M2 Max, macOS 27.2, Xcode 27.2, Metal toolchain 32023.922.4 Also reported on an M3 Pro ~33k render pipelines over ~16k vertex functions, up to ~1,000 pipelines in a row sharing one vertex function MTL4RenderPipelineDescriptor with MTL4LibraryFunctionDescriptor stages (sometimes wrapped in MTL4SpecializedFunctionDescriptor), all colour attachments Unspecialized The serializer gets written out with serializeAsArchiveAndFlushToURL: every so often, then we start a fresh compiler + serializer Crash EXC_BAD_ACCESS (SIGSEGV), "possible pointer authentication failure": objc_msgSend MultiLevelAirntBinaryCache::getElementAtDestination(MTLBinaryKey*, ...) -[MTLCompiler compileFunctionRequestInternal:frameworkLinking:linkDataSize:reflectionOnly:compilerTask:completionHandler:] ... AGX::Compiler::compileProgram<AGX::VertexProgramKey, ...>It looks like a use-after-free: two compiles capturing the same vertex program at the same time, and an entry in the destination archive gets freed while the other one is still using it. It only happens with capture on, and it goes away if we never let two capture compiles that share a stage function run at once. Workaround we ship (not great) A lock per stage function around each capture compile, plus sending runs of pipelines that share a vertex shader to the same thread. That stopped the crashes, but we had to put a timeout on the lock, because the compiler occasionally never returns, so in theory it can still race. Second, related crash If serializeAsArchiveAndFlushToURL: runs while other threads are still compiling into that serializer, we get a different crash (null deref at 0x30) in dispatch_data_create_map ← -[_MTLBinaryArchive newArchiverIdWithBinaryKey:entry:srcArchiverId:functionType:]. Our standalone repro hits this one reliably. The getElementAtDestination crash only shows up with our real workload so far. Repro + details https://github.com/Aeonitis/metal4-capture-race It's a small Objective-C repro (Metal + Foundation only) with a bunch of modes, the results so far, and a comparison with how Vulkan (VkPipelineCache) and D3D12 spell out the thread-safety rules for this same pattern. Questions Is capturing from multiple threads into one MTL4PipelineDataSetSerializer meant to be supported? If not, could the docs say so? Is one compiler + one serializer per thread a safe workaround, or is there shared state between compilers in the same process that would still race? Is calling serializeAsArchiveAndFlushToURL: while compiles are still in flight on that serializer allowed? Is this already fixed in a seed
Replies
0
Boosts
0
Views
306
Activity
3d
Request for PROLED display to support Adaptive Sync (VRR)
I was hoping with an OLED display that the new MBP might support true Adaptive Sync aka VRR. Alas, it only gets ProMotion, with the same huge gap between supported refresh rates at the top end (60hz and 120hz). Hooking up an HDMI VRR display such as any LG TV results in a display that works much better with variable framerates than the built-in display, since you can deliver frames at any interval between the min and max. VRR greatly improves games and already works great in macOS with an external display, so I humbly request that Apple consider another option next to ProMotion, for Adaptive Sync/VRR.
Replies
0
Boosts
0
Views
206
Activity
4d
Xcode 27.0 Metal Toolchain download stuck at "Beginning asset download..."
Hi everyone, I downloaded Xcode 27.0 today because I need the Metal Toolchain for Unreal Engine 5.6 on my Mac. Xcode itself is installed and working, but Unreal is having issues because the Metal Toolchain doesn't seem to be installed. When I run: xcrun metal --version I get: error: error: cannot execute tool 'metal' due to missing Metal Toolchain; use: xcodebuild -downloadComponent MetalToolchain I tried: sudo xcodebuild -downloadComponent MetalToolchain but it just stays on: Beginning asset download... If I cancel it, I get: xcodebuild: error: Download was cancelled. (Asset download for com.apple.MobileAsset.MetalToolchain ...) I've also run: xcodebuild -runFirstLaunch and installed the latest Command Line Tools for Xcode 27.0 through Software Update. xcrun --find metal does find the executable, but the actual Metal Toolchain appears to be missing. I'm mainly trying to get this working so I can use Unreal Engine 5.6. Has anyone had this issue with Xcode 27, or know the correct way to install the Metal Toolchain when xcodebuild -downloadComponent gets stuck at "Beginning asset download"? Thanks!
Replies
1
Boosts
1
Views
501
Activity
6d
D3DMetal: present hangs permanently when the game window becomes occluded
Environment macOS 27.0.1 (26A434), Apple M4 Pro Apple Game Porting Toolkit 4.0b2 (D3DMetal.framework) via CrossOver 27 D3D12 game, borderless fullscreen window (2540x1650) Symptom When the game window becomes occluded — Mission Control, a system OSD (volume/brightness), or another window being brought to the front — the game never renders another frame. It does not crash: GPU utilization drops to 0%, the process stays alive with near-zero CPU, and it never recovers. Killing and restarting the game is the only way out. D3DMetal log at the moment it happens: [D3DMetal:FATAL:%llX][%s:%d] Failed to get drawable for HWND. Run with D3DM_NO_WINDOW=1 env. Question When CAMetalLayer/nextDrawable has no drawable (window not visible), is there a supported way for D3DMetal to skip frames instead of blocking forever? Is there an environment variable or configuration that makes D3DMetal treat "no drawable" as a recoverable condition? Does D3DM_NO_WINDOW=1 apply here, or is it only for headless/offscreen use? I understand apps should handle a nil drawable and continue, but this path is inside D3DMetal. Any guidance would help.
Replies
3
Boosts
0
Views
1.2k
Activity
6d
iOS app running on macOS 27.0.1 terminated on suspension due to a Metal archiveUsage.db file lock (0xDEAD10CC)
Hello, We are investigating an issue where our iOS app, running as a “Designed for iPad” app on an Apple Silicon Mac, is terminated when it is suspended in the background. Environment OS: macOS 27.0.1 Hardware: Apple Silicon Mac App type: An iOS/iPadOS app running on Mac, rather than a separately developed macOS app The app uses Core Image’s CIContext to generate some image effects Steps to reproduce Launch the app on Mac and start playing a podcast. Move the app to the background. Start playback on another device. Our app’s cross-device playback coordination pauses playback on the Mac. After a short period, the Mac app is terminated. We have also observed similar termination after visiting a screen that uses Core Image image effects. Actual result The debugger reports: Terminated due to signal 9 RunningBoard logs indicate that the app was terminated with code 0xDEAD10CC because it was suspended while holding a disallowed shared file lock. The file identified in the log is: /var/folders/.../C/app.podcast.cosmos/app.podcast.cosmos/com.apple.metal/archiveUsage.db/lock.mdb The relevant log message includes: suspended while holding a shared file lock Investigation so far Using breakpoints on open and openat, we captured a call stack corresponding to: App creates a CIContext → Core Image CI::MetalContext → Metal recordBinaryArchiveUsage → archiveUsage.db/lock.mdb is opened We also tried: CIContext(options: [.useSoftwareRenderer: true]) However, in our test environment, we still observed Metal initialization and access to the relevant cache files. This option did not avoid the issue. For the gradient-mask generation code we had identified, we replaced the Core Image implementation with a CPU implementation using CGContext and CGGradient, avoiding CIContext creation on that path. After repeatedly testing the same reproduction steps, we could no longer reproduce the termination in that scenario. This suggests a connection to the Metal cache file lock accessed during Core Image initialization. However, we have not established the exact lifetime or internal owner of the lock, and we have not confirmed a thread deadlock. Questions Is this a known issue involving Core Image/Metal and suspension of iOS apps running on macOS 27.0.1? Why would a system-managed Metal cache file lock cause the app to be terminated with 0xDEAD10CC when suspended? Is there a recommended context configuration, resource cleanup procedure, or lifecycle handling approach that allows us to continue using Core Image safely in this environment?
Replies
1
Boosts
0
Views
76
Activity
1w
Xcode 27 - Device Hub - backboardd crashes when displaying multiple apps under iPadOS 27
I'm getting persistent backboardd crashes in the iPadOS 27 simulator when my app and another app (Safari, Photos or Calendar, etc.) has windows visible on the screen at the same time, including slide over, etc. Steps to reproduce: Launch my app in an iPadOS 27 simulator. Open Safari or Photos in another window. Arrange both app windows so they're visible. Within a second backboardd crashes with EXC_CRASH (SIGABRT) in a com.apple.coreanimation.display.primary thread. The simulator log reports… -[MTLTextureDescriptorInternal validateWithDevice:]: error 'Texture Descriptor Validation MTLTextureDescriptor has invalid pixelFormat (0).' Xcode subsequently reports connection-interrupted messages and ends the app's debug session with: “Debug session ended with code 9: Terminated due to signal 9.” This works correctly in iPadOS 26.x simulators and on a physical device running iPadOS 27. I can reproduce this with both Xcode 27 and 27.1. Resetting the simulator and testing different simulated iPad models does not resolve it. Is anyone else seeing anything like this? It's presumably something I'm doing wrong, so I've double checked my UIScene handling and all seems well (it's been in there for a while) so I'm somewhat at a loss as to where to look next. This is on a Mac Mini M4 Pro running macOS 27.0.1 (26A434).
Replies
1
Boosts
0
Views
264
Activity
1w
Metal Compositor Service & Persona (VisionOS)
Hello, I'm currently trying to make a collaborative app. But it just works only on Reality View, when I tried to use Compositor Layer like below, the personas disappeared. ImmersiveSpace(id: "ImmersiveSpace-Metal") { CompositorLayer(configuration: MetalLayerConfiguration()) { layerRenderer in SpatialRenderer_InitAndRun(layerRenderer) } } Is there any potential solution too see Personas in Metal view? Thanks in advance!
Replies
3
Boosts
0
Views
2k
Activity
3w
Photogrammetry with Object masks hangs and terminates with masks of objects
[PhotogrammetrySample]) with objectMask set and traps on my ios26.5.2, see the attached screenshot on feedback FB24379913 , it gets to the function and hangs . Even the folder reconstruction with lazy sequence as recommended from your video sample also doesn;t complete as it can't find alignment and displays the CoreOC.PhotogrammetrySession.Error error 6 means alignment failed. What can be done or do you guys expose any functions that can be used to check or trace or handle these internally The ObjectMasks are actually segmentation masks from an segmentation algorithm I will appreciate a timely response and willing to provide more clarity and informations, thank you so much for your understanding
Replies
3
Boosts
0
Views
1.5k
Activity
3w
Clarifying Fence Behavior in Metal 4
I have some clarifying questions on how fences work in Metal 4. If two encoders A, B update the same fence and encoder C waits on that fence, is the work encoded by C guaranteed to execute after both A and B have completed, or after either A or B have completed? The following quotes in the waitForFence() documentation seem contradictory. "Encodes a command that instructs the GPU to pause before starting one or more stages of the pass until a pass updates a fence." "When encoding a pass that reuses a fence, wait for other passes to update the fence before repurposing that fence..." Are fences unsignaled when they are waited on? Specifically, can two (or more) encoders wait on a fence that is only updated in a single prior encoder? If an encoder updates a fence, no subsequent encoders in the command queue wait on the fence, and the command queue is committed, is that fence still signaled when used in a subsequent command buffer?
Replies
1
Boosts
0
Views
2.1k
Activity
Oct ’26
On-screen RealityView starves CADisplayLink to 30 Hz on ProMotion (Mac Catalyst)
FB24536235 On Mac Catalyst under macOS 27, a plain CADisplayLink asking for CAFrameRateRange(minimum: 60, maximum: 60, preferred: 60) gets serviced at 30 Hz for as long as a RealityView is on screen in the same window. The link does nothing per tick but count, so there's nothing of mine to blame it on. RealityKit's own statistics overlay reads 60.41 fps in the same frame. Click a segmented control that removes the RealityView and the same link goes straight back to 60. Nothing else changes. That's the whole reproducer, and I've attached it to the radar. It only happens while the display panel is in ProMotion mode. Set the built-in to a fixed 60 Hz and it's correct again. With an external 60 Hz display attached the roles swap: the built-in is fine and the external drops to somewhere between 18 and 30, and setting the built-in to 60 Hz fixes that one too without touching the external's own settings. A raw MTKView presenting continuously at 60, at 120, and on a 120 Hz link presenting every second callback are all fine, so it isn't continuous presentation and it isn't the requested rate. It's RealityKit specifically. Worth knowing if you're testing: RealityView on Catalyst is an ARView underneath, so both paths give you the same answer. This is VERY rough for anything that puts RealityKit next to a UI. In an editor that's the sidebar, the inspector, gizmos, drag handles, every display-link-driven or UIKit animation in the window running at half rate around a viewport that stays smooth. Likely Related to FB24091347, which is the same defect seen as SwiftUI scroll judder. If you can reproduce either, please file a duplicate. Attached two screenshots; first with promotion enabled, second with promotion off. PLEASE fix this, it drives me crazy and there seems to be no workaround. On release day of macOS 27 our app will likely be blamed for it by users and my hands are tied. Thank you!
Replies
2
Boosts
1
Views
1.4k
Activity
Oct ’26
Contribution route for native Metal point-cloud operators used with PyTorch MPS
I maintain mps-pointops, an open-source library implementing farthest-point sampling (FPS), k-nearest-neighbor search, and radius / Ball Query operations with custom Metal kernels for PyTorch MPS tensors. The intended use is 3D point-cloud pipelines whose existing extensions depend on CUDA. I would like to prepare a contribution proposal at the appropriate layer. Apple’s PyTorch/MPS page links to the PyTorch MPS backend contribution guide and directs backend issues to PyTorch’s GitHub tracker with the module: mps label. The current implementation has numerical-contract tests for selection order, ties, radius boundaries, and padding, plus benchmark records from physical M1 and M5 Pro Macs. The repository contains the kernels, Python interfaces, tests, and measurement conditions: https://github.com/gamzerA/mps-pointops My questions are: For specialized operators such as FPS and radius search, is the recommended contribution structure to keep the operator/API in a domain library (for example, a point-cloud or graph library), while proposing only reusable backend primitives to PyTorch MPS? Is there a public Apple-maintained contribution or technical-review channel relevant to this work, beyond the PyTorch route linked from Apple’s page? For an external library that dispatches custom Metal work alongside PyTorch MPS operations, which documented extension/interoperability path should be the long-term target for command-stream ordering and buffer lifetime? I would like to minimize reliance on backend internals when preparing an upstream proposal. Pointers to the appropriate public interface, maintainers, or contribution requirements would help me scope a small, reviewable proposal. I understand that inclusion in an upstream project is subject to its maintainers’ review. Apple documentation referenced: https://developer.apple.com/metal/pytorch/
Replies
0
Boosts
0
Views
208
Activity
Oct ’26
How to restrict an App Store distribution exclusively to Metal 4 devices?
I am developing an application exclusively for devices that support Metal 4. Is it possible to publish this app on the App Store and restrict its availability solely to compatible hardware? The iphone-performance-gaming-tier key is not an option, as we need to maintain support for older devices.
Replies
0
Boosts
0
Views
264
Activity
Sep ’26
Is there a way to directly go from VideoToolbox to Metal for 10-bit/BT.2020 YCbCr HEVC?
tl;dr how can I get raw YUV in a Metal fragment shader from a VideoToolbox 10-bit/BT.2020 HEVC stream without any extra/secret format conversions? With VideoToolbox and 10-bit HEVC, I've found that it defaults to CVPixelBuffers w/ formats kCVPixelFormatType_Lossless_420YpCbCr10PackedBiPlanarFullRange or kCVPixelFormatType_Lossy_420YpCbCr10PackedBiPlanarFullRange. To mitigate this, I have the following snippet of code to my application: // We need our pixels unpacked for 10-bit so that the Metal textures actually work var pixelFormat:OSType? = nil let bpc = getBpcForVideoFormat(videoFormat!) let isFullRange = getIsFullRangeForVideoFormat(videoFormat!) // TODO: figure out how to check for 422/444, CVImageBufferChromaLocationBottomField? if bpc == 10 { pixelFormat = isFullRange ? kCVPixelFormatType_420YpCbCr10BiPlanarFullRange : kCVPixelFormatType_420YpCbCr10BiPlanarVideoRange } let videoDecoderSpecification:[NSString: AnyObject] = [kVTVideoDecoderSpecification_EnableHardwareAcceleratedVideoDecoder:kCFBooleanTrue] var destinationImageBufferAttributes:[NSString: AnyObject] = [kCVPixelBufferMetalCompatibilityKey: true as NSNumber, kCVPixelBufferPoolMinimumBufferCountKey: 3 as NSNumber] if pixelFormat != nil { destinationImageBufferAttributes[kCVPixelBufferPixelFormatTypeKey] = pixelFormat! as NSNumber } var decompressionSession:VTDecompressionSession? = nil err = VTDecompressionSessionCreate(allocator: nil, formatDescription: videoFormat!, decoderSpecification: videoDecoderSpecification as CFDictionary, imageBufferAttributes: destinationImageBufferAttributes as CFDictionary, outputCallback: nil, decompressionSessionOut: &decompressionSession) In short, I need kCVPixelFormatType_420YpCbCr10BiPlanar so that I have a straightforward MTLPixelFormat.r16Unorm/MTLPixelFormat.rg16Unorm texture binding for Y/CbCr. Metal, seemingly, has no direct pixel format for 420YpCbCr10PackedBiPlanar. I'd also rather not use any color conversion in VideoToolbox, in order to save on processing (and to ensure that the color transforms/transfer characteristics match between streamer/client, since I also have a custom transfer characteristic to mitigate blocking in dark scenes). However, I noticed that in visionOS 2, the CVPixelBuffer I receive is no longer a compressed render target (likely a bug), which caused GPU texture read bandwidth to skyrocket from 2GiB/s to 30GiB/s. More importantly, this implies that VideoToolbox may in fact be doing an extra color conversion step, wasting memory bandwidth. Does Metal actually have no way to handle 420YpCbCr10PackedBiPlanar? Are there any examples for reading 10-bit HDR HEVC buffers directly with Metal?
Replies
3
Boosts
0
Views
2.2k
Activity
Sep ’26
macOS 27 beta: ProMotion refresh cadence is unstable, causing constant scroll judder
FB24091347 On macOS 27.0 beta (26A5388g), MacBook Pro M4 Pro, the built-in ProMotion display never settles on a stable refresh cadence. Scrolling in SwiftUI judders constantly. The same app binary was smooth on macOS 26, and is smooth on a 120 Hz ProMotion iPad. I captured two 60-second Instruments traces — same app, same scene, same scrolling, no external display — changing only the display's refresh-rate setting. On ProMotion the vsync interval standard deviation is 4.093 ms across six different cadences, mostly flip-flopping between 120 Hz and 60 Hz. Forced to a fixed 60 Hz it drops to 0.391 ms with a single cadence. The app presented an identical 59 fps median in both runs — frame production is perfectly steady, the display just holds each frame for an unpredictable length of time. That's what makes this nasty: it's invisible to every frame-rate metric, so it looks like the app got slow when nothing about the app changed. I spent most of a day profiling my own code before realising the app was never the problem. Workaround: force the built-in display to 60 Hz. Worth noting, because it complicates the picture: attaching a 60 Hz Studio Display makes the built-in smooth, but the Studio itself then judders — despite its own vsync cadence measuring perfectly stable. So refresh rate alone isn't the whole story, and there may be a second mechanism. The clean, reproducible, single-variable result is the ProMotion vs forced-60 Hz comparison on the built-in panel. If you can reproduce this on an M-series MacBook Pro on 27 beta, please file a duplicate referencing FB24091347.
Replies
6
Boosts
2
Views
2.7k
Activity
Sep ’26
Metal FP32 arithmetic rounding-mode and denormal controls for deterministic shaders
Metal already provides a strong precise floating-point contract. With fast math disabled (mathMode = .safe and mathFloatingPointFunctions = .precise, or -fno-fast-math), the Metal Shading Language specification requires correctly rounded FP32 add, subtract, multiply, reciprocal, divide, sqrt, rsqrt, and fma. I am looking for clarification and, if necessary, API support for the two remaining pieces needed for portable bit-exact numerical shaders: Arithmetic rounding mode MSL §8.2 says either round-to-nearest-ties-to-even or round-toward-zero may be supported for floating-point operations. I cannot find a way to select or query the arithmetic rounding mode. The newer MTLCompileOptions.floatingPointConversionRoundingMode appears to apply only to narrowing float-to-float conversions, not arithmetic operations. Do all currently supported Apple GPU families use round-to-nearest-ties-to-even for precise FP32 add/subtract/multiply/divide/sqrt? If so, could that be made a documented guarantee? Otherwise, could Metal expose an arithmetic rounding-mode compile option and a corresponding MTLDevice capability query? Denormal behavior MSL §8.1 and §8.5 permit denormalized FP32 operands and results to be flushed to zero, including with fast math disabled. I cannot find a control or capability query for preserving denormal inputs and results. Do current Apple GPU families support denormal-preserving FP32 arithmetic? Could Metal expose a preserve/flush mode and a MTLDevice query? A convenient end state would be a queryable strict FP32 configuration combining: safe math; precise FP32 functions; contraction disabled when separate rounding points are required; round-to-nearest-ties-to-even arithmetic; preserved FP32 denormal inputs and results; defined signed-zero, infinity, and NaN behavior. My use case is deterministic GPU numerical simulation. A small compute-shader probe can identify effective rounding and denormal behavior on one GPU/OS/compiler combination, but it cannot provide the portable or future-proof contract needed by applications and higher-level APIs such as WebGPU. Related cross-API work: SPIR-V/Vulkan: https://github.com/KhronosGroup/SPIRV-Registry/issues/448 HLSL/DXIL/D3D12: https://github.com/microsoft/hlsl-specs/issues/926 WebGPU/WGSL umbrella issue: https://github.com/gpuweb/gpuweb/issues/2259 Relevant Metal documentation: Metal Shading Language Specification, §§1.6.3 and 8.1–8.5: https://developer.apple.com/metal/Metal-Shading-Language-Specification.pdf MTLCompileOptions.mathMode: https://developer.apple.com/documentation/metal/mtlcompileoptions/mathmode MTLCompileOptions.floatingPointConversionRoundingMode: https://developer.apple.com/documentation/metal/mtlcompileoptions/floatingpointconversionroundingmode
Replies
2
Boosts
0
Views
3.2k
Activity
Sep ’26
Data corruption when using MTLBlitCommandEncoder.copy from buffer to texture
When we use MTLBlitCommandEncoder.copy to copy from buffer to textures with specific size, we find that the texture content is corruppt. The specific rules we found are: On Apple GPUs (reproduced on Apple M5, macOS 26.3 and iPhone 13 / A15), for ASTC textures uploaded to private storage via the blit encoder, the sampler / texel-fetch mis-addresses the tail of a mip level when ALL of the following hold: that level's texel width is an exact multiple of the 16 KB page width (32 blocks) — 128 texels for ASTC 4x4, 192 for 6x6, 256 for 8x8; the texture width is not divisible by 2^level (a partial block column exists in the chain); that level is taller than one page (>32 block rows). The level's first 128 texel rows (first page row, 4x4) are always read correctly; everything beyond is mis-addressed. Reads landing on unmapped/invalid memory decode as opaque magenta (1,0,1); reads aliasing valid memory show wrong image content. This produced visible purple artifacts in a shipping game (impostor tree atlas, 514x1024 ASTC 4x4, mip2 bottom half). The minimum reproduce code is: ASTCMipTailBugRepro.swift The output is: 514x1024 blit(private) mip2: CORRUPTED (bottom of mip2 mis-addressed) 512x1024 blit(private) mip2: INTACT 514x1024 replaceRegion mip2: INTACT
Replies
1
Boosts
0
Views
1.6k
Activity
Sep ’26