Skip to content
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

[Unity][BYOC] Cache cuBlasLt handle with thread entry #15030

Merged
merged 3 commits into from
Jun 7, 2023

Conversation

vinx13
Copy link
Member

@vinx13 vinx13 commented Jun 6, 2023

Added CuBlasLtThreadEntry to cache cublasLt handle. Without it, cudaDeviceSynchonize is inserted and causes overhead.

cc @masahi @yelite

@tvm-bot
Copy link
Collaborator

tvm-bot commented Jun 6, 2023

Thanks for contributing to TVM! Please refer to the contributing guidelines https://tvm.apache.org/docs/contribute/ for useful information and tips. Please request code reviews from Reviewers by @-ing them in a comment.

Generated by tvm-bot

@github-actions github-actions bot requested a review from masahi June 6, 2023 10:18
@masahi masahi changed the title [BYOC] Cache cuBlasLt handle with thread entry [Unity][BYOC] Cache cuBlasLt handle with thread entry Jun 6, 2023
@masahi
Copy link
Member

masahi commented Jun 6, 2023

I get this compile error using g++11:

/home/masa/projects/dev/tvm/src/runtime/contrib/cublas/cublas_json_runtime.cc:45:23: error: using-declaration for non-member at class scope
   45 |   using tvm::contrib::CuBlasLtThreadEntry;
      |                       ^~~~~~~~~~~~~~~~~~

// TODO(masahi): Reuse the same handle across different subgraphs
cublasLtHandle_t handle;
cublasLtCreate(&handle);
CuBlasLtThreadEntry* entry_ptr = CuBlasLtThreadEntry::ThreadLocal();
Copy link
Member

Choose a reason for hiding this comment

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

I removed using and replaced this line with

auto entry_ptr = tvm::contrib::CuBlasLtThreadEntry::ThreadLocal();

to workaround the compile error.

@masahi
Copy link
Member

masahi commented Jun 7, 2023

@vinx13 Need to fix lint.

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

Successfully merging this pull request may close these issues.

4 participants