Stream-Ordered Memory Allocator

The allocator maintains independent memory pools for each CUDA stream. Allocations are made from a pool dedicated to the stream on which the allocation call is issued. This design ensures that memory used by operations on one stream does not interfere with or fragment memory used by another, providing strong isolation and predictable performance for concurrent workloads.

• Isolation: Prevents cross-stream fragmentation.
• Predictability: Memory availability is consistent for a given stream's workload.

Memory is freed automatically and asynchronously when the stream reaches the point where the memory is no longer needed, based on the order of operations. The allocator inserts events into the stream to track when an allocation's lifetime ends. This eliminates the need for explicit cudaFree calls and the host-device synchronization they often require, dramatically reducing latency.

• No Manual Freeing: Removes error-prone manual memory management.
• Reduced Synchronization: Frees happen asynchronously on the GPU timeline.

Freed memory is reused on a first-in, first-out (FIFO) basis for new allocations within the same stream. Because operations on a single stream execute in order, memory freed earlier in the stream's history is guaranteed to be safe for reuse by a later allocation. This pattern maximizes memory utilization and minimizes the need to request new memory from the global pool, which is a costly operation.

• High Utilization: Keeps memory actively in use.
• Low Overhead: Reuses existing blocks instead of searching for new ones.

By strictly adhering to stream ordering and FIFO reuse, the allocator prevents the external fragmentation common in general-purpose allocators. Since allocations and frees happen in the same order, free memory blocks naturally coalesce into large, contiguous segments. This is critical for deep learning inference where models may require large, contiguous buffers for weights or activations that a fragmented heap cannot satisfy.

• Contiguous Availability: Enables large tensor allocations.
• Stable Performance: Prevents gradual degradation of allocation times.

Stream-ordered allocation is a foundational primitive for CUDA Graphs, NVIDIA's technology for capturing and replaying sequences of kernels and memory operations. A graph's memory operations are inherently ordered, making the stream-ordered allocator's semantics a perfect fit. This allows graphs to be instantiated with pre-allocated, stable memory addresses, enabling ultra-low-latency graph launches.

• Graph Optimization: Essential for efficient graph capture and replay.
• Deterministic Addresses: Enables aggressive kernel optimizations.

This allocator is exceptionally well-suited for high-throughput inference servers using continuous batching. In such systems, requests (batches) are processed on dedicated streams. The allocator's behavior ensures that the memory for a processed batch is automatically and efficiently recycled for the next batch on the same stream, creating a steady-state memory footprint with minimal management overhead.

• Batch Processing: Ideal for the repetitive, ordered workload of inference.
• Steady-State Memory: Predictable memory usage under load.

The fundamental principle of a stream-ordered allocator is binding each memory allocation to a specific CUDA stream. This creates a deterministic lifetime scope. Memory is not freed with an explicit cudaFree call; instead, it becomes eligible for reuse automatically when all preceding work on that stream has completed. This eliminates the need for global synchronization (e.g., cudaDeviceSynchronize()) just to safely free memory, a major source of latency in dynamic inference workloads.

Traditional allocators suffer from external fragmentation when allocations and frees of varying sizes are interleaved randomly. A stream-ordered allocator guarantees that memory is freed in the exact reverse order of allocation within a stream. This LIFO (Last-In, First-Out) behavior allows the allocator to simply maintain a stack pointer for each stream, reusing the most recently freed block for the next allocation. This pattern is highly efficient for the sequential, batched nature of inference requests, preventing the creation of small, unusable gaps in memory.

Because memory lifecycle is tied to stream progress, free operations are non-blocking and asynchronous. The allocator can reclaim a block for a new allocation as soon as the kernel or memory copy that last used it finishes, without waiting for later work on other streams. This enables compute/memory management overlap. While a new batch is processing on one stream, memory from a completed batch on another stream can be simultaneously prepared for reuse, maximizing GPU utilization and hiding allocator latency.

This allocator is critical for inference servers using continuous batching. As requests of varying sequence lengths arrive and complete asynchronously, memory for KV caches and activations is constantly allocated and freed. A stream-ordered allocator assigned to the inference stream allows:

• Predictable deallocation as each request finishes within the batch.
• Near-zero fragmentation for the sequential workload.
• Elimination of synchronization points between batches, allowing the GPU to run at full throughput. This directly translates to lower latency and higher queries per second (QPS).

