From 43a97878ce14b72f0981164f87f2e35e14151312 Mon Sep 17 00:00:00 2001 From: Daniel Baumann Date: Sun, 7 Apr 2024 11:22:09 +0200 Subject: Adding upstream version 110.0.1. Signed-off-by: Daniel Baumann --- .../metal/examples/compute/compute-argument-buffer.metal | 14 ++++++++++++++ 1 file changed, 14 insertions(+) create mode 100644 third_party/rust/metal/examples/compute/compute-argument-buffer.metal (limited to 'third_party/rust/metal/examples/compute/compute-argument-buffer.metal') 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 + +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); +} -- cgit v1.2.3