CU-MEMO0001 · synchronous_transfer_in_stream_scope

Count blocking transfers issued where stream work is already in flight.

This is a deterministic rule for cuda. Read its implementation.

Definition

Report a synchronous transfer entry point in a translation unit that also creates or uses a non-default stream. A blocking copy synchronizes the whole device with the host, so it drains every stream that was overlapping compute with transfer and undoes the reason those streams exist. The asynchronous entry point with an explicit stream keeps that overlap.

Evidence

Each finding records the call range, the entry point, and the stream calls that established the scope. The value is the number of blocking transfers.

Exceptions

A translation unit that never touches a stream is left alone, because a blocking copy in a purely sequential program costs nothing extra. Setup and teardown transfers outside the hot path are legitimate, and a project can narrow this rule to its kernel sources.

Examples

Bad

cudaStreamCreate(&stream);
cudaMemcpy(device, host, bytes, cudaMemcpyHostToDevice);

Good

cudaStreamCreate(&stream);
cudaMemcpyAsync(device, host, bytes, cudaMemcpyHostToDevice, stream);

References

• Cites “CUDA C++ Best Practices Guide”, asynchronous transfers and overlapping. Open reference
• Cites “CUDA C++ Programming Guide”, streams. Open reference
• Cites “The NVIDIA Technical Blog”, how to overlap data transfers in CUDA C++. Open reference