Contrasting with the default device allocator highlights the optimization:

• Synchronization: cudaMalloc/Free often require device-wide syncs for safety. cudaMallocAsync is stream-local.
• Fragmentation: Traditional allocators use complex, global free lists prone to fragmentation. Stream-ordered uses simple per-stream stacks.
• Performance: The async allocator can provide >10x faster allocation/deallocation for high-frequency patterns common in model serving.
• Overhead: The traditional allocator has higher per-operation overhead due to thread-safe, global data structure management. The key trade-off is that memory cannot be shared between streams without explicit synchronization.

A memory pool is a pre-allocated, contiguous block of memory from which smaller, variable-sized allocations are sub-allocated. This strategy is foundational to a stream-ordered allocator.

• Reduces System Overhead: By calling cudaMalloc once for a large pool, it eliminates the latency and potential synchronization of frequent system-level allocations and frees.
• Mitigates Fragmentation: Sub-allocations from a pool are managed by a custom algorithm (like stream-ordering), which can be designed to minimize internal fragmentation compared to the default system allocator.
• Core Building Block: A stream-ordered allocator typically manages one or more memory pools, applying its ordering and reuse logic within these pre-reserved regions.

Memory fragmentation is a state where free memory is broken into many small, non-contiguous blocks, preventing the allocation of a larger contiguous segment despite sufficient total free memory.

• Problem for Inference: Dynamic batching with variable sequence lengths can cause rapid allocation/deallocation of differently sized tensors, leading to severe fragmentation in default allocators.
• Stream-Ordering as a Solution: By enforcing allocation and free order per CUDA stream, a stream-ordered allocator can implement a bump allocator pattern within its pool. Freed blocks are reused in the exact order they become available, preventing the random scattering of free blocks that causes fragmentation.
• Impact: High fragmentation leads to out-of-memory errors even when free memory exists, forcing expensive fallbacks like device-to-host transfers.

A CUDA stream is a sequence of operations (kernel launches, memory copies) that execute in issue-order on the GPU. Operations in different streams can execute concurrently.

• Allocation Context: A stream-ordered allocator ties each allocation's lifecycle to a specific stream. The memory is only considered free and reusable after all prior work in that stream has completed.
• Enables Asynchrony: This association allows for asynchronous, zero-overhead reclamation. The allocator can reuse a block for a new allocation in the same stream immediately after its previous user's kernel launches, without requiring a device-wide synchronization (cudaDeviceSynchronize).
• Concurrency Management: Proper use of streams and a stream-ordered allocator allows memory from finished workloads in one stream to be reused while kernels in other streams are still executing, maximizing GPU utilization.

Unified Virtual Memory is a memory management architecture that creates a single, contiguous virtual address space shared between CPU and GPU, allowing both processors to access the same memory pointers.

• Contrast with Stream-Ordered Allocation: UVM simplifies programming by automating data movement but can introduce performance overhead from page fault handling and suboptimal placement. A stream-ordered allocator works with device memory for peak performance, explicitly managing data locality.
• Complementary Use: In systems with UVM, a stream-ordered allocator can be used to manage a pool of device-resident memory for performance-critical tensors (like KV caches), while less frequently accessed data can reside in UVM space and be migrated on demand.

Page-locked memory is host (CPU) memory that is prevented from being paged to disk by the OS, enabling high-bandwidth Direct Memory Access transfers to/from GPU device memory.

• Interaction with Allocators: Stream-ordered allocators typically manage device memory. However, efficient inference pipelines also require fast host-to-device transfers for inputs and device-to-host for outputs. Using page-locked memory for these staging buffers is critical.
• Pipeline Optimization: A complete system uses a stream-ordered allocator for device tensors and a separate pool of page-locked host memory. Overlapping copies (using CUDA streams) and computations hides transfer latency, a principle known as copy/compute overlap.

Peer-to-Peer access enables GPUs within the same system to directly read from and write to each other's memory over a high-speed interconnect like NVLink, bypassing host memory.

• Multi-GPU Inference Context: For large models split across multiple GPUs (model parallelism), tensors must be transferred between devices. P2P access makes this transfer significantly faster.
• Allocator Consideration: A sophisticated, multi-device stream-ordered allocator could manage pools across GPUs with P2P enablement. It would understand that freeing memory on GPU 1 allows it to be directly allocated for a kernel on GPU 2, optimizing inter-GPU communication patterns within a stream's task graph.