When constructing a MTLIndirectAccelerationStructureInstanceDescriptor, one specifies an accelerationStructureID. In the equivalents in both Vulkan and DirectX12, one can set it to zero to be an "inactive" instance. However, on Metal, this field does not appear to have any documentation, and thus it is difficult to figure out if there is similar behavior (and no other Metal documentation seems to mention this). Does setting this field to 0 (i.e. null) disable the instance? If not, is there any other way to have an equivalent effect?

I have been using a small macOS research project to exercise Metal with a workload that differs from rendering and dense machine-learning kernels. Autolykos v2 is useful for this because it combines a large, height-dependent working set with pseudo-random reads, integer-heavy hashing, sustained execution, and periodic replacement of the dataset. The project happens to be a miner, but my question here is strictly about Metal compute behaviour. On the Apple M4 system used for these measurements, the full dataset contained 216,430,305 elements of 32 bytes each: 6.93 GB, or about 6.45 GiB, held in a .storageModePrivate buffer. For every nonce, the search kernel: computes an index seed, performs 32 pseudo-random 32-byte dataset reads, accumulates the eight 32-bit limbs into a 256-bit sum, and applies a final BLAKE2b compression and target comparison. The normal dispatch uses 128 threads per threadgroup. The wider pipeline also builds the next height's dataset in chunks on a separate command queue while search continues, and keeps two search command buffers in flight. I record command-buffer wall time, gpuStartTime/gpuEndTime, unions of overlapping intervals, and thermal state. To estimate the ceiling imposed by the random gathers, I added a non-consensus microbenchmark. It retains the normal seed calculation, index distribution, all 32 dataset reads, and the complete accumulation, but omits the final BLAKE2b compression. The accumulated result remains observable through a comparison, so the gather loop cannot simply disappear. I expected this stripped kernel to be at least slightly faster. Instead, an order-balanced campaign on an M4 produced: complete search kernel: 3.108 million nonces/s median active throughput gather-only kernel: 2.952 million nonces/s ratio: 105.3% All four same-round ratios were between 103.18% and 105.74%. Each measured run used the full dataset, a 30-second search interval, an excluded warm-up, and a start-temperature gate below 50 °C. Both compute pipeline states reported maxTotalThreadsPerThreadgroup == 1024. My conservative conclusion was not to pursue speculative register-pressure or manual memory-level-parallelism rewrites. The access pattern appears sufficiently dominant, while the supposedly simpler microbenchmark may have changed the compiled pipeline in a way that makes it a poor upper-bound model. My questions are: Can removing the trailing arithmetic legitimately make a memory-latency-heavy Metal kernel slower by changing register allocation, instruction scheduling, or the amount of useful latency hiding? Or would you first suspect a flaw in this kind of gather-only benchmark construction? Also, which Metal GPU counters are the most reliable way to distinguish memory-latency saturation from register-limited occupancy in a long-running compute kernel? I am looking at compute occupancy, buffer and ALU limiters, bandwidth, and cache behaviour, but maxTotalThreadsPerThreadgroup alone is clearly too coarse to explain the result. This is one hardware-specific observation rather than a general claim about Apple GPUs. If useful, I can reduce the workload to a smaller standalone reproducer. The source code, benchmark driver, and complete campaign report are available here: https://github.com/giffeler/ergometal The detailed measurements and validation procedure for this comparison are documented here: https://github.com/giffeler/ergometal/blob/main/Benchmarks/2026-08-15-search-gather-ceiling-ab.md

I know this might be a peeve of mine, but looking into programming a simple Metal4 Compute Shader example, essentially updating the Performing Calculations on a GPU example code to work with CPP and Metal4. I found that MTL::Allocation and MTL::Buffer pointers can't be used interchangeably when you are trying to add allocations to a MTL::ResidencySet, this is forcing you to: Know by heart that they inherit from each other and that you can just cast them (this is a bit suspicious though, it did work for me). Forcefully either C-cast or reinterpret_cast the MTL::Buffer pointer to a MTL::Allocation pointer as the MTL::ResidencySet will only accept that type. I might as well just be plain wrong about how this is used, any tips on correct usage in that case? Is there any expectation to either provide a typecast operator or add inheritance to support the expected behaviour seen in Swift and ObjC, which is just passing the thing? Opened a report with # FB24534953 with some extra information. Bear in mind that the example code uses Premake5, but it can generate an Xcode solution easily.

