# Residency

> USearch for C

An index lives in RAM, VRAM or on NVMe, and every lifecycle verb names its action and its residency instead of taking a tag a residency might not support:

- `init` starts empty and mutable: `usearch_dense_init_ram`, `usearch_dense_init_vram`.
- `load` copies a checkpoint in: `usearch_dense_load_ram`, `usearch_dense_load_vram`, and their `_buffer` forms.
- `view` serves one in place, read-only: `usearch_dense_view_ram`, `usearch_dense_view_nvme`.

The options name where an index lives, never who walks it.
`usearch_ram_options_t` names NUMA nodes, one shard per node, so a dual-socket box gets two graphs whose walks stay local; payload blocks claim huge pages where a pool is provisioned.
`usearch_vram_options_t` names device ordinals, one shard per device and never a replica, so an index outgrowing them fails rather than paging to the host.
Both take a `shard_capacity`, the entries one shard may hold, zero meaning the 32-bit ID space: an insertion past it opens another generation of shards on the same nodes or devices, and only the newest generation takes insertions.
`usearch_nvme_options_t` names paths: a filesystem path selects a kernel I/O engine, a transport identifier like `trtype:PCIe traddr:0000:5e:00.0` selects SPDK in a `-D USEARCH_WITH_SPDK=1` build, and a FUSE mount over a bucket reads as an ordinary path.

Memory follows residency: an index keeps its payload where its lifecycle verb put it, an executor keeps its scratch next to its threads or devices, and an arena lives where the executor that created it computes.
So outputs come back in RAM from a RAM executor and in VRAM from a VRAM executor, and moving them elsewhere is the caller's own copy.
Callers lend memory, never a way to allocate it: no allocator callback crosses the interface, because large systems hold too many kinds of memory for one ABI to describe them all.

Every dense verb is exported for every residency, and the engine answers what a residency lacks with `usearch_status_unsupported_k`:

