Struct MetalLibrary

Source

pub struct MetalLibrary { /* private fields */ }

Expand description

Apple’s id<MTLLibrary> — compiled MSL source.

Implementations§

Source §

impl MetalLibrary

Source

pub fn new_function(&self, name: &str) -> Option<MetalFunction>

Look up a kernel function by its source name.

Examples found in repository ?

examples/04_compute_shader.rs (line 30)

21fn main() {
22    let device = MetalDevice::system_default().expect("MTLCreateSystemDefaultDevice");
23    println!("Device unified={}", device.has_unified_memory());
24
25    let lib = device
26        .new_library_with_source(KERNEL_SRC)
27        .expect("compile MSL source");
28    println!("✅ Compiled library {:p}", lib.as_ptr());
29
30    let func = lib.new_function("mul2").expect("locate function 'mul2'");
31    println!("✅ Found function mul2 {:p}", func.as_ptr());
32
33    let pso = device
34        .new_compute_pipeline_state(&func)
35        .expect("build compute pipeline state");
36    println!("✅ Compute pipeline state {:p}", pso.as_ptr());
37
38    let byte_len = N * core::mem::size_of::<f32>();
39    let buffer = device
40        .new_buffer(byte_len, resource_options::STORAGE_MODE_SHARED)
41        .expect("allocate buffer");
42
43    {
44        let mut mapping = unsafe { buffer.map_write().expect("map compute input") };
45        for (i, bytes) in mapping.chunks_exact_mut(4).take(N).enumerate() {
46            bytes.copy_from_slice(&(i as f32).to_ne_bytes());
47        }
48        let input: Vec<f32> = mapping
49            .chunks_exact(4)
50            .take(N)
51            .map(|bytes| f32::from_ne_bytes(bytes.try_into().expect("four-byte float")))
52            .collect();
53        drop(mapping);
54        println!("Input : {input:?}");
55    }
56
57    let queue = device.new_command_queue().expect("MTLCommandQueue");
58    let cb = queue.new_command_buffer().expect("MTLCommandBuffer");
59    cb.dispatch_compute_1d(&pso, &[&buffer], N, 1)
60        .expect("dispatch compute");
61    cb.commit().expect("commit compute");
62    cb.wait_until_completed().expect("complete compute");
63
64    let mapping = unsafe { buffer.map_read().expect("map compute output") };
65    let output: Vec<f32> = mapping
66        .chunks_exact(4)
67        .take(N)
68        .map(|bytes| f32::from_ne_bytes(bytes.try_into().expect("four-byte float")))
69        .collect();
70    drop(mapping);
71    println!("Output: {output:?}");
72
73    for (i, &v) in output.iter().enumerate() {
74        let expected = (i as f32) * 2.0;
75        assert_eq!(v, expected, "element {i} expected {expected} got {v}");
76    }
77    println!("✅ All {N} elements correctly doubled by the GPU kernel");
78}

More examples

Hide additional examples

examples/05_render_and_explicit_encoders.rs (line 55)

