Asynchronous Execution

The fundamental mechanism where a computational kernel is enqueued for execution on the NPU without stalling the host CPU. The host receives a handle (e.g., a CUDA stream, a SYCL queue event, or a Metal command buffer) immediately, allowing it to proceed with other work.

• Key Benefit: Eliminates CPU idle time while the NPU computes.
• Implementation: Managed via vendor-specific APIs like cudaStream_t, sycl::queue::submit, or MTLCommandBuffer.
• Consequence: Requires explicit synchronization later to ensure kernel completion before using results.

Logical execution channels that maintain the order of operations (kernels and memory transfers) submitted to the NPU. Multiple streams/queues can execute concurrently, enabling fine-grained task parallelism.

• Function: Provide a sequence guarantee within a stream but allow out-of-order completion across different streams.
• Use Case: Overlap data transfer for kernel B in Stream 1 with the execution of kernel A in Stream 0.
• Hardware Support: Modern NPUs/GPUs have multiple hardware queues to service concurrent streams.

The primary performance goal of asynchronous execution. By using multiple streams, data transfer (Host-to-Device or Device-to-Host) for a subsequent operation can occur simultaneously with the kernel execution of a prior operation.

• Pattern: Double-buffering or multi-buffering, where data is staged in pinned host memory.
• Requirement: Requires pinned (page-locked) host memory for DMA transfers; standard malloc memory cannot be used for concurrent transfers.
• Metric: Achievable throughput is often measured by how effectively this overlap hides PCIe or memory bus latency.

Mechanisms to coordinate between the host and device or between concurrent streams, ensuring data dependencies are respected.

• Barriers (cudaStreamSynchronize, queue.wait()): Block the host thread until all preceding commands in a specified stream complete.
• Events (cudaEvent_t, sycl::event): Can be recorded into a stream and later queried by the host or used to synchronize between streams via cudaStreamWaitEvent.
• Implicit Synchronization: Certain operations, like a default-stream memory copy, cause a full device-wide barrier, breaking concurrency.

Advanced model where a workflow of kernels and memory operations is defined as a graph of nodes with explicit dependencies before launch. The runtime can then schedule the entire graph asynchronously and optimally.

• Benefit: Reduces launch overhead and enables the driver to see the full workflow for advanced optimizations.
• Representation: Nodes are operations (kernel, memcpy, memset); edges are dependencies.
• Execution: The entire graph is launched as a single, asynchronous unit using APIs like CUDA Graphs or SYCL Graph.

Asynchronous execution transforms performance bottlenecks and profiling requirements.

• New Bottleneck: Shifts focus from kernel runtime to pipeline latency and achieved concurrency. The goal is to keep all hardware units (compute engines, copy engines) busy.
• Profiling Tools: Requires timeline visualizers (Nsight Systems, Intel VTune, ROCm rocprof) to see overlap, not just kernel timers.
• Key Metrics: Kernel Occupancy (within a stream) and Engine Utilization (across all streams) become critical. Poor overlap often indicates insufficient stream parallelism or improper dependency management.

The canonical use case for asynchronous execution is to overlap computation on the NPU with data transfer between the host CPU and the NPU's memory. This hides the latency of PCIe or on-chip fabric transfers.

• Pattern: Launch an asynchronous memory copy (e.g., H2D), then immediately launch a compute kernel that operates on previously transferred data, while the copy proceeds in parallel.
• Benefit: Eliminates the memory-bound idle time of the NPU, moving the system toward being compute-bound.
• Implementation: Requires the use of non-default streams and pinned (page-locked) host memory to enable true concurrency.

Modern NPUs can execute multiple independent kernels concurrently across different streaming multiprocessors (SMs) or compute units. Asynchronous execution via separate streams enables this parallelism.

• Use Case: Running a pre-processing kernel (e.g., normalization) alongside a main inference kernel on different data batches.
• Requirement: Kernels must be resource-independent (not competing for the same hardware registers or shared memory per SM) to avoid resource contention and serialization.
• Tooling: A kernel profiler (like Nsight Compute or vendor tools) is essential to verify true concurrency and identify serialization points.

Asynchronous execution enables the construction of software pipelines for high-throughput batched inference. The pipeline stages (e.g., copy, compute, copy-back) operate on different data batches simultaneously.

• Pattern: Use multiple streams. While Stream 1 executes the kernel for Batch N, Stream 2 copies results for Batch N-1 back to host, and Stream 3 copies input for Batch N+1 to the NPU.
• Metric: Maximizes aggregate compute throughput (e.g., inferences per second) by keeping all hardware units busy.
• Auto-Tuning Link: The optimal number of concurrent batches and stream count is a key target for auto-tuning systems.