- RAM answers every verb but `usearch_dense_view_ram` and `usearch_dense_search_stream_init`, and exact scans serve `multiply` and `attend` too.
- VRAM answers insertion, refinement, absorbing, counting, `get`, search, `multiply`, `maxsim`, attention and the serialization family, `save_buffer` included, while cursors answer `usearch_status_unsupported_k`, as [GPUs](#gpus) explains.
  NVIDIA devices answer exact scans, while AMD and Apple ones refuse them with `usearch_status_unsupported_k`.
  Apple's Metal answers attention `usearch_status_unsupported_k` too, since its kernels cannot call the caller's value lookup.
  A sequence, shape or geospatial VRAM index answers `save_buffer` with `usearch_status_unsupported_k` for now, and those modalities load from memory into RAM alone, through `load_ram_buffer`.
- NVMe answers every read verb as RAM does, `sample` and `health` included, refuses every mutating one with `usearch_status_read_only_k`, and refuses `save` and `save_buffer`, since a view already is its checkpoint.

### GPUs

A VRAM index lives on the devices its `vram` options name, one shard per device, and a VRAM executor drives it through one stream per device.
Passing your own streams lets USearch work join the rest of your pipeline, though the engine refuses borrowed streams with `usearch_status_unsupported_k` for now:

```c
#include <cuda_runtime.h>

cudaStream_t stream;
cudaStreamCreateWithFlags(&stream, cudaStreamNonBlocking);
int32_t const device = 0;
usearch_executor_t gpu = NULL;
usearch_executor_for_vram(&(usearch_executor_for_vram_t) {
    .executor = &gpu, .outcome = &outcome, .vram = {&device, 1}, .streams = (void* const*)&stream});

usearch_dense_index_t index = NULL;
usearch_dense_init_vram(&(usearch_dense_init_vram_t) {.index = &index,
                                                      .outcome = &outcome,
                                                      .metric = usearch_dense_metric_inner_k,
                                                      .quantization = usearch_scalar_bf16_k,
                                                      .dimensions = 1024,
                                                      .vram = {&device, 1}});
usearch_dense_reserve(&(usearch_dense_reserve_t) {
    .index = index, .executor = gpu, .outcome = &outcome, .capacity = 1000000});
```

The library classifies every pointer it is given through the driver, so inputs may live anywhere:
rows already on the device are read in place, pinned host rows upload directly, and only pageable host rows pass through the executor's pinned bounce buffer.
Pin the rows you upload often, and the bounce disappears:

```c
float* rows = NULL;
cudaHostAlloc((void**)&rows, 4096 * 1024 * sizeof(float), cudaHostAllocDefault);
// ... fill 4096 rows ...
usearch_dense_insert(&(usearch_dense_insert_t) {.index = index,
                                                .executor = gpu,
                                                .outcome = &outcome,
                                                .count = 4096,
                                                .keys = {row_keys, sizeof(usearch_key_t)},
                                                .vectors = {rows, 1024 * sizeof(float), 1024, usearch_scalar_f32_k}});
```

An arena lives where the executor that created it computes, so on a VRAM executor every output an arena holds is __device memory__.
Feed it straight into your next kernel, or bring it back yourself, into the layout you want.
Here the packed rows `get` returns land in a pinned host table with a wider row pitch, in one strided copy:

```c
usearch_rows_view_t rows_on_device = {0};
size_t rows_count = 0;
usearch_arena_t arena = NULL;
usearch_arena_init(&(usearch_arena_init_t) {.arena = &arena, .executor = gpu, .outcome = &outcome});
usearch_dense_get(&(usearch_dense_get_t) {.index = index,
                                          .executor = gpu,
                                          .outcome = &outcome,
                                          .key = row_keys[0],
                                          .vectors = &rows_on_device,
                                          .written = &rows_count,
                                          .arena = arena,
                                          .outputs_scalar = usearch_scalar_f32_k});

size_t const table_pitch = 1280 * sizeof(float);
float* table = NULL;
cudaHostAlloc((void**)&table, rows_count* table_pitch, cudaHostAllocDefault);
cudaMemcpy2DAsync(table, table_pitch, rows_on_device.data, rows_on_device.stride, rows_on_device.stride, rows_count,
                  cudaMemcpyDeviceToHost, stream);
cudaStreamSynchronize(stream);
usearch_arena_free(arena);
```

A search's matches land in its arena too, so on a VRAM executor they are device memory: every shard ranks its own entries and the first device merges those rankings into the arena, so on NVIDIA and AMD no match crosses through the host, while on Apple the host merges them in the unified memory both share.
A device that runs out of memory for an arena reports `usearch_status_out_of_memory_k` rather than falling back to the host, so where an output lands never depends on how full a device is.
Free a device arena once your own device work on its outputs has finished, and before the executor that made it.

`#include <usearch/dense_index.h>`

`#include <usearch/geospatial_index.h>`

`#include <usearch/sequence_index.h>`

`#include <usearch/shapes_index.h>`

`#include <usearch/types.h>`

## usearch_gpu_vendor_t

The GPU vendors one binary drives at once, each through a runtime loaded on first use.

```c
enum usearch_gpu_vendor_t
```

- `usearch_gpu_vendor_unknown_k` = `0`
- `usearch_gpu_vendor_nvidia_k`: NVIDIA, through the CUDA driver when one is installed.
- `usearch_gpu_vendor_amd_k`: AMD, through the HIP runtime when one is installed.
- `usearch_gpu_vendor_apple_k`: Apple, through Metal, present on every Apple silicon Mac.

## usearch_gpu_device_t

One GPU as USearch numbers it: its position in [`usearch_gpu_devices`](#usearch_gpu_devices) is the USearch device ordinal every VRAM call names.

```c
struct usearch_gpu_device_t
```

- `vendor` `usearch_gpu_vendor_t`: The vendor whose runtime drives the device.
- `vendor_ordinal` `uint32_t`: The device's ordinal in its vendor's runtime, as `nvidia-smi` or `rocm-smi` list it.
- `gpu_class` `usearch_gpu_class_t`: The class whose kernels run on it.
- `compute_units` `uint32_t`: Streaming multiprocessors, compute units or GPU cores.
- `memory_bytes` `uint64_t`: Device memory in bytes, or the working set Metal recommends for unified memory.
- `name` `char`: The vendor's own name for the device, NUL-terminated and truncated to fit.

## usearch_residency_t

Where an index's payload lives and which engine traverses it - the vocabulary the lifecycle verbs spell in their names, as [`usearch_dense_init_vram`](#usearch_dense_init_vram) does.

```c
enum usearch_residency_t
```

Each modality names a verb per action and residency rather than taking this as a tag, so a residency a modality lacks has no verb to call. The enumerators remain for the name tables bindings parse user text with.

- `usearch_residency_unknown_k` = `0`: No residency named, which is what `usearch_residency_from_name` reads unknown text as.
- `usearch_residency_ram_k` = `1`: Volatile-memory index traversed by CPU threads - the concurrent baseline.
- `usearch_residency_vram_k` = `2`: Accelerator-memory index traversed by GPU warps on a chosen device.
- `usearch_residency_nvme_k` = `3`: Persisted index traversed through pluggable I/O engines, sized past RAM.

Spellings: ```
char const * usearch_residency_name(usearch_residency_t residency);
```

```
usearch_residency_t usearch_residency_from_name(
    usearch_bytes_view_t name);
```

## usearch_ram_options_t

The host memory domains a RAM-resident index places its shards in - one shard per named NUMA node, so a dual-socket box names two and gets two graphs whose traversals stay local to the cores that walk them.

```c
struct usearch_ram_options_t
```

Names where the index lives, never who walks it: threads belong to the executor the verbs take. Every modality's RAM lifecycle verbs and the RAM executor take this struct.

The library's own payload blocks of 1 GiB or more claim 1 GiB huge pages while the host's 1 GiB pool covers them, and every other block base pages; 2 MiB pages are never claimed.

- `memory_domains` `int32_t const *`: NUMA node IDs to place shards on, borrowed for the duration of the call; `NULL` derives every node the process may allocate on.
- `memory_domains_count` `size_t`: Number of entries in `memory_domains`; zero derives.
- `shard_capacity` `size_t`: Entries one shard may hold before the index opens another shard generation on the same domains; zero, or anything past the 32-bit ID space, means 2^32 - 2.
- `cpu_class` `usearch_cpu_class_t`: The class to dispatch as if the CPU were of it, admitting every rung whose features it holds; [`usearch_cpu_class_unknown_k`](https://www.unum.cloud/docs/usearch/c/introspection.md#usearch_cpu_class_unknown_k) derives the CPU's own.
- `memory_limit` `size_t`: Bytes an NVMe view's key tables may hold on each named domain; zero sets no ceiling.

## usearch_vram_options_t

The devices a VRAM-resident index places its shards on - one shard per named ordinal, sharded and never replicated, so eight devices host eight slices of one graph rather than eight copies of it.

```c
struct usearch_vram_options_t
```

Names where the index lives, never who walks it: streams belong to the executor the verbs take, one per device. A checkpoint of more parts than named devices keeps one shard per part, part `p` on device `p` modulo their count, and shards sharing a device queue on its stream in order.

Flat matrices are device-resident, never migratable: an index outgrowing the named devices fails rather than paging to the host, and the answer is to name more devices.

One index spans one vendor and one class, so naming devices of two vendors or two generations refuses with [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k); separate indexes may sit on different vendors in one process.

- `devices` `int32_t const *`: USearch device ordinals to place shards on, NVIDIA devices first, then AMD, then Apple, as [`usearch_gpu_devices`](#usearch_gpu_devices) lists them, borrowed for the call; `NULL` picks device zero.
- `devices_count` `size_t`: Number of entries in `devices`; zero derives.
- `shard_capacity` `size_t`: Entries one shard may hold before the index opens another shard generation on the same devices, while their memory lasts; zero, or anything past the 32-bit ID space, means 2^32 - 2.
- `gpu_class` `usearch_gpu_class_t`: The class to dispatch as if every named device were of it, admitting only a class the devices run; [`usearch_gpu_class_unknown_k`](https://www.unum.cloud/docs/usearch/c/introspection.md#usearch_gpu_class_unknown_k) derives the devices' own.

## usearch_nvme_options_t

The devices an NVMe-resident index serves its shards from - one shard per named path, so striping across four drives names four.

```c
struct usearch_nvme_options_t
```

Names where the index lives, never who walks it: submission depth belongs to the executor the verbs take. Naming nothing serves every shard where its manifest already points.

A path is whatever identifies a device to the engine that can reach it, so the string is the extension point rather than a field per transport. A filesystem path names a kernel device; a transport identifier like `trtype:PCIe` `traddr:0000`:5e:00.0 names a namespace the kernel no longer owns and selects a userspace engine. A FUSE mount is just a path too, which is how object storage stages in without an HTTP stack in this library - point `mountpoint-s3` at a bucket and the ordinary engines read it.

Transfer geometry is never an input: alignment is probed from the device, and the record granularity a checkpoint was written at is read back from its own header.

Reads always bypass the page cache where the filesystem admits `O_DIRECT`.

- `paths` `usearch_bytes_view_t const *`: Devices or directories to serve shards from, borrowed for the duration of the call; `NULL` leaves every shard where its manifest points.
- `paths_count` `size_t`: Number of entries in `paths`; zero derives.
- `cpu_class` `usearch_cpu_class_t`: The class to dispatch as if the CPU were of it, admitting every rung whose features it holds; [`usearch_cpu_class_unknown_k`](https://www.unum.cloud/docs/usearch/c/introspection.md#usearch_cpu_class_unknown_k) derives the CPU's own.

## usearch_dense_init_ram

Creates an empty, mutable index in host memory, its geometry frozen, so it takes vectors without a reservation.

```c
void usearch_dense_init_ram(usearch_dense_init_ram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The metric kind used for distance calculations between vectors.
- `quantization`: The scalar type vectors are stored & quantized into.
- `dimensions`: The number of dimensions in the indexed vectors.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Upper bound on entries sharing one key; zero is unbounded.
- `ram`: The memory domains to place one shard on apiece.

Statuses: - [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) Zero `dimensions`, `dimensions` a packed quantization like [`usearch_scalar_u1x8_k`](https://www.unum.cloud/docs/usearch/c/inputs-and-outputs.md#usearch_scalar_u1x8_k) leaves a storage value partly filled by, or a `connectivity` that is not a power of two.

## usearch_dense_init_vram

Creates an empty, mutable index in device memory, its geometry frozen, so it takes vectors without a reservation.

```c
void usearch_dense_init_vram(usearch_dense_init_vram_t *arguments);
```

GPU Acceleration & Constraints

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The metric kind used for distance calculations between vectors.
- `quantization`: The scalar type vectors are stored & quantized into.
- `dimensions`: The number of dimensions in the indexed vectors.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Upper bound on entries sharing one key; zero is unbounded.
- `vram`: The devices to place one shard on apiece, by USearch ordinal.

Statuses: - [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) Zero `dimensions`, a `connectivity` that is not a power of two, or on NVIDIA and AMD a geometry whose walks cannot fit the device's shared memory.

## usearch_*_free

Frees the resources associated with the index; `NULL` is safe, and it cannot fail, so takes no outcome.

```c
void usearch_dense_free(usearch_dense_index_t index);
```

- `usearch_dense_free`: index `usearch_dense_index_t`
- `usearch_geospatial_free`: index `usearch_geospatial_index_t`
- `usearch_sequence_free`: index `usearch_sequence_index_t`
- `usearch_shapes_free`: index `usearch_shapes_index_t`

Waits until every deferred call naming the handle is final.

- `index` (inout): The handle to be freed; invalid afterwards.

## usearch_dense_reserve

Grows the shards ahead of a number of incoming vectors - a capacity hint, since insertions grow them on demand and init fixed the geometry already.

```c
void usearch_dense_reserve(usearch_dense_reserve_t *arguments);
```

- `index`: The handle to be resized.
- `executor`: The executor the growth runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.
- `capacity`: The desired total capacity, including the current size.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_uninitialized_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_uninitialized_k) No shard is placed.
- [`usearch_status_unknown_kind_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unknown_kind_k) The stored scalar's kits cannot finalize the metric.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) On VRAM, a shard's neighbors live on another device or no walk fits the device.
- [`usearch_status_overflow_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_overflow_k) `capacity` passes a shard's ceiling, its ID space, its chunk directory or its key column.

## usearch_*_clear

Erases all entries, keeping the reserved capacity and frozen geometry.

```c
void usearch_dense_clear(usearch_dense_clear_t *arguments);
```

- `usearch_dense_clear`: index `usearch_dense_index_t`
- `usearch_shapes_clear`: index `usearch_shapes_index_t`

An insertion in flight, or a refinement, absorb, compaction, split or load holding the handle, refuses it with [`usearch_status_contended_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_contended_k); a search it reaches under repeats or reports the same.

- `index`: The handle to be cleared.
- `executor`: The executor the erasure runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_contended_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_contended_k) A call in flight, a batch or load holding claims, or a plan holding the handle closed.

## usearch_geospatial_init_ram

Creates an empty, mutable geospatial index in host memory.

```c
void usearch_geospatial_init_ram(
    usearch_geospatial_init_ram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `terrain`: The __optional__ terrain lines of sight are tested over; a zeroed one means none.
- `connectivity`: The __optional__ connections-per-node bound; zero derives to twelve - eight tangent-plane cones and four long-range links.
- `values_per_key`: Upper bound on observations sharing one key; zero is unbounded.
- `ram`: The memory domains to place one shard on apiece.

Statuses: - [`usearch_status_invalid_argument_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_invalid_argument_k) A terrain both borrowing `heights` and naming a `callback`.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A terrain `path`, which no residency reads yet, or a pinned class the CPU does not admit.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) A terrain whose row or column step is not positive.

## usearch_geospatial_init_vram

Creates an empty, mutable geospatial index in device memory.

```c
void usearch_geospatial_init_vram(
    usearch_geospatial_init_vram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `terrain`: The __optional__ terrain lines of sight are tested over; a zeroed one means none.
- `connectivity`: The __optional__ connections-per-node bound; zero derives to twelve.
- `values_per_key`: Upper bound on observations sharing one key; zero is unbounded.
- `vram`: The devices to place one shard on apiece.

Statuses: - [`usearch_status_invalid_argument_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_invalid_argument_k) A terrain both borrowing `heights` and naming a `callback`.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A terrain `path` or `callback`, a `values_per_key` bound, or devices that do not run the pinned class.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) Devices of two classes, or a terrain whose row or column step is not positive.
- [`usearch_status_device_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_device_k) A device past the listing, one this build carries no kernels for, or a build without GPU kernels.

## usearch_geospatial_reserve

Grows the shards ahead of incoming entries - a capacity hint, since insertions grow them on demand.

```c
void usearch_geospatial_reserve(
    usearch_geospatial_reserve_t *arguments);
```

- `index`: The handle to be resized.
- `executor`: The executor the growth runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.
- `entries`: The desired total entry capacity, including the current size.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_overflow_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_overflow_k) `entries` passes a shard's ceiling, its ID space or its key column.

## usearch_*_clear

Erases all entries, keeping the reserved capacity.

```c
void usearch_geospatial_clear(usearch_geospatial_clear_t *arguments);
```

- `usearch_geospatial_clear`: index `usearch_geospatial_index_t`
- `usearch_sequence_clear`: index `usearch_sequence_index_t`

- `index`: The handle to be cleared.
- `executor`: The executor the erasure runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_contended_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_contended_k) A call in flight, a batch or load holding claims, or a plan holding the handle closed.

## usearch_sequence_init_ram

Creates an empty, mutable sequence index in host memory, for the uniform edit distances or the weighted alignments alike.

```c
void usearch_sequence_init_ram(usearch_sequence_init_ram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The scoring kind.
- `weights`: The weights an alignment kind needs; `NULL` for the Levenshtein kinds, which refuse any.
- `window_widths`: Window widths in bytes for the overlap kind, which needs at least one; every other kind refuses any.
- `window_widths_count`: Number of entries in `window_widths`.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Upper bound on entries sharing one key; zero is unbounded.
- `ram`: The memory domains to place one shard on apiece.

Statuses: - [`usearch_status_invalid_argument_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_invalid_argument_k) `NULL` `window_widths` for a nonzero count, a zero window width, or `weights` without a substitution matrix.
- [`usearch_status_unknown_kind_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unknown_kind_k) An unknown `metric`.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) `weights` for a kind other than Needleman-Wunsch, Needleman-Wunsch without them, window widths missing for the overlap kind or given to another, or weights folding into more than 32 classes, a negative gap or a pair above its mean self-score.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A pinned class the CPU does not admit.

## usearch_sequence_init_vram

Creates an empty, mutable sequence index in device memory.

```c
void usearch_sequence_init_vram(
    usearch_sequence_init_vram_t *arguments);
```

Insertions and searches on it run on an executor driving its device, and a search writes its matches in device memory in an arena made for that executor, or on host pages in a host one.

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The scoring kind; byte and UTF-8 Levenshtein, window overlap and Needleman-Wunsch serve a GPU.
- `weights`: The weights an alignment kind needs; `NULL` for the Levenshtein kinds, which refuse any.
- `window_widths`: Window widths in bytes for the overlap kind; every other kind refuses any.
- `window_widths_count`: Number of entries in `window_widths`.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Zero alone: a GPU index keeps no key lookup, so it bounds no key's entries and refuses any bound it is given.
- `vram`: The devices to place one shard on apiece; naming none picks device zero.

Statuses: - [`usearch_status_invalid_argument_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_invalid_argument_k) `NULL` window widths for a nonzero count, a `NULL` substitution matrix, or an empty window.
- [`usearch_status_unknown_kind_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unknown_kind_k) A kind other than byte or UTF-8 Levenshtein, window overlap or Needleman-Wunsch.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) Weights for a kind that does not align, an alignment kind without weights, window widths for any kind but overlap or none for overlap, or devices of two classes.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A `values_per_key` bound, or devices that do not run the pinned class.
- [`usearch_status_device_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_device_k) A device past the listing, one whose vendor no kernel serves, or a build without GPU kernels.

## usearch_sequence_reserve

Grows the shards ahead of incoming entries - a capacity hint, since insertions grow them on demand.

```c
void usearch_sequence_reserve(usearch_sequence_reserve_t *arguments);
```

- `index`: The handle to be resized.
- `executor`: The executor the growth runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.
- `entries`: The desired total entry capacity, including the current size.
- `payload_bytes`: Expected total payload volume; zero lets the payload storage grow lazily.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_overflow_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_overflow_k) `entries` passes a shard's ceiling, its ID space or its key column.

## usearch_shapes_init_ram

Creates an empty, mutable shape index in host memory.

```c
void usearch_shapes_init_ram(usearch_shapes_init_ram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The kind comparing two shapes - [`usearch_shapes_metric_wasserstein_procrustes_k`](https://www.unum.cloud/docs/usearch/c/metrics.md#usearch_shapes_metric_wasserstein_procrustes_k) for conformers and pockets, [`usearch_shapes_metric_chamfer_k`](https://www.unum.cloud/docs/usearch/c/metrics.md#usearch_shapes_metric_chamfer_k) where speed comes first.
- `quantization`: The scalar type points are stored in: `f16` or `f32`, any other refused with [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k).
- `points`: The frozen point count, when every entry has one - a fixed-length backbone, a landmark set.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Upper bound on entries sharing one key; zero is unbounded.
- `ram`: The memory domains to place one shard on apiece.

Statuses: - [`usearch_status_unknown_kind_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unknown_kind_k) An unknown `metric`.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A `quantization` other than `f16` or `f32`, or a pinned class the CPU does not admit.

## usearch_shapes_init_vram

Creates an empty, mutable shape index in device memory.

```c
void usearch_shapes_init_vram(usearch_shapes_init_vram_t *arguments);
```

- `index`: Output for the new index.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `metric`: The kind comparing two shapes.
- `quantization`: The scalar type points are stored in: `f16` or `f32`.
- `points`: The frozen point count, when every entry has one; zero admits any count.
- `connectivity`: The __optional__ connections-per-node bound; zero derives.
- `values_per_key`: Upper bound on entries sharing one key; zero is unbounded.
- `vram`: The devices to place one shard on apiece.

Statuses: - [`usearch_status_unknown_kind_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unknown_kind_k) An unknown `metric`.
- [`usearch_status_unsupported_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_unsupported_k) A kind other than Chamfer or Hausdorff, a `quantization` other than `f16` or `f32`, a `values_per_key` bound, or devices that do not run the pinned class.
- [`usearch_status_config_mismatch_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_config_mismatch_k) Devices of two classes.
- [`usearch_status_device_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_device_k) A device past the listing, one this build carries no kernels for, or a build without GPU kernels.

## usearch_shapes_reserve

Grows the shards ahead of incoming entries - a capacity hint, since insertions grow them on demand.

```c
void usearch_shapes_reserve(usearch_shapes_reserve_t *arguments);
```

- `index`: The handle to be resized.
- `executor`: The executor the growth runs on.
- `outcome`: Output for the outcome, final once its status leaves [`usearch_status_deferred_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_deferred_k).
- `completion`: Blocking, the default, or deferred.
- `entries`: The desired total entry capacity, including the current size.
- `points`: Expected total point volume across all entries; zero grows lazily.

Statuses: - [`usearch_status_read_only_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_read_only_k) The handle is an NVMe view.
- [`usearch_status_overflow_k`](https://www.unum.cloud/docs/usearch/c/errors.md#usearch_status_overflow_k) `entries` passes a shard's ceiling, its ID space or its key column.

## usearch_gpu_devices

Lists every GPU this process can drive, NVIDIA devices first, then AMD, then Apple, so the position of a device is the USearch ordinal `[usearch_vram_options_t](#usearch_vram_options_t)` names it by.

```c
void usearch_gpu_devices(usearch_gpu_devices_t *arguments);
```

Each vendor's runtime loads on first use, and a vendor whose runtime is absent lists no devices, so the call succeeds with a zero count on a machine without GPUs. The listing is taken once per process and never changes after.

- `outcome`: Output for the outcome, final when the call returns.
- `devices`: Output for up to `devices_capacity` devices, in USearch ordinal order; may be `NULL`.
- `devices_capacity`: Entries `devices` holds.
- `devices_count`: Output for every device this process can drive, which may exceed `devices_capacity`.