11fn main() {
12    let device = MetalDevice::system_default().expect("Metal device available");
13    println!(
14        "device: {} (registry id {})",
15        device.name(),
16        device.registry_id()
17    );
18
19    let queue = device.new_command_queue().expect("command queue");
20    let status_buffer = unsafe {
21        queue
22            .new_command_buffer_with_unretained_references()
23            .expect("scratch command buffer")
24    };
25    println!("scratch command buffer status={}", status_buffer.status());
26
27    let src = device
28        .new_buffer(64, resource_options::STORAGE_MODE_SHARED)
29        .expect("source buffer");
30    let dst = device
31        .new_buffer(64, resource_options::STORAGE_MODE_SHARED)
32        .expect("destination buffer");
33    let blit_cb = queue.new_command_buffer().expect("blit command buffer");
34    let mut blit = blit_cb
35        .new_blit_command_encoder()
36        .expect("blit command encoder");
37    blit.fill_buffer(&src, 0..64, b'Z').expect("fill source");
38    blit.copy_buffer(&src, 0, &dst, 0, 64)
39        .expect("copy buffers");
40    blit.end_encoding().expect("end blit encoder");
41    blit_cb.commit().expect("commit blit");
42    blit_cb.wait_until_completed().expect("complete blit");
43    let copied = {
44        let mapping = unsafe { dst.map_read().expect("map destination") };
45        let copied = mapping[..8].to_vec();
46        drop(mapping);
47        copied
48    };
49    println!("blit copied bytes: {copied:?}");
50
51    let library = device
52        .new_library_with_source(common::COMPUTE_SRC)
53        .expect("compile compute library");
54    let increment = library
55        .new_function("increment")
56        .expect("increment function");
57    let pipeline = device
58        .new_compute_pipeline_state(&increment)
59        .expect("compute pipeline");
60
61    let buffer = device
62        .new_buffer(16, resource_options::STORAGE_MODE_SHARED)
63        .expect("compute buffer");
64    common::write_u32_words(&buffer, &[10, 20, 30, 40]);
65    let compute_cb = queue.new_command_buffer().expect("compute command buffer");
66    let mut compute = compute_cb
67        .new_compute_command_encoder()
68        .expect("compute command encoder");
69    compute
70        .set_compute_pipeline_state(&pipeline)
71        .expect("bind compute pipeline");
72    compute
73        .set_buffer(&buffer, 0, 0)
74        .expect("bind compute buffer");
75    compute
76        .dispatch_threads((4, 1, 1), (1, 1, 1))
77        .expect("dispatch compute");
78    compute.end_encoding().expect("end compute encoder");
79    compute_cb.commit().expect("commit compute");
80    compute_cb.wait_until_completed().expect("complete compute");
81    println!("compute output: {:?}", common::read_u32_words(&buffer, 4));
82
83    let render_library = device
84        .new_library_with_source(common::RENDER_SRC)
85        .expect("compile render library");
86    let vertex = render_library
87        .new_function("fullscreen_vertex")
88        .expect("vertex function");
89    let fragment = render_library
90        .new_function("solid_fragment")
91        .expect("fragment function");
92    let render_pipeline = device
93        .new_render_pipeline_state(&vertex, &fragment, pixel_format::BGRA8UNORM, 1)
94        .expect("render pipeline");
95    println!("render pipeline label: {:?}", render_pipeline.label());
96
97    let render_target = device
98        .new_texture(common::shared_render_target(4, 4))
99        .expect("render target");
100    let vertex_buffer = device
101        .new_buffer(16, resource_options::STORAGE_MODE_SHARED)
102        .expect("vertex buffer");
103    let render_cb = queue.new_command_buffer().expect("render command buffer");
104    let mut render = render_cb
105        .new_render_command_encoder(
106            &render_target,
107            load_action::CLEAR,
108            store_action::STORE,
109            [0.0, 0.0, 0.0, 1.0],
110        )
111        .expect("render command encoder");
112    render
113        .set_render_pipeline_state(&render_pipeline)
114        .expect("bind render pipeline");
115    render
116        .set_vertex_buffer(&vertex_buffer, 0, 0)
117        .expect("bind vertex buffer");
118    render
119        .draw_primitives(primitive_type::TRIANGLE, 0, 3)
120        .expect("draw triangle");
121    render.end_encoding().expect("end render encoder");
122    render_cb.commit().expect("commit render");
123    render_cb.wait_until_completed().expect("complete render");
124
125    let mut rendered = vec![0_u8; 4 * 4 * 4];
126    unsafe {
127        render_target
128            .read_bytes_2d(&mut rendered, 16, (0, 0), (4, 4), 0)
129            .expect("read render target");
130    }
131    println!("first rendered pixel: {:?}", &rendered[..4]);
132
133    let shared_texture = device
134        .new_texture(TextureDescriptor::new_2d(4, 4, pixel_format::BGRA8UNORM))
135        .expect("shared texture");
136    let upload = vec![0x22_u8; 4 * 4 * 4];
137    unsafe {
138        shared_texture
139            .replace_region_2d(&upload, 16, (0, 0), (4, 4), 0)
140            .expect("upload texture");
141    }
142    let mut download = vec![0_u8; upload.len()];
143    unsafe {
144        shared_texture
145            .read_bytes_2d(&mut download, 16, (0, 0), (4, 4), 0)
146            .expect("read texture");
147    }
148    let view = shared_texture
149        .new_view(pixel_format::BGRA8UNORM)
150        .expect("texture view");
151    println!(
152        "texture {}x{} usage={} storage_mode={} view_width={}",
153        shared_texture.width(),
154        shared_texture.height(),
155        shared_texture.usage(),
156        shared_texture.storage_mode(),
157        view.width(),
158    );
159}

