From 6c5036b596c028402db51149d948defd987bf95e Mon Sep 17 00:00:00 2001 From: Changho Hwang Date: Wed, 15 Jan 2025 19:53:07 +0000 Subject: [PATCH] Fix PR #449 --- src/connection.cc | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/src/connection.cc b/src/connection.cc index 23c744db..1e8d972e 100644 --- a/src/connection.cc +++ b/src/connection.cc @@ -76,7 +76,7 @@ void CudaIpcConnection::write(RegisteredMemory dst, uint64_t dstOffset, Register char* dstPtr = (char*)dst.data(); char* srcPtr = (char*)src.data(); - if (!env().cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); + if (!env()->cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); MSCCLPP_CUDATHROW(cudaMemcpyAsync(dstPtr + dstOffset, srcPtr + srcOffset, size, cudaMemcpyDeviceToDevice, *stream_)); INFO(MSCCLPP_P2P, "CudaIpcConnection write: from %p to %p, size %lu", srcPtr + srcOffset, dstPtr + dstOffset, size); @@ -96,7 +96,7 @@ void CudaIpcConnection::updateAndSync(RegisteredMemory dst, uint64_t dstOffset, *src = newValue; uint64_t* dstPtr = reinterpret_cast(reinterpret_cast(dst.data()) + dstOffset); - if (!env().cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); + if (!env()->cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); MSCCLPP_CUDATHROW(cudaMemcpyAsync(dstPtr, src, sizeof(uint64_t), cudaMemcpyHostToDevice, *stream_)); INFO(MSCCLPP_P2P, "CudaIpcConnection atomic write: from %p to %p, %lu -> %lu", src, dstPtr + dstOffset, oldValue, @@ -116,7 +116,7 @@ void CudaIpcConnection::flush(int64_t timeoutUsec) { INFO(MSCCLPP_P2P, "CudaIpcConnection flush: timeout is not supported, ignored"); } - if (!env().cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); + if (!env()->cudaIpcUseDefaultStream && stream_->empty()) stream_->set(cudaStreamNonBlocking); AvoidCudaGraphCaptureGuard guard; MSCCLPP_CUDATHROW(cudaStreamSynchronize(*stream_));