torch/lib/c10d/ProcessGroupGloo.cpp
// synchronizeStreams ensures that the private streams associated with
// an algorithm entry wait for the public streams to complete.
synchronizeStreams ensures that the private streams associated with
an algorithm entry wait for the public streams to complete.
synchronizeStreams ensures that the private streams associated with an algorithm entry wait for the public streams to complete .
void synchronizeStreams(THCState* thcState, AlgorithmEntry* entry) {
  at::DeviceGuard deviceGuard;
  const auto& key = entry->key;
  for (size_t i = 0; i < key.devices.size(); i++) {
    const auto& device = key.devices[i];
    auto publicStream = THCState_getCurrentStreamOnDevice(thcState, device);
    auto privateStream = entry->streams[i].getStream();
    auto event = entry->events[i].getEvent();
     We must use the device guard to cover the case where the public
     stream is stream 0 and cudaEventRecord relies on the current
     device to find the right one.
    deviceGuard.set_index(key.devices[i]);
    C10D_CUDA_CHECK(cudaEventRecord(event, publicStream));
    C10D_CUDA_CHECK(cudaStreamWaitEvent(privateStream, event, 0));
  }
}
void synchronizeStreams ( THCState * thcState , AlgorithmEntry * entry ) { at :: DeviceGuard deviceGuard ; const auto & key = entry -> key ; for ( size_t i = 0 ; i < key . devices . size ( ) ; i ++ ) { const auto & device = key . devices [ i ] ; auto publicStream = THCState_getCurrentStreamOnDevice ( thcState , device ) ; auto privateStream = entry -> streams [ i ] . getStream ( ) ; auto event = entry -> events [ i ] . getEvent ( ) ; deviceGuard . set_index ( key . devices [ i ] ) ; C10D_CUDA_CHECK ( cudaEventRecord ( event , publicStream ) ) ; C10D_CUDA_CHECK ( cudaStreamWaitEvent ( privateStream , event , 0 ) ) ; } }
synchronizeStreams ensures that the private streams associated with
an algorithm entry wait for the public streams to complete.
synchronizeStreams ensures that the private streams associated with an algorithm entry wait for the public streams to complete .
// Synchronize private stream with public stream. // // We must use the device guard to cover the case where the public // stream is stream 0 and cudaEventRecord relies on the current // device to find the right one. //