cuda-gen: Add CUDA Graph capture and replay for composite operators - #2006
Open
jeremylt wants to merge 5 commits into
Open
cuda-gen: Add CUDA Graph capture and replay for composite operators#2006jeremylt wants to merge 5 commits into
jeremylt wants to merge 5 commits into
Conversation
…1987) * CUDA graph implemented * Implement CUDA graphs with fallback system in libCEED * Cleaner and Simplified Implementation of CUDA graph * add vector pointer tracking to detect memory changes in Graph * Add PETSc vector setup for CUDA Graph compatibility in CUDA-gen backend * Clean CUDA-graph implementation * Simple and clean implementation of CUDA graph but numerically incorrect * per-operator CUDA Graph implementation * replace cudaMemset with cudaMemsetAsync * Auto-detect graph capture and use async memset with cudaStreamPerThread * CUDA Graph support for composite operators in cuda-gen backend * CUDA Graph is working fine for composite operators * cuda-gen: CUDA Graph capture and replay working for composite operators * style: apply clang-format-22 * cuda-gen: address review feedback for composite CUDA graphs * cuda: add CeedOperatorSetEnableCudaGraph and address review feedback * cuda: add docs for when graph/CUfunction setters aren't supported * cuda-gen: add output pointer check and address review fixes * cuda-gen: address review feedback for ceed handling and async memset * address review feedback * cuda: update CHANGELOG and fix style * cuda-gen: apply clang-format-22
jeremylt
commented
Aug 18, 2026
jeremylt
commented
Aug 18, 2026
jeremylt
commented
Aug 18, 2026
jeremylt
commented
Aug 18, 2026
cuda-ref: use cudaMemcpyAsync for QFunction context
jeremylt
commented
Aug 19, 2026
| } | ||
| CeedCallCuda(ceed, cudaMemcpy(impl->d_data, impl->h_data, ctx_size, cudaMemcpyHostToDevice)); | ||
|
|
||
| CeedCallCuda(ceed, cudaMemcpyAsync(impl->d_data, impl->h_data, ctx_size, cudaMemcpyHostToDevice, cudaStreamPerThread)); |
Member
Author
There was a problem hiding this comment.
Does this change need to happen with the vector memory transfers to the device too? @Nafees01
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.
Squash of #1987 merged into staging branch for Noether testing. See original PR for discussion - this is only a testing branch