Asynchronous execution allows the host CPU to perform auxiliary tasks—like preparing the next batch of data, running control logic, or handling I/O—without blocking the NPU.

• Pattern: After launching an asynchronous kernel, the host thread is free to execute its own code. It later synchronizes (e.g., with streamSynchronize) to harvest results.
• Benefit: Enables efficient heterogeneous computing where the CPU and NPU work in tandem on different aspects of a problem.
• Caution: Requires careful synchronization to avoid data races where the host reads data before the NPU has finished writing it.

Asynchronous execution is critical for online auto-tuning systems where the performance of multiple kernel variants must be evaluated rapidly.

• Pattern: An auto-tuner (e.g., a Kernel Tuner framework) can launch hundreds of kernel configurations asynchronously across different streams to profile them, using the NPU's performance counters.
• Method: Techniques like Bayesian Optimization rely on launching candidate kernels asynchronously to gather performance data for the next iteration of the search.
• Outcome: Dramatically reduces the total time required to search a large configuration space for optimal parameters like workgroup size or tile size.

Advanced asynchronous workflows use events to create fine-grained dependencies between operations across different streams, enabling complex execution graphs.

• Pattern: Record an event after a memory copy in Stream A. A kernel launch in Stream B can then wait for that event, ensuring data is ready before computation begins, without full stream synchronization.
• Use Case: Essential for graphs with data dependencies where kernels consume outputs from other kernels running in different streams.
• Framework Support: Low-level APIs (like CUDA/OpenCL events) and higher-level graph APIs (like CUDA Graphs) build on this primitive to define and launch pre-defined asynchronous workflows efficiently.

The ability of an NPU or GPU to execute multiple independent kernels simultaneously on different hardware resources (e.g., different streaming multiprocessors). This is a primary mechanism for achieving high throughput and is enabled by asynchronous execution.

• Key Benefit: Hides latency by keeping all compute units busy.
• Requirement: Kernels must be independent or have carefully managed dependencies.
• Hardware Support: Modern NPUs have multiple command queues and hardware schedulers to enable true concurrency.

A critical classification for performance analysis that determines the primary limiter of a kernel's execution time.

• Compute-Bound: Execution time is limited by the speed of the arithmetic logic units (ALUs). The processor is constantly performing calculations, and asynchronous execution helps by overlapping other tasks.
• Memory-Bound: Execution time is limited by memory bandwidth or latency. The compute units idle while waiting for data. Asynchronous execution can help by prefetching data for subsequent kernels while the current one runs.

Identifying the bound is the first step in bottleneck analysis and guides optimization strategy.

A chronological, detailed record of the sequence of instructions, memory accesses, and function calls executed by a program on the NPU. It is used for deep behavioral analysis.

• Purpose: Provides a ground-truth view of runtime behavior, essential for debugging complex concurrency issues and validating performance models.
• Relation to Async: Traces reveal the precise timing and overlap of kernel launches, data transfers, and synchronization events, showing how effectively asynchronous execution is being utilized.
• Tooling: Generated by specialized hardware profilers or through instrumented runtime libraries.

A situation in a processor's execution pipeline where the progression of instructions is halted. Stalls destroy efficiency and are a key target for optimization enabled by asynchronous paradigms.

• Common Causes:
  • Data Dependency: An instruction needs the result of a prior, unfinished instruction.
  • Resource Conflict: Multiple instructions need the same hardware unit.
  • Memory Latency: A load instruction is waiting for data from memory.
• Async Mitigation: By allowing the CPU to continue issuing work while the NPU processes previous tasks, asynchronous execution helps keep the NPU's instruction and task queues full, reducing stalls caused by idle resources.

A key performance metric for parallel architectures, defined as the ratio of active warps (or wavefronts/thread groups) to the maximum number that can be resident on a streaming multiprocessor (SM).

• High Occupancy: Indicates many threads are available to hide latency (e.g., from memory accesses).
• Low Occupancy: The SM may be underutilized, leading to idle cycles.
• Interaction with Async: Asynchronous execution operates at a higher level than occupancy. High kernel concurrency (async) ensures the NPU's SMs are fed with work, while high occupancy ensures each SM uses its threads efficiently to hide pipeline stalls. Both are needed for peak utilization.

The automated process of searching a configuration space (e.g., workgroup size, tile dimensions, loop unroll factor) to find the optimal parameters for a kernel on specific hardware. It is tightly coupled with asynchronous execution strategies.

• Objective: Maximize metrics like compute throughput and minimize latency.
• Method: Uses techniques like Bayesian optimization or genetic algorithms to efficiently navigate the parameter search.
• Async Context: Auto-tuning must evaluate configurations under realistic asynchronous workloads, where resource contention and concurrent kernel execution can dramatically affect the optimal choice. A kernel tuned in isolation may not perform best in a concurrent, async pipeline.