-
Notifications
You must be signed in to change notification settings - Fork 424
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
UCM/CUDA: Added support for device-side variables. #10237
base: master
Are you sure you want to change the base?
Conversation
/** | ||
* cudaGetSymbolAddress is in ucm_cuda_driver_funcs list because | ||
* hook for cuModuleGetGlobal doesn't intercept cudaGetSymbolAddress call. | ||
* */ | ||
UCM_CUDA_FUNC_ENTRY(cudaGetSymbolAddress), |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
why not place it in runtime_api_hooks? we install both driver and runtime hooks anyway..
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
We install driver and runtime hooks only if UCX_MEM_CUDA_HOOK_MODE=reloc
. We install only driver hooks by default.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
maybe make some changes in the hooks logic or array names then? it seems weird that a runtime function is in the "driver" API list
src/ucm/cuda/cudamem.h
Outdated
@@ -49,5 +49,6 @@ cudaError_t ucm_cudaMallocFromPoolAsync(void **devPtr, size_t size, | |||
cudaMemPool_t memPool, | |||
cudaStream_t hStream); | |||
#endif | |||
cudaError_t ucm_cudaGetSymbolAddress(void **devPtr, const void *symbol); |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
maybe move it before ucm_cudaMallocAsync hooks?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Fixed.
shall we add some unit tests for this? |
/azp run UCX PR |
Azure Pipelines successfully started running 1 pipeline(s). |
What
Added hook for
cudaGetSymbolSize
to support device-side variables.