cccl-hide-from-abi

c/experimental/stf/include/cccl/c/experimental/stf/stf.h

@@ -1775,7 +1775,7 @@ stf_error_t stf_cuda_kernel_add_desc_cufunc(
 //!
 //! \see stf_cuda_kernel_add_desc_cufunc(), stf_cuda_kernel_get_arg()
 
-inline cudaError_t stf_cuda_kernel_add_desc(
+inline cudaError_t stf_cuda_kernel_add_desc( // NOLINT(cccl-hide-from-abi)
   stf_cuda_kernel_handle k,
   const void* func,
   dim3 grid_dim_,

c/experimental/stf/test/test_logical_data_with_place.cu

@@ -110,7 +110,7 @@ C2H_TEST("stf_logical_data_with_place - device place (data on current device)",
   stf_ctx_handle ctx = stf_ctx_create();
   REQUIRE(ctx != nullptr);
 
-  float* d_raw    = nullptr;
+  float* d_raw    = nullptr; // NOLINT(misc-const-correctness)
   cudaError_t err = cudaMalloc(&d_raw, N * sizeof(float));
   REQUIRE(err == cudaSuccess);
   const std::unique_ptr<void, decltype(&cudaFree)> d_data_owner(d_raw, cudaFree);

c/experimental/stf/test/test_placement.cpp

@@ -197,7 +197,7 @@ C2H_TEST("shaped allocation on composite data places", "[places][placement][allo
   REQUIRE(dp != nullptr);
 
   // A byte count alone cannot carry the tensor geometry: must fail cleanly
-  void* bad = stf_data_place_allocate(dp, static_cast<ptrdiff_t>(n * sizeof(int)), nullptr);
+  const void* bad = stf_data_place_allocate(dp, static_cast<ptrdiff_t>(n * sizeof(int)), nullptr);
   REQUIRE(bad == nullptr);
 
   void* ptr = stf_data_place_allocate_nd(dp, &dims, sizeof(int), nullptr);

c/experimental/stf/test/test_places.cpp

@@ -741,7 +741,7 @@ C2H_TEST("data_place_allocate_invalid_returns_null", "[places][allocate]")
 {
   stf_data_place_handle inv = stf_data_place_affine();
   REQUIRE(inv != nullptr);
-  void* ptr = stf_data_place_allocate(inv, 64, nullptr);
+  const void* ptr = stf_data_place_allocate(inv, 64, nullptr);
   REQUIRE(ptr == nullptr);
   stf_data_place_destroy(inv);
 }

cmake/clang_tidy_plugins/CMakeLists.txt

@@ -87,7 +87,8 @@ function(_cccl_parse_clang_tidy_config source_file out_var)
     [==[{
       "Checks": ["-*", "-clang-diagnostic-*"],
       "CheckOptions": {},
-      "ExtraArgs": [],
+      "ExtraArgs": ["-nocudainc", "-nocudalib"],
+      "RemovedArgs": ["-G"],
       "HeaderFileExtensions": ["", "h", "hh", "hpp", "hxx", "cuh", "inl", "ipp"],
       "SystemHeaders": false
     }]==]
@@ -209,24 +210,24 @@ function(cccl_create_clang_tidy_test plugin source)
 
   cmake_path(ABSOLUTE_PATH source NORMALIZE)
   cmake_path(GET source STEM test_name)
-  set(target "cccl.test.tidy.${plugin}.${test_name}")
-
+  set(test_target "cccl.test.tidy.${plugin}.${test_name}")
+  set(compile_target "${test_target}.compile")
   # We don't actually care about compiling this (and in fact, we never want to!). The
   # point of adding this here is to get the corresponding entries into the compilation
   # database so that clang-tidy knows how to interpret the file properly.
-  add_library("${target}.compile" OBJECT EXCLUDE_FROM_ALL "${source}")
+  add_library("${compile_target}" OBJECT EXCLUDE_FROM_ALL "${source}")
   target_link_libraries(
-    "${target}.compile"
+    "${compile_target}"
     PRIVATE cccl.compiler_interface libcudacxx::libcudacxx
   )
   target_compile_options(
-    "${target}.compile"
+    "${compile_target}"
     PRIVATE -Wno-pragma-clang-attribute
   )
 
   _cccl_parse_clang_tidy_config("${source}" config)
   add_test(
-    NAME "${target}"
+    NAME "${test_target}"
     COMMAND
       "${CMAKE_COMMAND}" "-DSOURCE=${source}"
       "-DOUTPUT=${CMAKE_CURRENT_BINARY_DIR}/${test_name}.out"
@@ -235,7 +236,7 @@ function(cccl_create_clang_tidy_test plugin source)
       "-DFILECHECK=${CCCL_FILECHECK}" -P
       "${CMAKE_CURRENT_FUNCTION_LIST_DIR}/run_filecheck.cmake"
   )
-  _cccl_test_requires_target("${target}" cccl.tidy.plugins)
+  _cccl_test_requires_target("${test_target}" cccl.tidy.plugins)
 endfunction()
 
 block()
@@ -273,3 +274,4 @@ add_subdirectory(modernize_type_traits)
 add_subdirectory(prefer_cuda_traits)
 add_subdirectory(simplify_type_traits)
 add_subdirectory(nodebug_on_aliases)
+add_subdirectory(hide_from_abi)

cmake/clang_tidy_plugins/avoid_internal_linkage_in_headers/avoid_internal_linkage_in_headers.cpp

@@ -6,38 +6,24 @@
 #include <clang-tidy/ClangTidyCheck.h>
 #include <clang-tidy/ClangTidyModule.h>
 #include <clang-tidy/FileExtensionsSet.h>
-#include <clang-tidy/utils/FileExtensionsUtils.h>
-#include <clang-tidy/utils/LexerUtils.h>
 #include <clang/AST/ASTContext.h>
 #include <clang/AST/Decl.h>
 #include <clang/ASTMatchers/ASTMatchers.h>
 #include <clang/ASTMatchers/ASTMatchFinder.h>
 #include <clang/Basic/Diagnostic.h>
 #include <clang/Basic/SourceManager.h>
 #include <clang/Lex/Lexer.h>
+#include <common/lexer.h>
+#include <common/matchers.h>
 #include <llvm/ADT/StringRef.h>
 #include <llvm/Support/ErrorHandling.h>
-#include <llvm/Support/Path.h>
 
 namespace clang::tidy::cccl
 {
 namespace
 {
 using namespace ast_matchers; // NOLINT(google-build-using-namespace)
 
-AST_MATCHER_P(Decl, isExpansionInHeaderFile, const FileExtensionsSet*, extensions)
-{
-  static_cast<void>(Builder);
-  static_cast<void>(Finder);
-
-  const auto begin    = Node.getBeginLoc();
-  const auto& sm      = Finder->getASTContext().getSourceManager();
-  const auto filename = sm.getFilename(sm.getExpansionLoc(begin));
-  // LLVM's header utility rejects extensionless files even when the empty extension is configured.
-  return utils::isExpansionLocInHeaderFile(begin, sm, *extensions)
-      || (extensions->contains("") && !filename.empty() && llvm::sys::path::extension(filename).empty());
-}
-
 AST_MATCHER(VarDecl, isFileVarDecl)
 {
   static_cast<void>(Builder);
@@ -65,89 +51,6 @@ AST_MATCHER(Decl, hasExplicitGlobalAttr)
   return attr && !attr->isImplicit();
 }
 
-#if LLVM_VERSION_MAJOR >= 23
-[[nodiscard]] CharSourceRange findTokenTextInRange(
-  CharSourceRange Range,
-  const SourceManager& SM,
-  const LangOptions& LangOpts,
-  llvm::function_ref<bool(const Token&)> Pred)
-{
-  return utils::lexer::findTokenTextInRange(std::move(Range), SM, LangOpts, std::move(Pred));
-}
-#else
-// Word-for-word backport of findTokenTextInRange()
-[[nodiscard]] CharSourceRange findTokenTextInRange(
-  CharSourceRange Range,
-  const SourceManager& SM,
-  const LangOptions& LangOpts,
-  llvm::function_ref<bool(const Token&)> Pred)
-{
-  if (Range.isInvalid())
-  {
-    return {};
-  }
-
-  // Normalize to a file-based char range so raw lexing can operate on one
-  // contiguous buffer and reject unmappable (e.g. macro) ranges.
-  const CharSourceRange FileRange = Lexer::makeFileCharRange(Range, SM, LangOpts);
-
-  if (FileRange.isInvalid())
-  {
-    return {};
-  }
-
-  const auto [BeginFID, BeginOffset] = SM.getDecomposedLoc(FileRange.getBegin());
-  const auto [EndFID, EndOffset]     = SM.getDecomposedLoc(FileRange.getEnd());
-
-  if (BeginFID != EndFID || BeginOffset > EndOffset)
-  {
-    return {};
-  }
-
-  bool Invalid           = false;
-  const StringRef Buffer = SM.getBufferData(BeginFID, &Invalid);
-  if (Invalid)
-  {
-    return {};
-  }
-
-  const char* LexStart = Buffer.data() + BeginOffset;
-  // Re-lex raw tokens in the bounded file buffer while preserving comments so
-  // callers can match tokens regardless of interleaved comments.
-  Lexer TheLexer{SM.getLocForStartOfFile(BeginFID), LangOpts, Buffer.begin(), LexStart, Buffer.end()};
-  TheLexer.SetCommentRetentionState(true);
-
-  while (true)
-  {
-    Token Tok{};
-    if (TheLexer.LexFromRawLexer(Tok))
-    {
-      return {};
-    }
-
-    if (Tok.is(tok::eof) || Tok.getLocation() == FileRange.getEnd()
-        || SM.isBeforeInTranslationUnit(FileRange.getEnd(), Tok.getLocation()))
-    {
-      return {};
-    }
-
-    if (!Pred(Tok))
-    {
-      continue;
-    }
-
-    Token NextTok{};
-    if (TheLexer.LexFromRawLexer(NextTok))
-    {
-      return {};
-    }
-    // Return a char range ending at the next token start so trailing trivia of
-    // the matched token is included (useful for fix-it removals).
-    return CharSourceRange::getCharRange(Tok.getLocation(), NextTok.getLocation());
-  }
-}
-#endif
-
 class AvoidInternalLinkageInHeadersCheck final : public ClangTidyCheck
 {
   static constexpr StringRef FUNCTION_BIND{"f"};
@@ -171,7 +74,7 @@ class AvoidInternalLinkageInHeadersCheck final : public ClangTidyCheck
     // misc-anonymous-namespace-in-header already diagnoses anonymous namespaces.
     const auto header_decl =
       allOf(unless(isExpansionInSystemHeader()),
-            isExpansionInHeaderFile(header_extensions_),
+            util::isExpansionInHeaderFile(header_extensions_),
             unless(isInAnonymousNamespace()));
 
     finder->addMatcher(
@@ -250,7 +153,7 @@ class AvoidInternalLinkageInHeadersCheck final : public ClangTidyCheck
   {
     const auto& sm        = context.getSourceManager();
     const auto& lang_opts = context.getLangOpts();
-    const auto range      = findTokenTextInRange(
+    const auto range      = util::findTokenTextInRange(
       CharSourceRange::getTokenRange(declaration.getBeginLoc(), declaration.getLocation()),
       sm,
       lang_opts,

cmake/clang_tidy_plugins/common/lexer.h

@@ -0,0 +1,103 @@
+// SPDX-FileCopyrightText: Copyright (c) 2026 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#pragma once
+
+#include <utility>
+
+#include <clang-tidy/utils/LexerUtils.h>
+#include <clang/Basic/LangOptions.h>
+#include <clang/Basic/SourceLocation.h>
+#include <clang/Basic/SourceManager.h>
+#include <clang/Basic/TokenKinds.h>
+#include <clang/Lex/Lexer.h>
+#include <clang/Lex/Token.h>
+#include <llvm/ADT/STLFunctionalExtras.h>
+#include <llvm/ADT/StringRef.h>
+#include <llvm/Config/llvm-config.h>
+
+namespace clang::tidy::cccl::util
+{
+#if LLVM_VERSION_MAJOR >= 23
+[[nodiscard]] inline CharSourceRange findTokenTextInRange(
+  CharSourceRange Range,
+  const SourceManager& SM,
+  const LangOptions& LangOpts,
+  llvm::function_ref<bool(const Token&)> Pred)
+{
+  return utils::lexer::findTokenTextInRange(std::move(Range), SM, LangOpts, std::move(Pred));
+}
+#else
+// Backport of https://github.com/llvm/llvm-project/blob/main/clang-tools-extra/clang-tidy/utils/LexerUtils.cpp.
+[[nodiscard]] inline CharSourceRange findTokenTextInRange(
+  CharSourceRange Range,
+  const SourceManager& SM,
+  const LangOptions& LangOpts,
+  llvm::function_ref<bool(const Token&)> Pred)
+{
+  if (Range.isInvalid())
+  {
+    return {};
+  }
+
+  // Normalize to a file-based char range so raw lexing can operate on one
+  // contiguous buffer and reject unmappable (e.g. macro) ranges.
+  const CharSourceRange FileRange = Lexer::makeFileCharRange(Range, SM, LangOpts);
+
+  if (FileRange.isInvalid())
+  {
+    return {};
+  }
+
+  const auto [BeginFID, BeginOffset] = SM.getDecomposedLoc(FileRange.getBegin());
+  const auto [EndFID, EndOffset]     = SM.getDecomposedLoc(FileRange.getEnd());
+
+  if (BeginFID != EndFID || BeginOffset > EndOffset)
+  {
+    return {};
+  }
+
+  bool Invalid           = false;
+  const StringRef Buffer = SM.getBufferData(BeginFID, &Invalid);
+  if (Invalid)
+  {
+    return {};
+  }
+
+  const char* LexStart = Buffer.data() + BeginOffset;
+  // Re-lex raw tokens in the bounded file buffer while preserving comments so
+  // callers can match tokens regardless of interleaved comments.
+  Lexer TheLexer{SM.getLocForStartOfFile(BeginFID), LangOpts, Buffer.begin(), LexStart, Buffer.end()};
+  TheLexer.SetCommentRetentionState(true);
+
+  while (true)
+  {
+    Token Tok{};
+    if (TheLexer.LexFromRawLexer(Tok))
+    {
+      return {};
+    }
+
+    if (Tok.is(tok::eof) || Tok.getLocation() == FileRange.getEnd()
+        || SM.isBeforeInTranslationUnit(FileRange.getEnd(), Tok.getLocation()))
+    {
+      return {};
+    }
+
+    if (!Pred(Tok))
+    {
+      continue;
+    }
+
+    Token NextTok{};
+    if (TheLexer.LexFromRawLexer(NextTok))
+    {
+      return {};
+    }
+    // Return a char range ending at the next token start so trailing trivia of
+    // the matched token is included (useful for fix-it removals).
+    return CharSourceRange::getCharRange(Tok.getLocation(), NextTok.getLocation());
+  }
+}
+#endif
+} // namespace clang::tidy::cccl::util

cmake/clang_tidy_plugins/common/matchers.h

@@ -0,0 +1,32 @@
+// SPDX-FileCopyrightText: Copyright (c) 2026 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#pragma once
+
+#include <clang-tidy/FileExtensionsSet.h>
+#include <clang-tidy/utils/FileExtensionsUtils.h>
+#include <clang/AST/ASTContext.h>
+#include <clang/AST/Decl.h>
+#include <clang/ASTMatchers/ASTMatchers.h>
+#include <clang/Basic/SourceManager.h>
+#include <llvm/Support/Path.h>
+
+namespace clang::tidy::cccl::util
+{
+AST_MATCHER_P(Decl, isExpansionInHeaderFile, const FileExtensionsSet*, extensions) // NOLINT
+{
+  const auto begin    = Node.getBeginLoc();
+  const auto& SM      = Finder->getASTContext().getSourceManager();
+  const auto filename = SM.getFilename(SM.getExpansionLoc(begin));
+  // LLVM's header utility rejects extensionless files even when the empty extension is configured.
+  return utils::isExpansionLocInHeaderFile(begin, SM, *extensions)
+      || (extensions->contains("") && !filename.empty() && llvm::sys::path::extension(filename).empty());
+}
+
+using file_matching_type = decltype(ast_matchers::isExpansionInFileMatching(""));
+
+AST_MATCHER_FUNCTION(file_matching_type, isExpansionInTestOrExampleFile)
+{
+  return ast_matchers::isExpansionInFileMatching(".*/(test|c2h|examples)/.*");
+}
+} // namespace clang::tidy::cccl::util

cmake/clang_tidy_plugins/hide_from_abi/CMakeLists.txt

@@ -0,0 +1,5 @@
+list(APPEND CMAKE_MESSAGE_CONTEXT "hide_from_abi")
+
+cccl_create_clang_tidy_plugin(hide_from_abi.cpp)
+
+add_subdirectory(test)