I've just downloaded the Xcode 27.0 beta 5 on a macOS 26.6.1 machine and tried to build my app (which includes Metal CoreImage kernels). I'm met with a new (to me) error; /Users/…/Developer/…/air-lld:1:1 symbol(s) not found for target 'air64_v28-apple-macosx26.0.0' and from the build log; air-lld: warning: ignoring file '/Applications/Xcode-beta.app/Contents/Developer/Platforms/MacOSX.platform/Developer/SDKs/MacOSX.sdk/System/Library/Frameworks/CoreImage.framework/Versions/A/CoreImage.metallib', file AIR version (2.9) is bigger than the one of the target being linked (2.8) air-lld: error: symbol(s) not found for target 'air64_v28-apple-macosx26.0.0' metal: error: air-lld command failed with exit code 1 (use -v to see invocation) I have no MTL_LANGUAGE_REVISION in my build settings. If I add one, with value Metal41 the app builds fine, but crashes at runtime as Metal 4.1 isn't supported on macOS 26. I imagine this is a beta Xcode and or macOS SDK bug, is there a workaround?

like: let mat = try await ShaderGraphMaterial(named: "matname", from: "reality") currently I use an USD file with materials like so and it works: try await ShaderGraphMaterial(named: "/Root/matname", from: "file.usda", in: appBundle) when i try it with .reality i get "NameNotFound". so is it possible or do i have to have a bunch of dummy entities with my materials assigned so i can find the entity>components>material? or what's the best way to author materials in RCP3 for quick access in realitykit?