examples/06_resources_and_archives.rs (line 27)

10fn main() {
11    let device = MetalDevice::system_default().expect("Metal device available");
12    println!("device: {}", device.name());
13
14    let queue = device
15        .new_command_queue_with_max_command_buffer_count(4)
16        .expect("bounded command queue");
17    let scratch = unsafe {
18        queue
19            .new_command_buffer_with_unretained_references()
20            .expect("bounded scratch command buffer")
21    };
22    println!("bounded queue scratch status={}", scratch.status());
23
24    let library = device
25        .new_library_with_source(common::COMPUTE_SRC)
26        .expect("compile compute library");
27    let args = library.new_function("use_args").expect("use_args function");
28    let mut argument_encoder = args.new_argument_encoder(0).expect("argument encoder");
29    let argument_buffer = device
30        .new_buffer(
31            argument_encoder.encoded_length(),
32            resource_options::STORAGE_MODE_SHARED,
33        )
34        .expect("argument buffer");
35    let payload = device
36        .new_buffer(16, resource_options::STORAGE_MODE_SHARED)
37        .expect("payload buffer");
38    let texture = device
39        .new_texture(TextureDescriptor::new_2d(4, 4, pixel_format::BGRA8UNORM))
40        .expect("argument texture");
41    unsafe {
42        let mut binding = argument_encoder
43            .bind_argument_buffer(&argument_buffer, 0)
44            .expect("bind argument buffer");
45        binding
46            .set_buffer_unchecked(&payload, 0, 0)
47            .expect("bind payload");
48        binding
49            .set_texture_unchecked(&texture, 1)
50            .expect("bind texture");
51    }
52    println!(
53        "argument encoder length={} alignment={}",
54        argument_encoder.encoded_length(),
55        argument_encoder.alignment(),
56    );
57
58    let backing = device
59        .new_buffer(256, resource_options::STORAGE_MODE_SHARED)
60        .expect("backing buffer");
61    let buffer_texture = backing
62        .new_texture_view_2d(pixel_format::BGRA8UNORM, 16, 4, 64, 0)
63        .expect("buffer-backed texture");
64    println!(
65        "buffer-backed texture {}x{} fmt={}",
66        buffer_texture.width(),
67        buffer_texture.height(),
68        buffer_texture.pixel_format(),
69    );
70
71    if let Some(heap) = device.new_heap(1 << 20, storage_mode::SHARED) {
72        let heap_buffer = heap
73            .new_buffer(256, resource_options::STORAGE_MODE_SHARED)
74            .expect("heap buffer");
75        let heap_texture = heap
76            .new_texture(TextureDescriptor::new_2d(4, 4, pixel_format::BGRA8UNORM))
77            .expect("heap texture");
78        println!(
79            "heap size={} used={} current={} max_available={}",
80            heap.size(),
81            heap.used_size(),
82            heap.current_allocated_size(),
83            heap.max_available_size(256),
84        );
85        println!(
86            "heap buffer len={} heap texture {}x{} purgeable={}",
87            heap_buffer.length(),
88            heap_texture.width(),
89            heap_texture.height(),
90            heap.set_purgeable_state(apple_metal::purgeable_state::KEEP_CURRENT),
91        );
92    } else {
93        println!("heaps are unavailable on this device");
94    }
95
96    match device.new_log_state(log_level::INFO, 1_024) {
97        Ok(log_state) => {
98            let _ = device
99                .new_command_queue_with_log_state(4, &log_state)
100                .expect("log-state queue");
101            println!("created queue with log state");
102        }
103        Err(error) => println!("log state unavailable on this OS: {error}"),
104    }
105
106    if device.supports_dynamic_libraries() {
107        let dynamic_path = common::artifact_path("example-dylib.metallib");
108        let dynamic_library = device
109            .new_dynamic_library_with_source(
110                common::DYNAMIC_LIB_SRC,
111                dynamic_path.to_string_lossy().as_ref(),
112            )
113            .expect("dynamic library from source");
114        dynamic_library
115            .serialize_to_file(&dynamic_path)
116            .expect("serialize dynamic library");
117        let reloaded = device
118            .load_dynamic_library(&dynamic_path)
119            .expect("reload dynamic library");
120        println!("dynamic library install name: {}", reloaded.install_name());
121
122        let render_library = device
123            .new_library_with_source(common::RENDER_SRC)
124            .expect("compile render library");
125        let vertex = render_library
126            .new_function("fullscreen_vertex")
127            .expect("vertex function");
128        let fragment = render_library
129            .new_function("solid_fragment")
130            .expect("fragment function");
131        let increment = library
132            .new_function("increment")
133            .expect("increment function");
134
135        let archive_path = common::artifact_path("example-archive.metalarc");
136        let archive = device.new_binary_archive(None).expect("binary archive");
137        archive
138            .add_compute_function(&increment)
139            .expect("archive compute pipeline");
140        archive
141            .add_render_functions(&vertex, &fragment, pixel_format::BGRA8UNORM, 1)
142            .expect("archive render pipeline");
143        archive
144            .serialize_to_file(&archive_path)
145            .expect("serialize binary archive");
146        let _ = device
147            .new_binary_archive(Some(&archive_path))
148            .expect("reload binary archive");
149        println!("binary archive written to {}", archive_path.display());
150    } else {
151        println!("dynamic libraries unsupported; skipping archive serialization");
152    }
153}

