fix: replace unsafe d2h/h2d MemcpyAsync calls with synchronous Memcpy#107
Closed
kilinchange wants to merge 2 commits intomasterfrom
Closed
fix: replace unsafe d2h/h2d MemcpyAsync calls with synchronous Memcpy#107kilinchange wants to merge 2 commits intomasterfrom
kilinchange wants to merge 2 commits intomasterfrom
Conversation
kilinchange
force-pushed
the
fix/memcpyasync-host-lifetime
branch
2 times, most recently
from
82a8322 to
616ad5b
Compare
Collaborator
Author
|
kilinchange
force-pushed
the
fix/memcpyasync-host-lifetime
branch
from
cf5b07f to
c35455d
Compare
Chamberlain0w0
requested changes
Feb 24, 2026
| CUDA_CHECK(cudaMallocAsync(&device_input_ptrs, sizeof(T *) * num_inputs, stream)); | ||
| CUDA_CHECK(cudaMemcpyAsync(device_input_ptrs, host_input_ptrs.data(), sizeof(T *) * num_inputs, | ||
| cudaMemcpyHostToDevice, stream)); | ||
| CUDA_CHECK(cudaMemcpy(device_input_ptrs, host_input_ptrs.data(), sizeof(T *) * num_inputs, |
Contributor
There was a problem hiding this comment.
如果最后加了 stream sync 的话，这里用 Async 应该也可以？
| // NOTE(dcj): | ||
| // Synchronize the stream here to ensure all preceding H2D/D2H memcpy | ||
| // operations have completed before the host buffers go out of scope. | ||
| CUDA_CHECK(cudaStreamSynchronize(stream)); |
Contributor
There was a problem hiding this comment.
这里 host 内存都还没被释放，这里加了 stream sync 的话，前面 memcpy 都可以保留 async
| stream)); | ||
| CUDA_CHECK(cudaMemcpy(out_dims_dev, idx_dims.data(), num_dims * sizeof(int64_t), cudaMemcpyHostToDevice)); | ||
| CUDA_CHECK(cudaMemcpy(in_strides_dev, in_strides.data(), num_dims * sizeof(int64_t), cudaMemcpyHostToDevice)); | ||
| CUDA_CHECK(cudaMemcpy(out_strides_dev, out_strides.data(), num_dims * sizeof(int64_t), cudaMemcpyHostToDevice)); |
kilinchange
force-pushed
the
fix/memcpyasync-host-lifetime
branch
3 times, most recently
from
40ce151 to
03a6c57
Compare
… + SynchronizeStream
kilinchange
force-pushed
the
fix/memcpyasync-host-lifetime
branch
from
03a6c57 to
d569171
Compare
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
2 participants
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Follow-up to #103 comment: replace unsafe d2h/h2d MemcpyAsync calls with synchronous Memcpy.