i can read the depth and mix it with the rendered image and get a linear fog effect but I can't get a true radial fog. I've been going back and forth with gemini and chatgpt and neither can do it. gemini got me a radial gradient but it's jsut projected flat across all 3d objects. it's been all day trial after trial this is what gives me the flat radial gradient using namespace metal; struct DepthFogEffectConstants { float4 fogColor; float density; // We no longer need the inverse projection matrix here! }; kernel void depthFogKernel( texture2d<half, access::read> inColor [[texture(0)]], texture2d<float, access::read> inDepth [[texture(1)]], texture2d<half, access::write> outColor [[texture(2)]], constant DepthFogEffectConstants& uniforms [[buffer(0)]], uint2 gid [[thread_position_in_grid]]) { float w = outColor.get_width(); float h = outColor.get_height(); if (gid.x >= w || gid.y >= h) { return; } half4 originalColor = inColor.read(gid); float rawDepth = inDepth.read(gid).r; // 1. Guard check for empty backgrounds/skyboxes if (rawDepth <= 0.00001f || rawDepth >= 0.9999f) { outColor.write(originalColor, gid); return; } // 2. Map screen pixels from the center of the lens (-1.0 to 1.0) float2 screenPos = float2( ((float(gid.x) / w) * 2.0f) - 1.0f, 1.0f - ((float(gid.y) / h) * 2.0f) ); // 3. Since rawDepth is already acting as a view-space Z proxy, // we use it to calculate the true spherical ray distance from the lens center. // The hypotenuse of screen offset (X, Y) and depth (Z) gives the radial distance. float radialDistance = sqrt(screenPos.x * screenPos.x + screenPos.y * screenPos.y + rawDepth * rawDepth); // 4. Calculate exponential fog matching your visual test float fogFactor32 = exp(-radialDistance * uniforms.density); half fogFactor = half(clamp(fogFactor32, 0.0f, 1.0f)); half4 fogColor = half4(uniforms.fogColor); // Mix and write colors out cleanly half4 finalColor = mix(fogColor, originalColor, fogFactor); outColor.write(radialDistance, gid); } in this code i'm just outputing the radial distance to see that calculation and it's just wrong. i don't know what to do anymore

AI was getting confused and kept correcting, but didn’t have a overriding translation.

Filed as FB24422691. Posting here as well because the crash leaves no usable stack in the application's own crash handler, and searching for "0x1BFFFFFFE4" or "FindClosestMatchingMode" returns nothing anywhere, so this may save someone else the debugging. SYMPTOM A Direct3D 11 Windows application (Unity 2021.3) running fullscreen under Game Porting Toolkit dies when you Cmd+Tab out of it and back in. Exclusive and borderless fullscreen both crash; windowed never does. It is intermittent: on my machine it takes 5 to 9 switches. Always the same signature: EXCEPTION_ACCESS_VIOLATION (0xC0000005), reading address 0x1BFFFFFFE4, faulting RIP at D3DMetal+0xFA8F. Identical in all 16 crash reports I collected, which is what makes it a deterministic underflow rather than heap corruption. CAUSE D3DMetal's DXGIOutput::FindClosestMatchingMode calls GetDisplayModeList and then reads the last element of the list without checking the count or the buffer pointer: call DXGIOutput::GetDisplayModeList(...) ; count returned in [rsp+0x3c] mov eax, [rsp+0x3c] ; eax = count dec eax ; 0xFFFFFFFF when count == 0 lea rcx, [rax + 8rax] ; rcx = 28 * eax (28 = sizeof DXGI_MODE_DESC) lea rcx, [rcx + 2rcx] add rcx, rax mov rax, [r14 + rcx] ; reads modes[-1] 28 * 0xFFFFFFFF = 0x1BFFFFFFE4, and r14 (the mode buffer) is NULL on that path, so the read lands exactly on the address seen in the crash reports. WHY THE LIST IS EMPTY winemac.drv rebuilds the display device list on every application activation. With WINEDEBUG=+display, four Cmd+Tab activations produce exactly four full rebuilds (macdrv_UpdateDisplayDevices: GPU count, adapter, monitor). An application that queries the closest matching mode mid-rebuild gets zero modes back. Fullscreen makes that query on focus changes; windowed does not, which is exactly why windowed never crashes. The empty-list condition is not unique to this race: GetDisplayModeList has also been reported returning 0 modes for DXGI_FORMAT_R16G16B16A16_FLOAT on D3DMetal 2.1 (github.com/vec715/enfusion-dxgi-fix). DXVK and DXMT both return DXGI_ERROR_NOT_FOUND for an empty list instead of dereferencing it. MEASUREMENTS Alternating applications automatically and verifying every focus change: borderless fullscreen: crash after 5, 6, 6 and 8 switches (4 of 4 runs) exclusive fullscreen: crash after 9 switches windowed: 105+ switches, no crash As a control I patched a local copy of D3DMetal so the read is skipped when count == 0 or the buffer is NULL, keeping the requested mode. Same machine, same setup: 140 switches in fullscreen with no crash, and the unpatched binary crashed again after 8 switches immediately afterwards. VERSIONS The unchecked read is present in both D3DMetal builds I have: 2.0 (built for macOS 13.3) at 0x1453A, and 3.0 (built for macOS 15.4) at 0xFA82. Environment: macOS 26.5.2 (25F84), Apple M5, D3DMetal 3.0 inside Game Porting Toolkit, Wine 7.7, 64-bit prefix, fullscreen at 2560x1664 with winemac.drv Retina mode on. WORKAROUND UNTIL IT IS FIXED Run the application windowed, or interpose a dxgi proxy DLL that returns DXGI_ERROR_NOT_FOUND from FindClosestMatchingMode when the mode count is 0. Suggested fix: return DXGI_ERROR_NOT_FOUND from FindClosestMatchingMode / FindClosestMatchingMode1 when the count is 0 or the buffer is NULL, instead of indexing modes[count-1].

Metal Performance Shaders produces from bf16/fp16 softmax for large attention tensors on , forcing the entire local generative-AI ecosystem to fall back to fp32. This is a . Environment Minimal reproduction Decomposed softmax on MPS — diffs = x - maxes produces NaN even though every element is identical (result should be all zeros): Real-world impact Running ComfyUI (the dominant local generative-AI UI) on M5 Max: Root cause (per PyTorch MPS maintainers) PyTorch maintainers (@drisspg, @albanD) have traced this to in the MPS/MPSGraph fused kernels. The NaN originates in the x - maxes subtraction inside softmax, then propagates through the attention block and the entire network. PyTorch cannot fix this — it is in the Metal/MPS kernel layer. Not isolated The M4-era fix ("fixed on macOS 15.1") did not survive onto M5, indicating the MPS fused-kernel precision fix was either reverted or not ported to the M5 GPU architecture. Request

Xcode hangs when I call PhotogrammetrySession(input: [PhotogrammetrySample]) with objectMask set and traps on some devices, see the attached screenshot, it gets to the function and hangs. Even the folder reconstruction also doesn't complete as it can't find alignment and displays the CoreOC.PhotogrammetrySession.Error 6 and I understand to mean alignment failed. In this case it failed while object masking was ON, so RealityKit could not find enough consistent feature tracks inside the masked pixels across the image set. What can be done or do you guys expose any functions that can be used to aid, or handle these internally, can't find any internally. The ObjectMasks are actually segmentation masks from an ML algorithm . To replicate try calling PhotogrammetrySession(input: [PhotogrammetrySample]) with contentsOf as captured on your documentation, even with like 30 image set or is there something I'm missing. I will appreciate a timely response and willing to provide more clarity and informations, thank you so much for your understanding

Hello Apple! We've got offline shader compilation from HLSL -> Metallib using DXC -> SPIR-V -> metal.exe. This works okay for the most part, but it requires the creation of intermediate files to pass to/from the metal.exe process and we've had some issues with metal.exe sometimes not launching (probably our fault). Then we noticed Metal Shader Converter (MSC) exists and has a DLL - this looks way better since there's no need to launch processes or store intermediate files. However, upon trying to replace metal.exe with it I quickly ran into rampant heap corruption. I was surprised because the docs claim this: Each thread in your program needs to create its own instance of IRCompiler to avoid race conditions. But once I start calling IRCompilerAllocCompileAndLink in parallel all hell breaks loose, whether or not each thread has its own IRCompiler. I figured I must be doing something wrong, so I removed my attempt and compiled DXC locally with the MSC integration and encountered the exact same heap corruption. So I'm inclined to think the library isn't actually thread safe, but I'm wondering if there's something I'm missing? I tried all 3 versions of MSC just in case it was a problem with 3.0, but I got the same result each time. The only way to make it work was to surround compilation with a mutex, which makes its use pointless in our case.

What's the point of residency sets if you can just make a buffer accessible to the GPU through storage mode in metal?

Hallo, mein Spiel Lockpicking ist exakt gleich programmiert für Android und Apple. Warum sehen die links rechts Tasten bei beiden Apps verschieden aus ? https://apps.apple.com/us/app/lockpicking/id6793054294

Hello ! Thank you guys for all the hard work you are doing on RealityKit For now in my studying Im trying to understand, in general, what will cause my Main app view to redraw everything? Or any view? If i have a struct that use the view protocol, and that struct is being called inside RealityView on the main app, or outside a reality view within a ZStack on the main app, if this view updates, does this will cause everything in my main app view to redraw as well? Or is it only true to @State variables of the main app? or the @State variables of the external view i try to implement? and what about @Observable ? Is there an accurate table that tells the developer what will cause a redraw of things? (not only the things the logic asks for, but all things sitting idle within the view) I try to understand how to separate UI updates, to avoid full redraw of things that havent been changed. to minimize UI compute in my games Before i play around with instruments i need to understand the general architecture.. if i work in a way that is counter designed to the way reality kit should work then the instrument results will not make sense to me. so i feel like i should ask you guys first. Depending on your answer i will know how to arrange my data in my game. And how to design and instantiate all my views Thanks

Game Center matchmaking fails for all non-default users on Apple TV — no IDS registration for secondary users (FB24156316) On a multi-user Apple TV (tvOS 26.5, Apple TV 4K 3rd gen), Game Center real-time matchmaking fails for every user except the default user, in every app I've tested — including Apple Arcade titles. Filed as FB24156316 with full logs and sysdiagnose; posting here for visibility and in case anyone has shipped multi-user GC multiplayer on tvOS successfully. My game adopts com.apple.developer.user-management (runs-as-current-user-with-user-independent-keychain). The entitlement itself works: on a secondary user's profile the app runs under that user's persona and GKLocalPlayer authenticates as them — the welcome banner shows the right account. But any GKMatchmakerViewController quickmatch hard-fails within ~9 seconds ("Failed to find players"), and accepting an invite fails with GKError 35 ("not signed in to iCloud") even though Settings shows that user's iCloud as signed in. Unified logs show the root cause. When matchmaking starts, gamed can't provision the player's pseudonym because the current user has no identity-services registration: gamed No URI found on any account -- returning nil gamed Failed to fetch pseudonym for local player. Error: GameDaemonCore.PseudonymManagerError.failedToProvision( internalError: Error Domain=com.apple.ids.IDSPseudonymErrorDomain Code=400 "Invalid URI") For the default user, the identical flow succeeds (identityservicesd … resultCode: 0). Across a full day of log capture — profile adds, a remove/re-add, multiple user switches — identityservicesd never once references the secondary users' accounts: registration for them is never attempted, not attempted-and-failed. Meanwhile gamed advertises the nearby-matchmaking Bonjour service with the default user's identity while the foreground app runs as the secondary user. Reproduction matrix: two apps (my shipping game Extreme Violence and Apple Arcade's Crossy Road Castle, which also runs under the correct persona), both sandbox and production Game Center, two unrelated secondary accounts (both healthy elsewhere). Persists across reboot and profile remove/re-add. Default user unaffected. The documentation says the entitlement is all that's needed ("each person who uses your app will have access to… their own Game Center… you don't have to make any code changes" — WWDC20 session 10645). As far as I can tell that promise is currently unfulfillable for online play: there is no API or Settings path that creates the missing IDS registration. Has anyone seen non-default-user matchmaking work on tvOS, on any version? Is there anything an app can do here, or is this purely an OS-side fix? (Related: thread 782163 — a different tvOS matchmaking failure that DTS confirmed as a bug.)

Hi! I'd like to share an implementation of 2D soft shadows using SpriteKit and custom Metal rendering. GitHub Repo SpriteKit-SoftShadows The demo app runs on iOS and Mac Catalyst. The soft shadow implementation is based on Scott Lembcke's algorithm. Pipeline The app uses MTKView to drive the rendering loop at the desired frame rate. Each frame: SpriteKit renders the scene into a Metal texture using SKRenderer. The CPU sends each light's properties and the relevant shape edges to a Metal vertex shader. The vertex shader projects shadow geometry from each edge. A fragment shader calculates the shadow opacity at each pixel, producing a soft shadow mask for each light. A final fragment shader combines the SpriteKit texture, lights, and shadow masks to produce the displayed image. The SwiftUI controls update variables inside the SpriteKit scene through @Observable. The rendering loop consumes the new values on its next cycle. Let me know if you have any feedback!

I'm writing GPU compute kernels (parallel prefix-sum and histogram) that rely on threadgroup-address-space atomics specifically atomic_fetch_add_explicit on a threadgroup atomic_uint, with memory_order_relaxed. Setup: Device: Apple M5, macOS 26.5 Toolchain reports MSL 4.0 / AIR 2.8 (i.e. the "Metal 4.0" feature level) Note: I'm generating AIR (Apple IR) directly rather than emitting MSL source this is through a custom compute backend (Julia's Metal.jl), not the standard MSL front end. What I observe: These threadgroup atomics compile and produce correct results, and give a meaningful speedup over a non-atomic (scan-based) fallback. I've validated correctness across a 256-bin histogram and a full multi-pass radix sort no mismatches. The question: Some capability checks gate threadgroup atomic support behind Metal 4.1, and my device reports 4.0 yet they clearly work. So: Are relaxed integer threadgroup atomics (atomic_fetch_add on threadgroup atomic_uint, relaxed ordering) officially supported on the Metal 4.0 feature set for current Apple Silicon, or is this unsupported behavior that happens to work? Is there a specific MSL version or GPU family that is the true minimum for these operations? Does the answer differ at the AIR / feature-set level (what I'm targeting) vs. the MSL front end, given I'm feeding AIR to the compiler directly? I ask because a downstream library is (reasonably) hesitant to enable this path unless it's officially supported rather than relying on undefined behavior. Any authoritative guidance or a pointer to the relevant feature-set/GPU-family documentation would be hugely appreciated. Thanks!

Update: Please disregard this report. The behavior was caused by enabling the hardware keyboard for the device in Xcode 27 beta 5’s Device Hub. This appears to expose a virtual “Generic Keyboard” to the connected physical device, causing GCKeyboardDidConnect to be posted even though no physical keyboard is attached directly to it. Simply disabling the hardware keyboard option was not sufficient. I needed to close Device Hub and restart Xcode for the virtual keyboard to be removed and the expected GCKeyboard behavior to resume. I’m seeing what appears to be a GameController regression on physical iPhone and iPad devices running iOS 27 beta 5. When my app launches with no Bluetooth, USB, or Smart Connector keyboard attached, GCKeyboard.coalesced is initially nil. Shortly afterward, the system posts .GCKeyboardDidConnect for a keyboard named “Generic Keyboard,” and GCKeyboard.coalesced becomes non-nil: === Initial state === GCKeyboard.coalesced: nil cannot add handler to 0 from 0 - dropping === GCKeyboardDidConnect #1 === notification.object: GCKeyboard: 0x10b01cd00 'Generic Keyboard'> GCKeyboard.coalesced: Optional( 0x10b01cd00>) The phantom keyboard then appears to remain connected indefinitely. After this happens: Attaching a real hardware keyboard does not post another .GCKeyboardDidConnect. Detaching the real keyboard does not post .GCKeyboardDidDisconnect. GCKeyboard.coalesced remains non-nil even though no hardware keyboard is attached. This may be interacting with the documented coalescing behavior: the connect notification is posted only for the first keyboard, and the disconnect notification only after the last keyboard disconnects. If the phantom “Generic Keyboard” remains present, a real keyboard is never considered the first or last keyboard. Minimal observer code: print("Initial GCKeyboard.coalesced:", GCKeyboard.coalesced as Any) NotificationCenter.default.addObserver( forName: .GCKeyboardDidConnect, object: nil, queue: .main ) { notification in print("GCKeyboardDidConnect") print("notification.object:", notification.object as Any) print("GCKeyboard.coalesced:", GCKeyboard.coalesced as Any) } NotificationCenter.default.addObserver( forName: .GCKeyboardDidDisconnect, object: nil, queue: .main ) { notification in print("GCKeyboardDidDisconnect") print("notification.object:", notification.object as Any) print("GCKeyboard.coalesced:", GCKeyboard.coalesced as Any) } I can reproduce this on both a physical iPhone and a physical iPad. This is not Simulator keyboard forwarding, and no keyboard is attached when the “Generic Keyboard” appears. Has anyone else observed this behavior on iOS 27 beta 5 or found another API that reliably reports whether a real hardware keyboard is attached? Filed as FB24269002.

Hi everyone I'm experiencing an issue with iPadOS 26 regarding multi-touch gesture detection. When performing a quick four-finger gesture (tap and swipe), the system often fails to recognize the input. This especially affects multi-touch gestures, such as rhythm games with difficult levels. Steps to Reproduce: Place four fingers on the screen. Perform a quick tap or a quick horizontal swipe (like the one used to switch apps). Observe whether the gesture is ignored or detected inconsistently. Expected Behavior: 4-finger multitouch gestures should be recognized regardless of gesture speed, just like previous iPadOS versions. Actual Behavior: Gestures fail to be detected when executed quickly—same gestures still work, and miss notes in rhythm games. You can check out my posts on Twitter/x and Facebook: [https://x.com/kokona_fwa/status/1978131164104728949?s=61] Facebook: [https://m.facebook.com/groups/idipad/permalink/24438964899058806/?]

I am developing a metal based ray tracing rendering application (running heavy GPU kernels). I am sometimes "forcefully quitting" my application and I can see the application is not in the activity monitor. But I can see the windowserver is using %97 the GPU. The mac gets hotter and hotter. I kill the windowserver, re-login it is still the case. The only way to fix is to restart the mac. I have checked if there are any zombie processes, there are none. I am 3-4 month into Mac development (I used many rendering APIs e.g. before under Windows and Linux, they release the resources automatically unless the driver is very broken), but I believe when you force quit or exit gracefully, regarding application should release resources. I may be missing some knowledge. Does anybody have an idea? I had added every corner a graceful exit code but once the kernel has some infinite loop the clean up cannot happen. In Windows there are some driver reload mechanisms to recover when GPU is stuck, is there a similar system ?