examples/07_advanced_objects.rs (line 107)

11fn main() {
12    let device = MetalDevice::system_default().expect("Metal device available");
13    let queue = device.new_command_queue().expect("command queue");
14    let counter_sets = device.counter_set_names();
15    println!("counter sets: {counter_sets:?}");
16
17    if let Some(event) = device.new_shared_event() {
18        event.set_signaled_value(1);
19        println!("event signaled value={}", event.signaled_value());
20        let signal = queue
21            .new_command_buffer()
22            .expect("event signal command buffer");
23        signal
24            .encode_signal_event(&event, 2)
25            .expect("encode event signal");
26        signal.commit().expect("commit event signal");
27        signal
28            .wait_until_completed()
29            .expect("complete event signal");
30        println!(
31            "event reached value 2: {}",
32            event.wait_until_signaled_value(2, 1_000),
33        );
34
35        let wait = queue
36            .new_command_buffer()
37            .expect("event wait command buffer");
38        wait.encode_wait_for_event(&event, 2)
39            .expect("encode event wait");
40        wait.commit().expect("commit event wait");
41        wait.wait_until_completed().expect("complete event wait");
42    }
43
44    let fence_a = device.new_fence();
45    let fence_b = device.new_fence();
46    let sample_buffer = counter_sets.first().and_then(|name| {
47        if device.supports_counter_sampling(counter_sampling_point::AT_BLIT_BOUNDARY) {
48            device
49                .new_counter_sample_buffer(name, 2, storage_mode::SHARED, Some("example-samples"))
50                .ok()
51        } else {
52            None
53        }
54    });
55
56    let src = device
57        .new_buffer(64, resource_options::STORAGE_MODE_SHARED)
58        .expect("source buffer");
59    let dst = device
60        .new_buffer(64, resource_options::STORAGE_MODE_SHARED)
61        .expect("destination buffer");
62    let blit = queue.new_command_buffer().expect("blit command buffer");
63    let mut encoder = blit.new_blit_command_encoder().expect("blit encoder");
64    encoder
65        .fill_buffer(&src, 0..64, b'Q')
66        .expect("fill source buffer");
67    if let Some(fence) = fence_a.as_ref() {
68        encoder.update_fence(fence).expect("update fence");
69    }
70    encoder.end_encoding().expect("end first blit encoder");
71    blit.commit().expect("commit first blit");
72    blit.wait_until_completed().expect("complete first blit");
73
74    let blit = queue
75        .new_command_buffer()
76        .expect("second blit command buffer");
77    let mut encoder = blit
78        .new_blit_command_encoder()
79        .expect("second blit encoder");
80    if let Some(fence) = fence_a.as_ref() {
81        encoder.wait_for_fence(fence).expect("wait for fence");
82    }
83    if let Some(sample_buffer) = sample_buffer.as_ref() {
84        encoder
85            .sample_counters(sample_buffer, 0, false)
86            .expect("sample counters");
87    }
88    encoder
89        .copy_buffer(&src, 0, &dst, 0, 64)
90        .expect("copy buffers");
91    encoder.end_encoding().expect("end second blit encoder");
92    blit.commit().expect("commit second blit");
93    blit.wait_until_completed().expect("complete second blit");
94    if let Some(sample_buffer) = sample_buffer.as_ref() {
95        println!(
96            "resolved counter bytes={}",
97            sample_buffer
98                .resolve_range(0..1)
99                .map_or(0, |bytes| bytes.len())
100        );
101    }
102
103    let library = device
104        .new_library_with_source(common::COMPUTE_SRC)
105        .expect("compile compute library");
106    let increment = library
107        .new_function("increment")
108        .expect("increment function");
109    let pipeline = device
110        .new_compute_pipeline_state(&increment)
111        .expect("compute pipeline");
112    let visible_table = pipeline.new_visible_function_table(1);
113    let intersection_table = if device.supports_raytracing() {
114        pipeline.new_intersection_function_table(1)
115    } else {
116        None
117    };
118    if let Some(table) = intersection_table.as_ref() {
119        table.set_opaque_triangle_intersection_function(intersection_function_signature::NONE, 0);
120    }
121    let acceleration_structure = if device.supports_raytracing() {
122        device.new_acceleration_structure_with_size(256)
123    } else {
124        None
125    };
126
127    let buffer = device
128        .new_buffer(16, resource_options::STORAGE_MODE_SHARED)
129        .expect("compute buffer");
130    common::write_u32_words(&buffer, &[1, 2, 3, 4]);
131    let texture = device
132        .new_texture(apple_metal::TextureDescriptor::new_2d(
133            4,
134            4,
135            apple_metal::pixel_format::BGRA8UNORM,
136        ))
137        .expect("compute texture");
138    let compute = queue.new_command_buffer().expect("compute command buffer");
139    let mut encoder = compute
140        .new_compute_command_encoder()
141        .expect("compute command encoder");
142    encoder
143        .set_compute_pipeline_state(&pipeline)
144        .expect("bind compute pipeline");
145    encoder
146        .set_buffer(&buffer, 0, 0)
147        .expect("bind compute buffer");
148    encoder
149        .set_texture(&texture, 1)
150        .expect("bind compute texture");
151    if let Some(fence) = fence_a.as_ref() {
152        encoder.wait_for_fence(fence).expect("wait for fence");
153    }
154    if let Some(table) = visible_table.as_ref() {
155        encoder
156            .set_visible_function_table(table, 2)
157            .expect("bind visible function table");
158    }
159    if let Some(table) = intersection_table.as_ref() {
160        encoder
161            .set_intersection_function_table(table, 3)
162            .expect("bind intersection function table");
163    }
164    if let Some(acceleration_structure) = acceleration_structure.as_ref() {
165        encoder
166            .set_acceleration_structure(acceleration_structure, 4)
167            .expect("bind acceleration structure");
168    }
169    encoder
170        .dispatch_threadgroups((1, 1, 1), (4, 1, 1))
171        .expect("dispatch compute");
172    if let Some(fence) = fence_b.as_ref() {
173        encoder.update_fence(fence).expect("update fence");
174    }
175    encoder.end_encoding().expect("end compute encoder");
176    compute.commit().expect("commit compute");
177    compute.wait_until_completed().expect("complete compute");
178    println!(
179        "compute buffer after dispatch: {:?}",
180        common::read_u32_words(&buffer, 4)
181    );
182
183    if let Some(indirect) = device.new_indirect_command_buffer(
184        indirect_command_type::CONCURRENT_DISPATCH,
185        1,
186        0,
187        0,
188        4,
189        resource_options::STORAGE_MODE_PRIVATE,
190    ) {
191        indirect.reset_range(0..1);
192        println!("indirect command buffer size={}", indirect.size());
193    }
194
195    if let Some(heap) = device.new_heap(1 << 20, storage_mode::SHARED) {
196        if let Ok(residency_set) = device.new_residency_set(Some("example-residency"), 4) {
197            let heap_buffer = heap
198                .new_buffer(256, resource_options::STORAGE_MODE_SHARED)
199                .expect("heap buffer");
200            residency_set.add_buffer(&heap_buffer);
201            residency_set.add_heap(&heap);
202            residency_set.commit();
203            residency_set.request_residency();
204            queue.add_residency_set(&residency_set);
205            queue.remove_residency_set(&residency_set);
206            residency_set.end_residency();
207            residency_set.remove_all_allocations();
208            residency_set.commit();
209            println!(
210                "residency allocation count={}",
211                residency_set.allocation_count()
212            );
213        } else {
214            println!("residency sets unavailable on this OS");
215        }
216    }
217
218    if let Some(capture_manager) = CaptureManager::shared() {
219        println!(
220            "capture supported for developer tools={} active={}",
221            capture_manager.supports_destination(capture_destination::DEVELOPER_TOOLS),
222            capture_manager.is_capturing(),
223        );
224        if let Some(scope) = capture_manager.new_capture_scope_with_device(&device) {
225            scope.begin();
226            scope.end();
227        }
228        if let Some(scope) = capture_manager.new_capture_scope_with_command_queue(&queue) {
229            scope.begin();
230            scope.end();
231        }
232    }
233}

