summaryrefslogtreecommitdiffstats
path: root/third_party/rust/metal/examples/compute/compute-argument-buffer.metal
diff options
context:
space:
mode:
Diffstat (limited to 'third_party/rust/metal/examples/compute/compute-argument-buffer.metal')
-rw-r--r--third_party/rust/metal/examples/compute/compute-argument-buffer.metal14
1 files changed, 14 insertions, 0 deletions
diff --git a/third_party/rust/metal/examples/compute/compute-argument-buffer.metal b/third_party/rust/metal/examples/compute/compute-argument-buffer.metal
new file mode 100644
index 0000000000..1dcc79daf5
--- /dev/null
+++ b/third_party/rust/metal/examples/compute/compute-argument-buffer.metal
@@ -0,0 +1,14 @@
+#include <metal_stdlib>
+
+using namespace metal;
+
+struct SumInput {
+ device uint *data;
+ volatile device atomic_uint *sum;
+};
+
+kernel void sum(device SumInput& input [[ buffer(0) ]],
+ uint gid [[ thread_position_in_grid ]])
+{
+ atomic_fetch_add_explicit(input.sum, input.data[gid], memory_order_relaxed);
+}