| Ben Clayton | f2ec7f3 | 2021-06-29 11:53:15 +0000 | [diff] [blame] | 1 | #include <metal_stdlib> |
| Ben Clayton | 6a77236 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 2 | |
| Ben Clayton | f2ec7f3 | 2021-06-29 11:53:15 +0000 | [diff] [blame] | 3 | using namespace metal; |
| Ben Clayton | 6a77236 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 4 | struct SB_RW { |
| Ben Clayton | f2ec7f3 | 2021-06-29 11:53:15 +0000 | [diff] [blame] | 5 | /* 0x0000 */ atomic_int arg_0; |
| Ben Clayton | 0a32a72 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 6 | }; |
| 7 | |
| Natalie Chouinard | e902718 | 2024-06-13 18:50:19 +0000 | [diff] [blame] | 8 | int atomicXor_c1b78c(device SB_RW* const tint_symbol) { |
| James Price | e548db9 | 2021-10-28 15:00:39 +0000 | [diff] [blame] | 9 | int res = atomic_fetch_xor_explicit(&((*(tint_symbol)).arg_0), 1, memory_order_relaxed); |
| Natalie Chouinard | e902718 | 2024-06-13 18:50:19 +0000 | [diff] [blame] | 10 | return res; |
| Ben Clayton | 0a32a72 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 11 | } |
| 12 | |
| Natalie Chouinard | e902718 | 2024-06-13 18:50:19 +0000 | [diff] [blame] | 13 | fragment void fragment_main(device int* tint_symbol_1 [[buffer(0)]], device SB_RW* tint_symbol_2 [[buffer(1)]]) { |
| 14 | *(tint_symbol_1) = atomicXor_c1b78c(tint_symbol_2); |
| Ben Clayton | f2ec7f3 | 2021-06-29 11:53:15 +0000 | [diff] [blame] | 15 | return; |
| Ben Clayton | 0a32a72 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 16 | } |
| 17 | |
| Natalie Chouinard | e902718 | 2024-06-13 18:50:19 +0000 | [diff] [blame] | 18 | kernel void compute_main(device int* tint_symbol_3 [[buffer(0)]], device SB_RW* tint_symbol_4 [[buffer(1)]]) { |
| 19 | *(tint_symbol_3) = atomicXor_c1b78c(tint_symbol_4); |
| Ben Clayton | f2ec7f3 | 2021-06-29 11:53:15 +0000 | [diff] [blame] | 20 | return; |
| Ben Clayton | 0a32a72 | 2021-06-18 18:56:13 +0000 | [diff] [blame] | 21 | } |
| 22 | |