Source

pub const fn as_ptr(&self) -> *mut c_void

Raw id<MTLLibrary> pointer.

Examples found in repository ?

examples/04_compute_shader.rs (line 28)

21fn main() {
22    let device = MetalDevice::system_default().expect("MTLCreateSystemDefaultDevice");
23    println!("Device unified={}", device.has_unified_memory());
24
25    let lib = device
26        .new_library_with_source(KERNEL_SRC)
27        .expect("compile MSL source");
28    println!("✅ Compiled library {:p}", lib.as_ptr());
29
30    let func = lib.new_function("mul2").expect("locate function 'mul2'");
31    println!("✅ Found function mul2 {:p}", func.as_ptr());
32
33    let pso = device
34        .new_compute_pipeline_state(&func)
35        .expect("build compute pipeline state");
36    println!("✅ Compute pipeline state {:p}", pso.as_ptr());
37
38    let byte_len = N * core::mem::size_of::<f32>();
39    let buffer = device
40        .new_buffer(byte_len, resource_options::STORAGE_MODE_SHARED)
41        .expect("allocate buffer");
42
43    {
44        let mut mapping = unsafe { buffer.map_write().expect("map compute input") };
45        for (i, bytes) in mapping.chunks_exact_mut(4).take(N).enumerate() {
46            bytes.copy_from_slice(&(i as f32).to_ne_bytes());
47        }
48        let input: Vec<f32> = mapping
49            .chunks_exact(4)
50            .take(N)
51            .map(|bytes| f32::from_ne_bytes(bytes.try_into().expect("four-byte float")))
52            .collect();
53        drop(mapping);
54        println!("Input : {input:?}");
55    }
56
57    let queue = device.new_command_queue().expect("MTLCommandQueue");
58    let cb = queue.new_command_buffer().expect("MTLCommandBuffer");
59    cb.dispatch_compute_1d(&pso, &[&buffer], N, 1)
60        .expect("dispatch compute");
61    cb.commit().expect("commit compute");
62    cb.wait_until_completed().expect("complete compute");
63
64    let mapping = unsafe { buffer.map_read().expect("map compute output") };
65    let output: Vec<f32> = mapping
66        .chunks_exact(4)
67        .take(N)
68        .map(|bytes| f32::from_ne_bytes(bytes.try_into().expect("four-byte float")))
69        .collect();
70    drop(mapping);
71    println!("Output: {output:?}");
72
73    for (i, &v) in output.iter().enumerate() {
74        let expected = (i as f32) * 2.0;
75        assert_eq!(v, expected, "element {i} expected {expected} got {v}");
76    }
77    println!("✅ All {N} elements correctly doubled by the GPU kernel");
78}