ClikaRT::Stream
class
Header: ClikaRT/compute/stream.h
An execution stream: an ordered queue of device work. Tensors are placed on a stream; operations submitted to it run in order, asynchronously of the host.
Put real work on a stream you create (Stream::create(device)): submission there is pipelined, so the host keeps enqueueing while the device runs, and two such streams overlap independent work. A bare Device placement never runs on a device's default stream: the work follows its inputs and otherwise takes the calling thread's CURRENT lane there, the stream StreamOrDevice(device).resolve() names. The per-device Stream::default_stream(device) dispatches each operation on the calling thread instead: the explicit opt-in to synchronous execution (pass it on purpose, or use SynchronousStreamScope); convenient for a one-off call or a quick script, but it gives up that pipelining, so it is not where a throughput-sensitive workload belongs.
A handle to a device execution stream. Cheap to copy (it is an identifier, not the stream's resources).
A moved-from Stream is an EMPTY handle that names no stream: is_default() is false, query_idle() is true, native_handle() is nullptr, synchronize() completes at once, device() reports no device, it compares equal to another empty handle and to nothing else, a copy of it is empty, and allocate raises InvalidArgument. Assign a live stream to use it again.
Every fallible operation returns its value and raises ClikaRT::Error on failure; opt back into a Result<T> at the call site with CLIKART_TRY(...).
Static member functions
default_stream()
The per-device default stream; always valid, never fails.
Declared in ClikaRT/compute/stream.h, line 54
create()
Allocate a fresh non-default stream on device (for overlapping work).
Declared in ClikaRT/compute/stream.h, line 57
from_native_handle()
Adopt a native stream you own (a cudaStream_t on CUDA, passed as an opaque pointer) as a stream on device. Work placed on the returned stream runs on the native stream and is submitted asynchronously of the calling thread, as on a stream from create. The runtime never destroys the native stream. Adopting the same native stream on the same device again returns the same stream.
Scope: an adopted stream is not released before the process ends. The runtime may synchronize it, query it and reclaim memory on it at any point until then, so keep the native stream alive until the process ends.
A null handle raises InvalidArgument. On CUDA a null handle is the legacy default stream, which the runtime cannot order against its own streams; torch reports it (torch.cuda.current_stream().cuda_stream is 0) until you set a stream. Pass a stream you created (cudaStreamCreate, torch.cuda.Stream()). A device whose backend cannot adopt a native stream raises Unsupported.
Declared in ClikaRT/compute/stream.h, line 78
Member functions
Stream(Stream)
Stream(const Stream& other)
Declared in ClikaRT/compute/stream.h, line 82
operator=(Stream)
Copy-assign: rebinds this handle to the same underlying stream.
Declared in ClikaRT/compute/stream.h, line 84
Stream(Stream)
Stream(Stream&& other) noexcept
Move: transfers the handle; other is left empty and stays usable.
Declared in ClikaRT/compute/stream.h, line 86
operator=(Stream)
Move-assign: transfers the handle; other is left empty and stays usable.
Declared in ClikaRT/compute/stream.h, line 88
~Stream()
~Stream()
Releases this handle (the stream itself outlives its handles).
Declared in ClikaRT/compute/stream.h, line 90
device()
Device device() const
The device this stream runs on.
Declared in ClikaRT/compute/stream.h, line 93
is_default()
bool is_default() const
True if this is the device's default stream.
Declared in ClikaRT/compute/stream.h, line 95
query_idle()
bool query_idle() const
True if the stream currently has no work in flight (a hint, not a barrier).
Declared in ClikaRT/compute/stream.h, line 97
operator==()
bool operator==(const Stream& other) const
True when both handles name the same lane: a pool lane compares equal to every handle of that lane and unequal to every other lane; a device's default stream equals itself and no other stream.
Declared in ClikaRT/compute/stream.h, line 102
operator!=()
bool operator!=(const Stream& other) const
Declared in ClikaRT/compute/stream.h, line 103
native_handle()
void* native_handle() const noexcept
The backend-native stream handle, as an opaque pointer, what you launch a hand-written kernel on. static_cast it to the backend's stream type (cudaStream_t for a CUDA stream; the queue handle for others) inside a custom op's compute():
auto cu = static_cast<cudaStream_t>(stream.native_handle());
my_kernel<<<grid, block, smem, cu>>>(...);
Returns nullptr for the CPU backend (host work has no device stream) and for a stream with no live backend resources. The handle is owned by the runtime; do not destroy it, and keep this Stream alive while using it.
Declared in ClikaRT/compute/stream.h, line 116
synchronize()
void synchronize()
Block until all work submitted on this stream has completed.
Declared in ClikaRT/compute/stream.h, line 120
wait_for()
void wait_for(const Tensor& tensor)
Make this stream wait, on the device, for tensor's pending writes: work submitted to this stream after the call runs once the producers of tensor's bytes have finished, with no host wait (a CUDA stream waits on the producer's recorded event; a device's default stream and a CPU stream complete the wait on the calling thread, which is what a synchronous lane is). This stream is also recorded as a reader of those bytes, so a free of tensor's buffer while the work queued here still reads it parks the buffer until this stream has run past that work. The DLPack handoff: a consumer's stream argument, adopted with from_native_handle, waits here before the capsule is handed over. tensor must be defined and live on this stream's device, and this handle must name a stream (InvalidArgument otherwise; an empty handle's code reads FAILED_PRECONDITION). A tensor with no pending work returns at once.
Declared in ClikaRT/compute/stream.h, line 137
allocate()
Tensor allocate(ClikaRT::Span<const std::int64_t> shape, DataType dtype)
Stream::allocate: defined here (not in stream.h) because a by-value Tensor return needs the complete Tensor type, and stream.h cannot include tensor.h (this header includes stream.h). The throw stays caller-side (inline).
Declared in ClikaRT/compute/stream.h, line 155