Skip to content

ComputeGraph iOS xcode27.0 b1

Alex Soto edited this page Jun 9, 2026 · 1 revision

#ComputeGraph.framework

diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraph.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraph.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraph.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraph.h	2026-05-29 23:12:38
@@ -0,0 +1,18 @@
+#pragma once
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphModuleTypes.h>
+
+#include <ComputeGraph/ComputeGraphComputeContext.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+#include <ComputeGraph/ComputeGraphElement.h>
+#include <ComputeGraph/ComputeGraphElementContext.h>
+#include <ComputeGraph/ComputeGraphEmissionContext.h>
+#include <ComputeGraph/ComputeGraphGroupContext.h>
+#include <ComputeGraph/ComputeGraphInitializationContext.h>
+#include <ComputeGraph/ComputeGraphOutputContext.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+#include <ComputeGraph/ComputeGraphSimulationContext.h>
+#include <ComputeGraph/ComputeGraphTextureContext.h>
+#include <ComputeGraph/ComputeGraphTextures.h>
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphComputeContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphComputeContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphComputeContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphComputeContext.h	2026-05-29 23:12:39
@@ -0,0 +1,60 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+
+#include <metal_types>
+
+namespace compute
+{
+struct context_t;
+
+namespace api
+{
+PS_API thread graph_context_t &graph_context(thread compute::context_t &);
+
+// TODO: deprecate these APIs
+PS_API ushort thread_index_in_simdgroup(thread compute::context_t &);
+PS_API ushort thread_position_in_threadgroup(thread compute::context_t &);
+PS_API uint thread_position_in_grid(thread compute::context_t &);
+
+PS_API ushort threads_per_simdgroup(thread compute::context_t &);
+PS_API uint threads_per_threadgroup(thread compute::context_t &);
+
+PS_API uint simdgroup_index_in_threadgroup(thread compute::context_t &);
+}
+}
+
+namespace computegraph
+{
+struct compute_context
+{
+    PS_ALWAYS_INLINE compute_context(thread compute::context_t &_data)
+    : data(_data)
+    {}
+
+    PS_ALWAYS_INLINE compute_context(const thread compute_context &_other)
+    : data(_other.data)
+    {}
+
+    PS_ALWAYS_INLINE computegraph::graph_context graph() const
+    {
+        return computegraph::graph_context(graph_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return computegraph::debug_draw(graph_context());
+    }
+
+    // TODO: private
+    thread compute::context_t &data;
+private:
+    PS_ALWAYS_INLINE thread graph_context_t &graph_context() const { return compute::api::graph_context(data); }
+};
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphContext.h	2026-05-29 23:12:39
@@ -0,0 +1,82 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+#include <ComputeGraph/ComputeGraphTextures.h>
+
+#include <metal_types>
+
+struct graph_context_t;
+
+namespace graph
+{
+struct context_t;
+
+namespace api
+{
+PS_API float age(thread graph_context_t &);
+PS_API float delta_time(thread graph_context_t &);
+
+PS_API metal::float4x4 local_to_world(thread graph_context_t &);
+PS_API metal::float4x4 world_to_local(thread graph_context_t &);
+}
+} // namepsace graph
+
+namespace api
+{
+PS_API thread random_context_t &random(thread graph_context_t &base);
+}
+
+namespace computegraph
+{
+struct graph_context
+{
+    PS_ALWAYS_INLINE graph_context(thread graph_context_t &ctx)
+    : context(ctx)
+    {}
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return api::random(context);
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return computegraph::debug_draw(context);
+    }
+
+    PS_ALWAYS_INLINE computegraph::global_textures textures() const
+    {
+        return computegraph::global_textures(context);
+    }
+
+    PS_ALWAYS_INLINE float age() const
+    {
+        return graph::api::age(context);
+    }
+
+    PS_ALWAYS_INLINE float delta_time() const
+    {
+        return graph::api::delta_time(context);
+    }
+
+    PS_ALWAYS_INLINE metal::float4x4 local_to_world() const
+    {
+        return graph::api::local_to_world(context);
+    }
+
+    PS_ALWAYS_INLINE metal::float4x4 world_to_local() const
+    {
+        return graph::api::world_to_local(context);
+    }
+
+private:
+    thread graph_context_t &context;
+};
+
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphDebugDraw.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphDebugDraw.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphDebugDraw.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphDebugDraw.h	2026-05-29 23:12:39
@@ -0,0 +1,63 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+#include <metal_types>
+#include <metal_array>
+
+struct graph_context_t;
+
+namespace debugdraw
+{
+namespace api
+{
+PS_API void lines(thread debug_draw_context_t &, metal::array_ref<float3> points, half4 color);
+PS_API void line_strip(thread debug_draw_context_t &, metal::array_ref<float3> points, half4 color);
+PS_API bool enabled(thread debug_draw_context_t &);
+
+PS_API thread debug_draw_context_t &graph_debug_draw_context(thread graph_context_t &);
+}
+}
+
+namespace computegraph
+{
+
+struct debug_draw
+{
+    PS_ALWAYS_INLINE debug_draw(thread graph_context_t & _context)
+    : context(debugdraw::api::graph_debug_draw_context(_context))
+    {}
+
+    PS_ALWAYS_INLINE debug_draw(thread debug_draw_context_t & _context)
+    : context(_context)
+    {}
+
+    PS_ALWAYS_INLINE void line(float3 from, float3 to, half4 color)
+    {
+        float3 points[2] = {from, to};
+        debugdraw::api::lines(context, points, color);
+    }
+
+    PS_ALWAYS_INLINE void lines(metal::array_ref<float3> points, half4 color)
+    {
+        debugdraw::api::lines(context, points, color);
+    }
+
+    PS_ALWAYS_INLINE void line_strip(metal::array_ref<float3> points, half4 color)
+    {
+        debugdraw::api::line_strip(context, points, color);
+    }
+
+    PS_ALWAYS_INLINE bool enabled() const
+    {
+        return debugdraw::api::enabled(context);
+    }
+private:
+    thread debug_draw_context_t &context;
+};
+
+}
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElement.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElement.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElement.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElement.h	2026-05-29 23:12:39
@@ -0,0 +1,136 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+#include <metal_types>
+
+struct element_context_t;
+
+namespace element
+{
+namespace api
+{
+PS_API float age(thread element_context_t &element);
+PS_API void set_age(thread element_context_t &element, float age);
+
+PS_API float lifetime(thread element_context_t &element);
+PS_API void set_lifetime(thread element_context_t &element, float lifetime);
+
+PS_API float3 position(thread element_context_t &element);
+PS_API void set_position(thread element_context_t &element, float3 position);
+
+PS_API float2 size(thread element_context_t &element);
+PS_API void set_size(thread element_context_t &element, float2 size);
+
+PS_API half4 color(thread element_context_t &element);
+PS_API void set_color(thread element_context_t &element, half4 color);
+
+PS_API float3 velocity(thread element_context_t &element);
+PS_API void set_velocity(thread element_context_t &element, float3 velocity);
+
+PS_API bool is_alive(thread element_context_t &element);
+}
+}
+
+namespace computegraph
+{
+/// Information about a single element, usable in multiple stages, such as
+/// initialization and simulation.
+struct element_info
+{
+    PS_ALWAYS_INLINE element_info(thread element_context_t &_element)
+    : element(_element)
+    {}
+
+    PS_ALWAYS_INLINE float age()
+    {
+        return element::api::age(element);
+    }
+
+    PS_ALWAYS_INLINE void set_age(float value)
+    {
+        element::api::set_age(element, value);
+    }
+
+    PS_ALWAYS_INLINE float lifetime()
+    {
+        return element::api::lifetime(element);
+    }
+
+    PS_ALWAYS_INLINE void set_lifetime(float value)
+    {
+        return element::api::set_lifetime(element, value);
+    }
+
+    PS_ALWAYS_INLINE float3 position()
+    {
+        return element::api::position(element);
+    }
+
+    PS_ALWAYS_INLINE void set_position(float3 value)
+    {
+        return element::api::set_position(element, value);
+    }
+
+    PS_ALWAYS_INLINE float2 size()
+    {
+        return element::api::size(element);
+    }
+
+    PS_ALWAYS_INLINE void set_size(float2 value)
+    {
+        return element::api::set_size(element, value);
+    }
+
+    PS_ALWAYS_INLINE half4 color() const
+    {
+        return element::api::color(element);
+    }
+
+    PS_ALWAYS_INLINE void set_color(half3 value)
+    {
+        set_color(half4(value, color().a));
+    }
+
+    PS_ALWAYS_INLINE half alpha() const
+    {
+        return element::api::color(element).a;
+    }
+
+    PS_ALWAYS_INLINE void set_alpha(half value)
+    {
+        set_color(half4(color().rgb, value));
+    }
+
+    PS_ALWAYS_INLINE void set_color(half4 color)
+    {
+        element::api::set_color(element, color);
+    }
+
+    PS_ALWAYS_INLINE float3 velocity() const
+    {
+        return element::api::velocity(element);
+    }
+
+    PS_ALWAYS_INLINE void set_velocity(float3 velocity)
+    {
+        element::api::set_velocity(element, velocity);
+    }
+
+    PS_ALWAYS_INLINE void add_velocity(float3 delta)
+    {
+        set_velocity(velocity() + delta);
+    }
+
+    PS_ALWAYS_INLINE bool is_alive() const
+    {
+        return element::api::is_alive(element);
+    }
+private:
+    thread element_context_t &element;
+};
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElementContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElementContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElementContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphElementContext.h	2026-05-29 23:12:39
@@ -0,0 +1,85 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+#include <ComputeGraph/ComputeGraphElement.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+
+#include <metal_types>
+
+struct system_element_context_t;
+
+namespace system_element
+{
+namespace api
+{
+PS_API void terminate(thread system_element_context_t &base);
+PS_API thread graph_context_t &graph_context(const thread system_element_context_t &base);
+
+PS_API uint element_index(thread system_element_context_t &);
+
+PS_API bool is_reader_thread(thread system_element_context_t &);
+PS_API bool is_authority_thread(thread system_element_context_t &);
+PS_API bool is_output_thread(thread system_element_context_t &);
+
+PS_API thread element_context_t &element_data(thread system_element_context_t &);
+
+PS_API uint element_capacity(thread system_element_context_t &);
+}
+}
+
+namespace computegraph
+{
+struct element_context
+{
+    PS_ALWAYS_INLINE element_context(thread system_element_context_t &data)
+    : context(data)
+    {}
+
+    PS_ALWAYS_INLINE uint element_index() const thread
+    {
+        return system_element::api::element_index(context);
+    }
+
+    /// Returns true if the thread executing had an element to read.
+    ///
+    /// @discussion This can be true for output threads that need to write null elements to their output.
+    PS_ALWAYS_INLINE bool is_reader() const thread
+    {
+        return system_element::api::is_reader_thread(context);
+    }
+
+    /// Returns true if the thread executing is the authority on its element
+    ///
+    /// @discussion This can be false for 'helper threads', when simulating strips. These threads
+    /// will calculate values, but not write data. This concept is similar to Metal's helper threads,
+    /// however Metal uses helper threads exclusively for render pipelines.
+    PS_ALWAYS_INLINE bool is_authority() const thread
+    {
+        return system_element::api::is_authority_thread(context);
+    }
+
+    PS_ALWAYS_INLINE computegraph::element_info element() const thread
+    {
+        return computegraph::element_info(system_element::api::element_data(context));
+    }
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return api::random(system_element::api::graph_context(context));
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return computegraph::debug_draw(system_element::api::graph_context(context));
+    }
+private:
+    thread system_element_context_t &context;
+};
+
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphEmissionContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphEmissionContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphEmissionContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphEmissionContext.h	2026-05-29 23:12:39
@@ -0,0 +1,86 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+
+#include <metal_types>
+
+namespace emitter
+{
+struct context_t;
+
+namespace api
+{
+PS_API void spawn(thread emitter::context_t &, int count);
+PS_API thread graph_context_t &graph_context(thread emitter::context_t &);
+
+PS_API float lifetime(thread emitter::context_t &);
+}
+}
+
+namespace computegraph
+{
+/// Context for emitting new particles.
+///
+/// @discussion Adding a ``emission_context`` parameter to a function designates it as
+/// an emission module.
+///
+/// @see ``ParticleEmitterInfo`` for read-only information about the current emitter.
+struct emission_context
+{
+    PS_ALWAYS_INLINE emission_context(thread emitter::context_t &_context)
+    : data(_context)
+    {
+    }
+
+    PS_ALWAYS_INLINE emission_context(const thread emission_context &other)
+    : data(other.data)
+    {
+    }
+
+    PS_ALWAYS_INLINE float delta_time() const
+    {
+        return graph::api::delta_time(emitter::api::graph_context(data));
+    }
+
+    PS_ALWAYS_INLINE float age() const
+    {
+        return graph::api::age(emitter::api::graph_context(data));
+    }
+
+    PS_ALWAYS_INLINE float lifetime() const thread
+    {
+        return emitter::api::lifetime(data);
+    }
+
+    /// Adds ``count`` elements to the current spawn request.
+    PS_ALWAYS_INLINE void spawn(uint count)
+    {
+        emitter::api::spawn(data, count);
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return computegraph::debug_draw(emitter::api::graph_context(data));
+    }
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return api::random(emitter::api::graph_context(data));
+    }
+
+    PS_ALWAYS_INLINE graph_context graph() thread
+    {
+        return graph_context(emitter::api::graph_context(data));
+    }
+
+    thread emitter::context_t &data;
+};
+
+} // namespace computegraph
+
+#endif
+
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphGroupContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphGroupContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphGroupContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphGroupContext.h	2026-05-29 23:12:39
@@ -0,0 +1,89 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+#include <metal_types>
+
+struct group_context_t;
+
+namespace group
+{
+struct context_t;
+
+namespace api
+{
+PS_API uint group_index_in_system(thread group_context_t &);
+PS_API uint group_maximum_in_system(thread group_context_t &);
+
+PS_API uint simdgroup_index_in_group(thread group_context_t &);
+PS_API uint simdgroups_per_group(thread group_context_t &);
+
+PS_API uint element_index_in_group(thread group_context_t &);
+PS_API uint element_active_count_in_group(thread group_context_t &);
+PS_API uint element_maximum_in_group(thread group_context_t &);
+}
+}
+
+namespace computegraph
+{
+/// A context for elements that are grouped together, either as a strip
+/// (like a ribbon) or as an unstructured group.
+struct group_context
+{
+    PS_ALWAYS_INLINE group_context(thread group_context_t &ctx)
+    : _context(ctx)
+    {}
+
+    /// Index of the current group within the system overall
+    PS_ALWAYS_INLINE ushort group_index_in_system() const
+    {
+        return group::api::group_index_in_system(_context);
+    }
+
+    /// The current simdgroup's index within the element group or ribbon
+    PS_ALWAYS_INLINE ushort simdgroup_index_in_group() const
+    {
+        return group::api::simdgroup_index_in_group(_context);
+    }
+
+    /// Returns the number of simdgroups that execute per element group or ribbon.
+    PS_ALWAYS_INLINE ushort simdgroups_per_group() const
+    {
+        return group::api::simdgroups_per_group(_context);
+    }
+
+    /// Returns the index of the current element within the current group.
+    PS_ALWAYS_INLINE uint element_index_in_group() const
+    {
+        return group::api::element_index_in_group(_context);
+    }
+
+    /// Number of elements active in the current group or ribbon.
+    PS_ALWAYS_INLINE uint active_element_count() const
+    {
+        return group::api::element_active_count_in_group(_context);
+    }
+
+    /// Maximum number of elements that can be active within a group.
+    ///
+    /// This value is configured in the simulation and is constant for the lifetime of the graph.
+    PS_ALWAYS_INLINE uint element_maximum_in_group() const
+    {
+        return group::api::element_maximum_in_group(_context);
+    }
+
+    /// Maximum number of groups that can exist in the system.
+    ///
+    /// The number of active groups will always be less than or equal to this number
+    PS_ALWAYS_INLINE uint maximum_groups_in_system() const
+    {
+        return group::api::group_maximum_in_system(_context);
+    }
+private:
+    thread group_context_t &_context;
+};
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphInitializationContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphInitializationContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphInitializationContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphInitializationContext.h	2026-05-29 23:12:39
@@ -0,0 +1,85 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+#include <ComputeGraph/ComputeGraphElement.h>
+#include <ComputeGraph/ComputeGraphElementContext.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+
+#include <metal_types>
+
+namespace initialize
+{
+struct context_t;
+
+namespace api
+{
+PS_API thread graph_context_t &graph_context(thread initialize::context_t &);
+PS_API thread system_element_context_t &system_element_context(thread initialize::context_t &);
+
+PS_API uint source_element_index(thread initialize::context_t &);
+PS_API uint spawn_index(thread initialize::context_t &);
+}
+}
+
+namespace computegraph
+{
+struct initialization_context
+{
+    PS_ALWAYS_INLINE initialization_context(thread initialize::context_t &_data)
+    : data(_data)
+    {}
+
+    PS_ALWAYS_INLINE initialization_context(const thread initialization_context &_context)
+    : data(_context.data)
+    {}
+
+    PS_ALWAYS_INLINE computegraph::element_info element() const
+    {
+        return computegraph::element_info(system_element::api::element_data(initialize::api::system_element_context(data)));
+    }
+
+    PS_ALWAYS_INLINE uint element_index() const
+    {
+        return system_element::api::element_index(initialize::api::system_element_context(data));
+    }
+
+    /// For a derived element spawned with EventSource, returns the index of the source
+    /// element. If this element was not spawned from an event source, 0 is returned.
+    PS_ALWAYS_INLINE uint source_element_index() const
+    {
+        return initialize::api::source_element_index(data);
+    }
+
+    /// For a derived element spawned from an emitter, returns the sequential index of
+    /// the spawn from that emitter. If this was spawned from an EventSource or from the CPU,
+    /// this value is the index within the current spawn request.
+    PS_ALWAYS_INLINE uint spawn_index() const
+    {
+        return initialize::api::spawn_index(data);
+    }
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return graph().random();
+    }
+
+    PS_ALWAYS_INLINE computegraph::graph_context graph() const
+    {
+        return computegraph::graph_context(initialize::api::graph_context(data));
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return graph().debug_draw();
+    }
+
+    // TODO: private
+    thread initialize::context_t &data;
+};
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphModuleTypes.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphModuleTypes.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphModuleTypes.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphModuleTypes.h	2026-05-29 23:12:39
@@ -0,0 +1,52 @@
+#pragma once
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+/// Declarations for types that are implemented by modules, but aren't
+/// needed as part of the ComputeGraph API
+
+
+/// Parameters describing mouse interaction in 3D space.
+///
+/// This structure captures both the position and direction of a mouse cursor
+/// projected into 3D coordinates, commonly used for ray casting or spatial interactions.
+struct MouseParams
+{
+    /// The 3D position of the mouse cursor in local space, converted to the
+    /// simulation's coordinate system.
+    simd_float3 position;
+    
+    /// The normalized direction vector of the mouse ray, converted to the
+    /// simulation's coordinate system.
+    simd_float3 direction;
+    
+    /// Indicates whether valid mouse parameters are available.
+    ///
+    /// When `false`, the `position` and `direction` values should not be used.
+    bool has_value;
+};
+
+/// Camera viewpoint parameters in 3D space.
+///
+/// This structure represents a camera or observer's position and viewing direction,
+/// with optional availability flags for each component. This allows partial viewpoint
+/// information to be represented when only position or direction is known.
+struct Viewpoint
+{
+    /// The 3D position of the camera or viewpoint, in the simulation's coordinate system.
+    simd_float3 position;
+    
+    /// The normalized direction vector indicating where the camera is looking,
+    /// in the simulation's coordinate system
+    simd_float3 direction;
+
+    /// Indicates whether a valid position is available.
+    ///
+    /// When `false`, the `position` value should not be used.
+    bool has_position;
+    
+    /// Indicates whether a valid direction is available.
+    ///
+    /// When `false`, the `direction` value should not be used.
+    bool has_direction;
+};
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphOutputContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphOutputContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphOutputContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphOutputContext.h	2026-05-29 23:17:16
@@ -0,0 +1,208 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+#include <ComputeGraph/ComputeGraphElementContext.h>
+
+#include <metal_types>
+
+namespace output
+{
+struct context_t;
+
+namespace api
+{
+PS_API uint output_index(thread output::context_t &);
+PS_API thread graph_context_t &graph_context(thread output::context_t &);
+PS_API thread system_element_context_t &system_element_context(thread output::context_t &);
+
+PS_API float3 position(thread output::context_t &);
+PS_API void set_position(thread output::context_t &, float3);
+
+PS_API float2 size(thread output::context_t &);
+PS_API void set_size(thread output::context_t &, float2);
+
+PS_API float2 scale(thread output::context_t &);
+PS_API void set_scale(thread output::context_t &, float2);
+
+PS_API half4 color(thread output::context_t &);
+PS_API void set_color(thread output::context_t &, half4);
+
+PS_API float2 uv_offset(thread output::context_t &);
+PS_API void set_uv_offset(thread output::context_t &, float2);
+
+PS_API float2 uv_scale(thread output::context_t &);
+PS_API void set_uv_scale(thread output::context_t &, float2);
+
+} // namespace api
+} // namespace update
+
+
+namespace computegraph
+{
+struct output_context
+{
+    PS_ALWAYS_INLINE output_context(thread output::context_t &_context,
+                                    thread ElementRenderData &__unused)
+    : context(_context)
+    , __unused(__unused)
+    {}
+
+    PS_ALWAYS_INLINE uint element_id() const
+    {
+        return system_element::api::element_index(output::api::system_element_context(context));
+    }
+
+    /// Returns true if the current context has a valid element to read.
+    PS_ALWAYS_INLINE bool is_reader() const
+    {
+        return system_element::api::is_reader_thread(output::api::system_element_context(context));
+    }
+
+    /// Returns true if the current context is the authority for its current element.
+    PS_ALWAYS_INLINE bool is_authority() const
+    {
+        return system_element::api::is_authority_thread(output::api::system_element_context(context));
+    }
+
+    /// Returns true if the current context is expected to write to its output, regardless of
+    /// whether it has an active element.
+    PS_ALWAYS_INLINE bool is_output() const
+    {
+        return system_element::api::is_output_thread(output::api::system_element_context(context));
+    }
+
+    // Returns the output index ordering.
+    //
+    // This may be different from elementId() in cases where particles are sorted.
+    PS_ALWAYS_INLINE uint output_index() const
+    {
+        return output::api::output_index(context);
+    }
+
+    PS_ALWAYS_INLINE float3 output_position() const
+    {
+        return output::api::position(context);
+    }
+    PS_ALWAYS_INLINE void set_output_position(float3 value)
+    {
+        output::api::set_position(context, value);
+    }
+
+    /// Returns the current output size.
+    ///
+    /// This begin's with the element's current value, and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE float2 output_size() const
+    {
+        return output::api::size(context);
+    }
+    PS_ALWAYS_INLINE void set_output_size(float2 value)
+    {
+        output::api::set_size(context, value);
+    }
+
+    /// Returns the current output scale.
+    ///
+    /// This begin's with the element's current value, and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE float2 output_scale() const
+    {
+        return output::api::scale(context);
+    }
+    PS_ALWAYS_INLINE void set_output_scale(float2 value)
+    {
+        output::api::set_scale(context, value);
+    }
+
+    /// Returns the current output color.
+    ///
+    /// This begin's with the element's current value, and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE half4 output_color() const
+    {
+        return output::api::color(context);
+    }
+    PS_ALWAYS_INLINE void set_output_color(half3 value)
+    {
+        output::api::set_color(context, half4(value, output_color().a));
+    }
+
+    /// Returns the current output alpha.
+    ///
+    /// This begin's with the element's current value, and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE half output_alpha() const
+    {
+        return output::api::color(context).a;
+    }
+    PS_ALWAYS_INLINE void set_output_alpha(half value)
+    {
+        output::api::set_color(context, half4(output_color().rgb, value));
+    }
+
+
+    /// Returns the current output uv offset.
+    ///
+    /// This begin's with the (0, 0), and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE float2 uv_offset() const
+    {
+        return output::api::uv_offset(context);
+    }
+    PS_ALWAYS_INLINE void set_uv_offset(float2 value)
+    {
+        output::api::set_uv_offset(context, value);
+    }
+
+    /// Returns the current output uv scale.
+    ///
+    /// This begin's with (1, 1), and can be customized during
+    /// the output stage without affecting the element.
+    PS_ALWAYS_INLINE float2 uv_scale() const
+    {
+        return output::api::uv_scale(context);
+    }
+
+    PS_ALWAYS_INLINE void set_uv_scale(float2 value)
+    {
+        output::api::set_uv_scale(context, value);
+    }
+
+    PS_ALWAYS_INLINE computegraph::graph_context graph() const
+    {
+        return computegraph::graph_context(output::api::graph_context(context));
+    }
+
+    PS_ALWAYS_INLINE computegraph::element_info element() const
+    {
+        return computegraph::element_info(system_element::api::element_data(output::api::system_element_context(context)));
+    }
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return base().random();
+    }
+
+    PS_ALWAYS_INLINE computegraph::element_context base() const
+    {
+        return computegraph::element_context(output::api::system_element_context(context));
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return base().debug_draw();
+    }
+
+    // TODO: private
+    thread output::context_t &context;
+    // Replaced with function-based interface. Temporarily has
+    // an extra member during transition
+    thread ElementRenderData &__unused;
+};
+
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphRandom.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphRandom.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphRandom.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphRandom.h	2026-05-29 23:12:39
@@ -0,0 +1,139 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <metal_stdlib>
+#include <metal_types>
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+namespace random
+{
+namespace api
+{
+PS_API uint16_t random_uint16(uint16_t seed);
+PS_API uint32_t random_uint32(uint32_t seed);
+
+PS_API half random01_half(uint16_t seed);
+PS_API float random01_float(uint32_t seed);
+
+PS_API uint32_t increment_seed(thread random_context_t &rng, uint32_t count);
+
+/// Returns the current seed without incrementing it
+PS_API uint32_t seed(const thread random_context_t &rng);
+}
+}
+namespace computegraph
+{
+struct random_generator
+{
+    PS_ALWAYS_INLINE random_generator(thread random_context_t &_context)
+    : context(_context)
+    {}
+    
+    PS_ALWAYS_INLINE uint16_t uint16()
+    {
+        return random::api::random_uint16(uint16_t(next_seed()));
+    }
+    
+    PS_ALWAYS_INLINE uint32_t uint32()
+    {
+        return random::api::random_uint32(next_seed());
+    }
+    
+    PS_ALWAYS_INLINE float float_01()
+    {
+        return random::api::random01_float(next_seed());
+    }
+    
+    PS_ALWAYS_INLINE float2 float2_01()
+    {
+        uint32_t seed = next_seed(2);
+        
+        return {
+            random::api::random01_float(seed),
+            random::api::random01_float(seed+1)
+        };
+    }
+    
+    PS_ALWAYS_INLINE float3 float3_01()
+    {
+        uint32_t seed = next_seed(3);
+        
+        return {
+            random::api::random01_float(seed),
+            random::api::random01_float(seed+1),
+            random::api::random01_float(seed+2)
+        };
+    }
+    
+    PS_ALWAYS_INLINE float4 float4_01()
+    {
+        uint32_t seed = next_seed(4);
+        
+        return {
+            random::api::random01_float(seed),
+            random::api::random01_float(seed+1),
+            random::api::random01_float(seed+2),
+            random::api::random01_float(seed+3)
+        };
+    }
+    
+    PS_ALWAYS_INLINE half half_01()
+    {
+        uint32_t seed = next_seed(1);
+        
+        return random::api::random01_half(seed);
+    }
+    
+    PS_ALWAYS_INLINE half2 half2_01()
+    {
+        uint32_t seed = next_seed(2);
+        
+        return {
+            random::api::random01_half(seed),
+            random::api::random01_half(seed+1),
+        };
+    }
+    
+    PS_ALWAYS_INLINE half3 half3_01()
+    {
+        uint32_t seed = next_seed(3);
+        
+        return {
+            random::api::random01_half(seed),
+            random::api::random01_half(seed+1),
+            random::api::random01_half(seed+2),
+        };
+    }
+    
+    PS_ALWAYS_INLINE half4 half4_01()
+    {
+        uint32_t seed = next_seed(4);
+        
+        return {
+            random::api::random01_half(seed),
+            random::api::random01_half(seed+1),
+            random::api::random01_half(seed+2),
+            random::api::random01_half(seed+3)
+        };
+    }
+    
+    /// Returns the current seed and increments it `count` times.
+    PS_ALWAYS_INLINE uint32_t next_seed(uint32_t count = 1)
+    {
+        return random::api::increment_seed(context, count);
+    }
+    
+    /// Returns the current seed without incrementing it
+    PS_ALWAYS_INLINE uint32_t seed() const
+    {
+        return random::api::seed(context);
+    }
+private:
+    thread random_context_t &context;
+};
+
+} // namespace computegraph
+
+#endif // __METAL_VERSION__
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphSimulationContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphSimulationContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphSimulationContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphSimulationContext.h	2026-05-29 23:12:39
@@ -0,0 +1,102 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphRandom.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+#include <ComputeGraph/ComputeGraphElement.h>
+#include <ComputeGraph/ComputeGraphElementContext.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+
+#include <metal_types>
+
+namespace simulate
+{
+struct context_t;
+
+namespace api
+{
+PS_API thread float3 &force(thread simulate::context_t &);
+PS_API thread graph_context_t &graph_context(thread simulate::context_t &);
+PS_API thread system_element_context_t &system_element_context(thread simulate::context_t &);
+} // namespace api
+} // namespace update
+
+namespace computegraph
+{
+struct simulation_context
+{
+    PS_ALWAYS_INLINE simulation_context(thread simulate::context_t &_data)
+    : data(_data)
+    {}
+
+    PS_ALWAYS_INLINE simulation_context(const thread simulation_context &_other)
+    : data(_other.data)
+    {}
+
+    PS_ALWAYS_INLINE uint element_index() const
+    {
+        return system_element::api::element_index(system_element_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::element_info element() const
+    {
+        return computegraph::element_info(system_element::api::element_data(system_element_context()));
+    }
+
+    PS_ALWAYS_INLINE computegraph::random_generator random() thread
+    {
+        return graph().random();
+    }
+
+    PS_ALWAYS_INLINE computegraph::global_textures textures() const
+    {
+        return computegraph::global_textures(graph_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::graph_context graph() const
+    {
+        return computegraph::graph_context(graph_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::element_context base() const
+    {
+        return computegraph::element_context(system_element_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return graph().debug_draw();
+    }
+
+    PS_ALWAYS_INLINE void terminate()
+    {
+        system_element::api::terminate(system_element_context());
+    }
+
+    PS_ALWAYS_INLINE float delta_time() const
+    {
+        return graph::api::delta_time(graph_context());
+    }
+
+    PS_ALWAYS_INLINE float3 force() const
+    {
+        return simulate::api::force(data);
+    }
+
+    PS_ALWAYS_INLINE void set_force(float3 force)
+    {
+        simulate::api::force(data) = force;
+    }
+
+    // TODO: make me private
+    thread simulate::context_t &data;
+private:
+    PS_ALWAYS_INLINE thread graph_context_t &graph_context() const { return simulate::api::graph_context(data); }
+    PS_ALWAYS_INLINE thread system_element_context_t &system_element_context() const { return simulate::api::system_element_context(data); }
+};
+
+} // namespace computegraph
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextureContext.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextureContext.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextureContext.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextureContext.h	2026-05-29 23:12:39
@@ -0,0 +1,59 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+#include <ComputeGraph/ComputeGraphContext.h>
+#include <ComputeGraph/ComputeGraphDebugDraw.h>
+
+#include <metal_types>
+
+namespace texture
+{
+struct context_t;
+
+namespace api
+{
+PS_API metal::texture2d<float, metal::access::read_write> texture(thread texture::context_t &);
+PS_API thread graph_context_t &graph_context(thread texture::context_t &);
+}
+}
+
+namespace computegraph
+{
+
+/// A context for creating a texture.
+struct texture_context
+{
+    PS_ALWAYS_INLINE texture_context(thread texture::context_t &_data)
+    : data(_data)
+    {}
+
+    PS_ALWAYS_INLINE texture_context(const thread texture_context &_other)
+    : data(_other.data)
+    {}
+
+    PS_ALWAYS_INLINE computegraph::graph_context graph() const
+    {
+        return computegraph::graph_context(graph_context());
+    }
+
+    PS_ALWAYS_INLINE computegraph::debug_draw debug_draw() const
+    {
+        return computegraph::debug_draw(graph_context());
+    }
+
+    PS_ALWAYS_INLINE metal::texture2d<float, metal::access::read_write> texture() const
+    {
+        return texture::api::texture(data);
+    }
+
+    // TODO: private
+    thread texture::context_t &data;
+private:
+    PS_ALWAYS_INLINE thread graph_context_t &graph_context() const { return texture::api::graph_context(data); }
+};
+} // namespace computegraph
+
+#endif
+
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextures.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextures.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextures.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTextures.h	2026-05-29 23:12:39
@@ -0,0 +1,58 @@
+#pragma once
+
+#if defined(__METAL_VERSION__)
+
+#include <ComputeGraph/ComputeGraphTypes.h>
+
+#include <metal_types>
+#include <metal_texture>
+
+struct graph_context_t;
+
+namespace textures
+{
+namespace api
+{
+PS_API metal::texture1d<float> noise1d(thread graph_context_t &);
+PS_API metal::texture2d<float> noise2d(thread graph_context_t &);
+PS_API metal::texture3d<float> noise3d(thread graph_context_t &);
+} // namespace api
+} // namespace textures
+
+namespace computegraph
+{
+
+// TODO: replace me with shared_uniform texture bindings
+struct global_textures
+{
+    PS_ALWAYS_INLINE global_textures(thread graph_context_t &_data)
+        : data(_data)
+    {
+    }
+
+    PS_ALWAYS_INLINE global_textures(const thread global_textures &other)
+        : data(other.data)
+    {
+    }
+
+    PS_ALWAYS_INLINE metal::texture1d<float> noise1d() const thread
+    {
+        return textures::api::noise1d(data);
+    }
+
+    PS_ALWAYS_INLINE metal::texture2d<float> noise2d() const thread
+    {
+        return textures::api::noise2d(data);
+    }
+
+    PS_ALWAYS_INLINE metal::texture3d<float> noise3d() const thread
+    {
+        return textures::api::noise3d(data);
+    }
+private:
+    thread graph_context_t &data;
+};
+} // namespace computegraph
+
+
+#endif
diff -ruN /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTypes.h /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTypes.h
--- /Applications/Xcode_26.5.0.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTypes.h	1969-12-31 19:00:00
+++ /Applications/Xcode_27.0.0-beta.app/Contents/Developer/Platforms/iPhoneOS.platform/Developer/SDKs/iPhoneOS.sdk/System/Library/Frameworks/ComputeGraph.framework/Headers/ComputeGraphTypes.h	2026-05-29 23:18:03
@@ -0,0 +1,481 @@
+#pragma once
+
+#include <simd/simd.h>
+
+#if defined(__METAL_VERSION__)
+#   include <metal_types>
+#   include <metal_logging>
+#   define PS_CONSTANT constant
+#   define PS_THREAD thread
+#   define PS_THREADGROUP threadgroup
+#   define PS_DEVICE device
+#   define ps_binding_type int32_t
+#else
+#   include <sys/types.h>
+#   include <stdbool.h>
+#   include <stddef.h>
+#   define PS_CONSTANT
+#   define PS_THREAD
+#   define PS_THREADGROUP
+#   define PS_DEVICE
+#   define ps_binding_type intptr_t
+#endif
+
+#define PS_AVAILABILE(X) __attribute__((availability(x)))
+#define PS_DEPRECATED(M...) __attribute__((deprecated(M)))
+#define PS_ALWAYS_INLINE inline __attribute__ ((always_inline))
+#define PS_ENUM(type) enum __attribute__((enum_extensibility(closed))) : type
+
+// Explicit annotation for public/exported
+#define PS_API [[visible]]
+
+// Explicit annotation of internal functions
+#define PS_INTERNAL
+
+// MARK: - Element structures
+
+#if defined(__METAL_VERSION__)
+
+struct graph_context_t;
+struct system_context_t;
+struct debug_draw_context_t;
+struct random_context_t;
+
+namespace emitter
+{
+struct context_t;
+}
+
+namespace initialize
+{
+struct context_t;
+}
+
+namespace simulate
+{
+struct context_t;
+}
+
+namespace output
+{
+struct context_t;
+}
+
+namespace compute
+{
+struct context_t;
+}
+
+namespace texture
+{
+struct context_t;
+}
+
+// TODO: shorter name. "access" would be ideal, but it conflicts with metal::access.
+enum class value_access
+{
+    read,
+    write,
+    read_write
+};
+
+constexpr bool is_read_only(value_access access)
+{
+    return access == value_access::read;
+}
+
+constexpr bool is_writable(value_access access)
+{
+    return access == value_access::write || access == value_access::read_write;
+}
+
+/// A uniform value which is global to *all nodes* binding a value of the same type.
+///
+/// @discussion Globals are matched by *type only* and are intended for use with
+/// compound structure types.
+template<typename T>
+struct shared_uniform
+{
+    constant T &value;
+
+    constant T *operator->() thread {
+        return &value;
+    }
+
+    operator T() const thread {
+        return value;
+    }
+};
+
+/// Per-emitter state which is read+write from the emitter phase and unavailable from later phases.
+///
+/// @discussion System State variables are matched by both type and name, so multiple
+/// attributes with the same type may be specified by using different names.
+template<typename T>
+struct emitter_state
+{
+    device T &value;
+
+    device T *operator->() thread {
+        return &value;
+    }
+
+    operator T() const thread {
+        return value;
+    }
+};
+
+/// Per-group state which is read+write during the simulation state.
+///
+/// @discussion Group State variables are matched by both type and name, so multiple
+/// attributes with the same type may be specified by using different names.
+template<typename T>
+struct group_state
+{
+    device T &value;
+
+    device T *operator->() thread {
+        return &value;
+    }
+
+    operator T() const thread {
+        return value;
+    }
+};
+
+/// Threadgroup memory binding with type and size parameters. Use ``group_temporary`` for a
+/// more convenient alternative.
+template<typename T, size_t N>
+struct threadgroup_memory
+{
+    threadgroup T &data;
+
+    operator threadgroup T &() {
+        return data;
+    }
+};
+
+template<typename T>
+using group_temporary = threadgroup_memory<T, sizeof(T)>;
+
+/// A per-vertex attribute which can be written from the output stage.
+///
+/// The ``output_attribute`` attribute allows accessing the entire buffer's contents.
+/// Use the OutputContext's``context.range`` to index into this attribute.
+///
+/// @discussion In order for the attribute to be usable from a Vertex shader,
+/// its name must match an existing semantic. See ``LowLevelMesh.VertexSemantic``.
+template<typename T, bool required = true>
+struct output_attribute
+{
+    struct attribute_ref
+    {
+        const bool has_value;
+        device T* value;
+
+        T operator=(T newValue) thread {
+            if (has_value) {
+                *value = newValue;
+            }
+            return newValue;
+        }
+
+        operator T() const thread {
+            if (has_value) {
+                return *value;
+            } else {
+                return {};
+            }
+        }
+    };
+
+    const bool _has_value;
+    const int _stride;
+    const int _count;
+    device T* value;
+
+    bool has_value() const thread {
+        return _has_value;
+    }
+
+    attribute_ref operator [](size_t index) {
+        if (_has_value) {
+            device uint8_t *bytes = (device uint8_t *)value;
+
+            return attribute_ref {
+                .has_value = true,
+                .value = (device T *)(bytes + _stride * index)
+            };
+        } else {
+            return {
+                .has_value = false,
+            };
+        }
+    }
+};
+
+/// A per-element attribute which can be read/written.
+///
+/// @discussion Attributes are matched by both type and name, so multiple attributes
+/// of the same type may be specified by using different names.
+///
+/// "Built-in" attributes can also be accessed using attribute, for example
+/// `element_state<float3> velocity`.
+template<typename T, value_access a = value_access::read, bool required = true, typename _Enable = void>
+struct element_state
+{
+    bool _has_value;
+    const size_t _offset;
+    thread T* value;
+};
+
+template<typename T, value_access a>
+struct element_state<T, a, true, typename metal::enable_if_t<is_read_only(a)> >
+{
+    const bool _has_value = true;
+    const size_t _offset;
+    const thread T* value;
+
+    bool has_value() const thread {
+        return true;
+    }
+
+    const thread T &operator->() thread {
+        return value;
+    }
+
+    operator T() const thread {
+        return *value;
+    }
+
+    /// Conversion operator from a required to optional attribute
+    operator element_state<T, a, false, typename metal::enable_if_t<is_read_only(a)> > () const thread {
+        return {true, value};
+    }
+};
+
+template<typename T, value_access a>
+struct element_state<T, a, true, typename metal::enable_if_t<is_writable(a)> >
+{
+    const bool _has_value = true;
+    const size_t _offset;
+    thread T* value;
+
+    constexpr bool has_value() const thread {
+        return true;
+    }
+
+    thread T &operator->() thread {
+        return value;
+    }
+
+    void operator=(T newValue) thread {
+        *value = newValue;
+    }
+
+    operator T() const thread {
+        return *value;
+    }
+};
+
+template<typename T, value_access a>
+struct element_state<T, a, false, typename metal::enable_if_t<is_read_only(a)> >
+{
+    const bool _has_value;
+    const size_t _offset;
+    const thread T* value;
+
+    bool has_value() const thread {
+        return _has_value;
+    }
+
+    T operator->() thread {
+        if (_has_value) {
+            return *value;
+        } else {
+            return {};
+        }
+    }
+
+    operator T() const thread {
+        if (_has_value) {
+            return *value;
+        } else {
+            return {};
+        }
+    }
+};
+
+template<typename T, value_access a>
+struct element_state<T, a, false, typename metal::enable_if_t<is_writable(a)> >
+{
+    const bool _has_value;
+    const size_t _offset;
+    thread T* value;
+
+    bool has_value() const thread {
+        return _has_value;
+    }
+
+    void operator=(T newValue) thread {
+        if (_has_value) {
+            *value = newValue;
+        }
+    }
+
+    operator T() const thread {
+        if (_has_value) {
+            return *value;
+        } else {
+            return {};
+        }
+    }
+};
+
+/// A typed buffer of unknown size
+template<typename T>
+struct pointer
+{
+    device T *value;
+    uint3 __unused;
+
+    device T *operator->() thread {
+        return value;
+    }
+};
+
+template<typename T>
+struct buffer1d
+{
+    device T *values;
+    uint3 count;
+
+    T read(uint x) thread {
+        if (x < count.x) {
+            return values[x];
+        } else {
+            return {};
+        }
+    }
+};
+
+template<typename T>
+struct buffer2d
+{
+    device T *values;
+    uint3 count;
+
+    T read(uint2 pos) thread {
+        return read(pos.x, pos.y);
+    }
+
+    T read(uint x, uint y) thread {
+        if (x < count.x && y < count.y) {
+            return values[x + y * count.x];
+        } else {
+            return {};
+        }
+    }
+};
+
+template<typename T>
+struct buffer3d
+{
+    device T *values;
+    uint3 count;
+
+    T read(uint3 pos) thread {
+        return read(pos.x, pos.y, pos.z);
+    }
+
+    T read(uint x, uint y, uint z) thread {
+        if (x < count.x && y < count.y && z < count.z) {
+            return values[x + y * count.x + z * (count.x * count.y)];
+        } else {
+            return {};
+        }
+    }
+};
+
+/// Optional buffer with stride and count. Used internally during compilation.
+template<typename T>
+struct strided_buffer
+{
+    const bool _has_value;
+    const int _stride;
+    const int count;
+    device T* value;
+
+    struct value_ref
+    {
+        const bool has_value;
+        device T* value;
+
+        PS_ALWAYS_INLINE T operator=(const thread value_ref &newValue) thread {
+            return (*this) = (T)newValue;
+        }
+
+        PS_ALWAYS_INLINE T operator=(T newValue) thread {
+            if (has_value) {
+                *value = newValue;
+            }
+            return newValue;
+        }
+
+        PS_ALWAYS_INLINE operator T() const thread {
+            if (has_value) {
+                return *value;
+            } else {
+                return {};
+            }
+        }
+    };
+
+    PS_ALWAYS_INLINE value_ref operator [] (size_t index) const {
+        if (_has_value && index < (size_t)count) {
+            device uint8_t *bytes = (device uint8_t *)value;
+
+            return value_ref {
+                .has_value = true,
+                .value = (device T *)(bytes + _stride * index)
+            };
+        } else {
+            return {
+                .has_value = false,
+            };
+        }
+    }
+
+    explicit operator bool () const { return _has_value; }
+};
+
+struct ElementRenderData;
+
+// Mark - Built-ins
+
+struct thread_position_in_grid {
+    uint3 value;
+
+    operator uint() const { return value.x; }
+    operator uint2() const { return value.xy; }
+    operator uint3() const { return value; }
+};
+
+struct threads_per_grid {
+    uint3 value;
+
+    operator uint() const { return value.x; }
+    operator uint2() const { return value.xy; }
+    operator uint3() const { return value; }
+};
+
+struct threads_per_threadgroup {
+    uint3 value;
+
+    operator uint() const { return value.x; }
+    operator uint2() const { return value.xy; }
+    operator uint3() const { return value; }
+};
+
+#endif // defined(__METAL_VERSION__)
+

Clone this wiki locally