Skip to content

Conversation

@Gasoonjia
Copy link
Contributor

@Gasoonjia Gasoonjia commented Feb 9, 2026

Stack from ghstack (oldest at bottom):

Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: D92193164

Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 9, 2026
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

ghstack-source-id: 339552916
Pull Request resolved: #17315
@pytorch-bot
Copy link

pytorch-bot bot commented Feb 9, 2026

🔗 Helpful Links

🧪 See artifacts and rendered test results at hud.pytorch.org/pr/pytorch/executorch/17315

Note: Links to docs will display an error until the docs builds have been completed.

❌ 11 New Failures, 1 Pending, 2 Unrelated Failures

As of commit d3b9ad7 with merge base 6ebaa05 (image):

NEW FAILURES - The following jobs have failed:

FLAKY - The following job failed but was likely due to flakiness present on trunk:

BROKEN TRUNK - The following job failed but was present on the merge base:

👉 Rebase onto the `viable/strict` branch to avoid these failures

This comment was automatically generated by Dr. CI and updates every 15 minutes.

@meta-cla meta-cla bot added the CLA Signed This label is managed by the Facebook bot. Authors need to sign the CLA before a PR can be reviewed. label Feb 9, 2026
@github-actions
Copy link

github-actions bot commented Feb 9, 2026

This PR needs a release notes: label

If your change should be included in the release notes (i.e. would users of this library care about this change?), please use a label starting with release notes:. This helps us keep track and include your important work in the next release notes.

To add a label, you can comment to pytorchbot, for example
@pytorchbot label "release notes: none"

For more information, see
https://github.com/pytorch/pytorch/wiki/PyTorch-AutoLabel-Bot#why-categorize-for-release-notes-and-how-does-it-work.

Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 9, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339642357
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339728013
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
@Gasoonjia Gasoonjia mentioned this pull request Feb 10, 2026
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339777492
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339784126
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339788761
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339802040
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 339914649
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Stack from [ghstack](https://github.com/ezyang/ghstack) (oldest at
bottom):
* #17315
* __->__ #17324

torchcodec we are using (0.10.0.dev20251211) has no longer existed in
https://download.pytorch.org/whl/nightly/torchcodec/, which leads to
lots of cis including all whisper cis crashed. this diff pin bump
torchcodec to bring ci back.

Differential Revision:
[D92797044](https://our.internmc.facebook.com/intern/diff/D92797044/)
Comment on lines 134 to +135
auto stream_result = executorch::backends::cuda::getCurrentCUDAStream(-1);
if (stream_result.ok()) {
ET_CUDA_LOG_WARN(cudaFreeAsync(ptr, stream_result.get()));
} else {
// Fallback to synchronous free if we can't get the stream
ET_CUDA_LOG_WARN(cudaFree(ptr));
}
ET_CHECK_MSG(stream_result.ok(), "Failed to get current CUDA stream");
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Just to be clear this is turning a warning into a fatal error. Just making sure this is your intention.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Yes that's my plan. We should guarantee that each tensor should get the correct cuda_stream

if (is_using_shared_cuda_stream()) {
// Shared stream mode: set handle's stream to nullptr.
// The stream will be retrieved from backend in execute().
handle->cuda_stream = nullptr;
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think it's better to set handle->cuda_stream to the only cuda stream.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

based on offline sync we tried to create a new CudaHandle class to inherit current aoti_handle, puting a shared_ptr<cuda_stream> inside cuda_handle, to make sure there's only one cuda stream in the whole pipeline.

Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)

[ghstack-poisoned]
Gasoonjia added a commit that referenced this pull request Feb 10, 2026
Pull Request resolved: #17315


Right now we always do cudasync before existing cudabackend.execution(). However we only need that when copying data from gpu to cpu; any actions happen inside a same stream do not need explicit sync.
ghstack-source-id: 340006830
@exported-using-ghexport

Differential Revision: [D92193164](https://our.internmc.facebook.com/intern/diff/D92193164/)
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

CLA Signed This label is managed by the Facebook bot. Authors need to sign the CLA before a PR can be reviewed. fb-exported meta-exported

Projects

None yet

Development

Successfully merging this pull request may close these issues.

2 participants