Skip to content

Use CUML_KERNEL for remaining kernels - #8465

Open
ahmedtaha100 wants to merge 2 commits into
NVIDIA:mainfrom
ahmedtaha100:enh-standardize-kernel-annotations
Open

Use CUML_KERNEL for remaining kernels#8465
ahmedtaha100 wants to merge 2 commits into
NVIDIA:mainfrom
ahmedtaha100:enh-standardize-kernel-annotations

Conversation

@ahmedtaha100

@ahmedtaha100 ahmedtaha100 commented Aug 11, 2026

Copy link
Copy Markdown

@divyegala @csadorf

Closes #6532

Summary

This fixes NVIDIA/cuml issue #6532, "Create preprocessor macros to standardize annotations." CUML_KERNEL already exists in cpp/include/cuml/common/utils.hpp, so this patch uses it instead of defining another macro.

  • Use CUML_KERNEL for all 16 remaining C++ CUDA kernels.
  • Include <cuml/common/utils.hpp> directly in each affected file.

Ten of these kernels were already declared static __global__. The other six gain internal linkage through CUML_KERNEL, so they no longer produce externally linked kernel symbols.

The Python NVRTC RawKernel strings remain unchanged because Python loads those kernels by name at runtime. Their extern "C" entry points must remain visible.

Kernel bodies and public APIs are unchanged.

Validation

  • All applicable changed-file pre-commit hooks passed.
  • CUDA 12.9 builds passed for cuml, SG_ISOLATION_FOREST_TEST, and SG_RF_TEST.
  • RAPIDS detect-weak-linking reported no externally linked CUDA kernel symbols in the built libcuml.so. Upstream conda-cpp-checks uses the same detector.
  • git diff --check passed.
  • A full C++ source search found no raw __global__ void kernel definitions.

This patch adds no unit test because runtime tests do not directly inspect linkage. The targeted builds compile every changed translation unit, and detect-weak-linking checks the resulting libcuml.so.

@ahmedtaha100
ahmedtaha100 requested a review from a team as a code owner August 11, 2026 00:42
@copy-pr-bot

copy-pr-bot Bot commented Aug 11, 2026

Copy link
Copy Markdown

This pull request requires additional validation before any workflows can run on NVIDIA's runners.

Pull request vetters can view their responsibilities here.

Contributors can view more details about this message here.

@ahmedtaha100

ahmedtaha100 commented Aug 11, 2026

Copy link
Copy Markdown
Author

Could a maintainer please add the improvement and non-breaking labels? Label Checker requires one category and one compatibility label, and GitHub does not allow outside contributors to apply them. Thanks!

@divyegala divyegala left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

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

Thank you for the contribution!

@divyegala divyegala added improvement Improvement / enhancement to an existing function non-breaking Non-breaking change labels Aug 11, 2026
@divyegala

Copy link
Copy Markdown
Contributor

/ok to test c74847c

@ahmedtaha100

Copy link
Copy Markdown
Author

Fixed the copyright lint failure in 64858f6. The follow-up only updates the two headers emitted by verify-copyright-cuml, and all applicable local hooks pass. Could you please run /ok to test 64858f6 for the new head? Thanks!

@divyegala

Copy link
Copy Markdown
Contributor

/ok to test 64858f6

@ahmedtaha100

Copy link
Copy Markdown
Author

The lint fix now passes, but the remaining jobs fail before building during AWS OIDC authentication. It looks infrastructure-related, but if there’s anything I can correct on the PR side, I’m happy to implement it. Thank you! @divyegala

@divyegala

Copy link
Copy Markdown
Contributor

Thanks @ahmedtaha100 , we just changed organizations for this repository so we may see some transient errors. I'll follow up and run CI when it is ready, and let you know if there are any real failures related to this PR.

@ahmedtaha100

Copy link
Copy Markdown
Author

Great thank you for your help!

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

CUDA/C++ improvement Improvement / enhancement to an existing function non-breaking Non-breaking change

Projects

None yet

Development

Successfully merging this pull request may close these issues.

Create preprocessor macros to standardize annotations

2 participants