1

When you perform a wait-on-value operation using the CUDA driver API call cuStreamWaitValue32(), you can specify the flag CU_STREAM_WAIT_VALUE_FLUSH. Here's what the documentation says it does:

Follow the wait operation with a flush of outstanding remote writes. This means that, if a remote write operation is guaranteed to have reached the device before the wait can be satisfied, that write is guaranteed to be visible to downstream device work.

My question is: What counts as a "remote write" in this context? Is it only calls to cuStreamWriteValue32() / cuStreamWriteValue64()? Is it any kind of write involving a different device or the host? Including cudaMemcpy() and friends?

einpoklum
  • 118,144
  • 57
  • 340
  • 684

1 Answers1

1

Remote Writes are writes issued by a third party device targeting the GPU device memory. This is related to GPUDirect RDMA. By extension, that also includes writes by issued by the CPU via GDRCopy mappings.

davide rossetti
  • 11
  • 1
  • 1
    Is this your common-sense reading of the definition, or do you have an "official basis" for this answer? – einpoklum Sep 19 '20 at 07:23