What are the new unique-id's for CUDA streams and contexts useful for?

Question

CUDA 12 introduces two new API calls, cuStreamGetId() and cuCtxGetId() which return "unique ID"s associated with a stream or a context respectively. I'm struggling to understand why this is useful, or how this would be used. Are the handles for streams and contexts not unique? i.e. does CUDA create copies of CUstream_st and CUctx_st structures with the same values, or which describe the same entities? If it does - under what circumstances?

Answer 1

A stream handle does not uniquely identify an active stream. Consider the following sequence:

cudaStream_t s, s2;
cudaStreamCreate(&s);
slow_kernel<<<grid, block, 0, s>>>(args);
cudaStreamDestroy(s);   // no synchronization - the kernel is still running!
cudaStreamCreate(&s2);  // almost certainly s2 == s
fast_kernel<<<grid, block, 0, s2>>>(args);
cudaStreamDestroy(s2);

In the example above, you'll almost certainly get s2 == s even though the slow_kernel is still running - it's even possible that fast_kernel finishes before slow_kernel.

On top of that there's an issue with special handles like default stream or per-thread default stream. Thread ids can be reused by the OS, too (albeit not as aggressively).

One place where uniquely identifying a stream is crucial is stream-ordered resource management (like cudaMallocAsync).

Assigning a stream name doesn't solve the issue, because the party interested in establishing stream identity/equality doesn't necessarily own the stream and assigning a name would constitute a breach of contract.