Move more implementations to precompiled shared library (part 2)

conda/recipes/librmm/recipe.yaml

@@ -78,6 +78,7 @@ outputs:
       host:
         - cuda-version =${{ cuda_version }}
         - cuda-cudart-dev
+        - rapids-logger =0.1
       run:
         - ${{ pin_compatible("cuda-version", upper_bound="x", lower_bound="x") }}
         - cuda-cudart

cpp/CMakeLists.txt

@@ -86,8 +86,18 @@ include(cmake/thirdparty/get_nvtx.cmake)
 # ##################################################################################################
 # * library targets --------------------------------------------------------------------------------
 
-add_library(rmm src/aligned.cpp src/cuda_device.cpp src/cuda_stream_pool.cpp
-                src/cuda_stream_view.cpp src/cuda_stream.cpp src/device_buffer.cpp)
+add_library(
+  rmm
+  src/aligned.cpp
+  src/cuda_device.cpp
+  src/cuda_stream_pool.cpp
+  src/cuda_stream_view.cpp
+  src/cuda_stream.cpp
+  src/device_buffer.cpp
+  src/error.cpp
+  src/exec_policy.cpp
+  src/logger.cpp
+  src/prefetch.cpp)
 add_library(rmm::rmm ALIAS rmm)
 
 target_include_directories(

cpp/include/rmm/error.hpp

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020-2024, NVIDIA CORPORATION.
+ * Copyright (c) 2020-2025, NVIDIA CORPORATION.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -59,19 +59,19 @@ class bad_alloc : public std::bad_alloc {
    *
    * @param msg Message to be associated with the exception
    */
-  bad_alloc(const char* msg) : _what{std::string{std::bad_alloc::what()} + ": " + msg} {}
+  bad_alloc(const char* msg);
 
   /**
    * @brief Constructs a bad_alloc with the error message.
    *
    * @param msg Message to be associated with the exception
    */
-  bad_alloc(std::string const& msg) : bad_alloc{msg.c_str()} {}
+  bad_alloc(std::string const& msg);
 
   /**
    * @briefreturn{The explanatory string}
    */
-  [[nodiscard]] const char* what() const noexcept override { return _what.c_str(); }
+  [[nodiscard]] const char* what() const noexcept override;
 
  private:
   std::string _what;
@@ -91,14 +91,14 @@ class out_of_memory : public bad_alloc {
    *
    * @param msg Message to be associated with the exception
    */
-  out_of_memory(const char* msg) : bad_alloc{std::string{"out_of_memory: "} + msg} {}
+  out_of_memory(const char* msg);
 
   /**
    * @brief Constructs an out_of_memory with the error message.
    *
    * @param msg Message to be associated with the exception
    */
-  out_of_memory(std::string const& msg) : out_of_memory{msg.c_str()} {}
+  out_of_memory(std::string const& msg);
 };
 
 /**

cpp/include/rmm/exec_policy.hpp

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020-2024, NVIDIA CORPORATION.
+ * Copyright (c) 2020-2025, NVIDIA CORPORATION.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -38,7 +38,7 @@ namespace RMM_NAMESPACE {
  */
 
 /**
- * @brief Synchronous execution policy for allocations using thrust
+ * @brief Synchronous execution policy for allocations using Thrust
  */
 using thrust_exec_policy_t =
   thrust::detail::execute_with_allocator<mr::thrust_allocator<char>,
@@ -57,21 +57,16 @@ class exec_policy : public thrust_exec_policy_t {
    * @param mr The resource to use for allocating temporary memory
    */
   explicit exec_policy(cuda_stream_view stream      = cuda_stream_default,
-                       device_async_resource_ref mr = mr::get_current_device_resource_ref())
-    : thrust_exec_policy_t(
-        thrust::cuda::par(mr::thrust_allocator<char>(stream, mr)).on(stream.value()))
-  {
-  }
+                       device_async_resource_ref mr = mr::get_current_device_resource_ref());
 };
 
-#if THRUST_VERSION >= 101600
-
 /**
- * @brief Asynchronous execution policy for allocations using thrust
+ * @brief Asynchronous execution policy for allocations using Thrust
  */
 using thrust_exec_policy_nosync_t =
   thrust::detail::execute_with_allocator<mr::thrust_allocator<char>,
                                          thrust::cuda_cub::execute_on_stream_nosync_base>;
+
 /**
  * @brief Helper class usable as a Thrust CUDA execution policy
  * that uses RMM for temporary memory allocation on the specified stream
@@ -80,22 +75,15 @@ using thrust_exec_policy_nosync_t =
  */
 class exec_policy_nosync : public thrust_exec_policy_nosync_t {
  public:
+  /**
+   * @brief Construct a new execution policy object
+   *
+   * @param stream The stream on which to allocate temporary memory
+   * @param mr The resource to use for allocating temporary memory
+   */
   explicit exec_policy_nosync(cuda_stream_view stream      = cuda_stream_default,
-                              device_async_resource_ref mr = mr::get_current_device_resource_ref())
-    : thrust_exec_policy_nosync_t(
-        thrust::cuda::par_nosync(mr::thrust_allocator<char>(stream, mr)).on(stream.value()))
-  {
-  }
+                              device_async_resource_ref mr = mr::get_current_device_resource_ref());
 };
 
-#else
-
-using thrust_exec_policy_nosync_t =
-  thrust_exec_policy_t;  ///< When used with Thrust < 1.16.0, thrust_exec_policy_nosync_t is an
-                         ///< alias for thrust_exec_policy_t
-using exec_policy_nosync =
-  exec_policy;  ///< When used with Thrust < 1.16.0, exec_policy_nosync is an alias for exec_policy
-#endif
-
 /** @} */  // end of group
 }  // namespace RMM_NAMESPACE

cpp/include/rmm/logger.hpp

@@ -31,42 +31,20 @@ namespace RMM_NAMESPACE {
  *
  * @return sink_ptr The sink to use
  */
-inline rapids_logger::sink_ptr default_sink()
-{
-  auto* filename = std::getenv("RMM_DEBUG_LOG_FILE");
-  if (filename != nullptr) {
-    return std::make_shared<rapids_logger::basic_file_sink_mt>(filename, true);
-  }
-  return std::make_shared<rapids_logger::stderr_sink_mt>();
-}
+rapids_logger::sink_ptr default_sink();
 
 /**
  * @brief Returns the default log pattern for the global logger.
  *
  * @return std::string The default log pattern.
  */
-inline std::string default_pattern() { return "[%6t][%H:%M:%S:%f][%-6l] %v"; }
+std::string default_pattern();
 
 /**
  * @brief Get the default logger.
  *
  * @return logger& The default logger
  */
-inline rapids_logger::logger& default_logger()
-{
-  static rapids_logger::logger logger_ = [] {
-    rapids_logger::logger logger_{"RMM", {default_sink()}};
-    logger_.set_pattern(default_pattern());
-#if RMM_LOG_ACTIVE_LEVEL <= RMM_LOG_LEVEL_DEBUG
-#ifdef CUDA_API_PER_THREAD_DEFAULT_STREAM
-    logger_.debug("----- RMM LOG [PTDS ENABLED] -----");
-#else
-    logger_.debug("----- RMM LOG [PTDS DISABLED] -----");
-#endif
-#endif
-    return logger_;
-  }();
-  return logger_;
-}
+rapids_logger::logger& default_logger();
 
 }  // namespace RMM_NAMESPACE

cpp/include/rmm/prefetch.hpp

@@ -47,13 +47,7 @@ namespace RMM_NAMESPACE {
 void prefetch(void const* ptr,
               std::size_t size,
               rmm::cuda_device_id device,
-              rmm::cuda_stream_view stream)
-{
-  auto result = cudaMemPrefetchAsync(ptr, size, device.value(), stream.value());
-  // InvalidValue error is raised when non-managed memory is passed to cudaMemPrefetchAsync
-  // We should treat this as a no-op
-  if (result != cudaErrorInvalidValue && result != cudaSuccess) { RMM_CUDA_TRY(result); }
-}
+              rmm::cuda_stream_view stream);
 
 /**
  * @brief Prefetch a span of memory to the specified device on the specified stream.

cpp/src/error.cpp

@@ -0,0 +1,33 @@
+/*
+ * Copyright (c) 2020-2025, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <rmm/error.hpp>
+
+#include <string>
+
+namespace rmm {
+
+bad_alloc::bad_alloc(const char* msg) : _what{std::string{std::bad_alloc::what()} + ": " + msg} {}
+
+bad_alloc::bad_alloc(std::string const& msg) : bad_alloc{msg.c_str()} {}
+
+const char* bad_alloc::what() const noexcept { return _what.c_str(); }
+
+out_of_memory::out_of_memory(const char* msg) : bad_alloc{std::string{"out_of_memory: "} + msg} {}
+
+out_of_memory::out_of_memory(std::string const& msg) : out_of_memory{msg.c_str()} {}
+
+}  // namespace rmm

cpp/src/exec_policy.cpp

@@ -0,0 +1,33 @@
+/*
+ * Copyright (c) 2020-2025, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <rmm/exec_policy.hpp>
+
+namespace rmm {
+
+exec_policy::exec_policy(cuda_stream_view stream, device_async_resource_ref mr)
+  : thrust_exec_policy_t(
+      thrust::cuda::par(mr::thrust_allocator<char>(stream, mr)).on(stream.value()))
+{
+}
+
+exec_policy_nosync::exec_policy_nosync(cuda_stream_view stream, device_async_resource_ref mr)
+  : thrust_exec_policy_nosync_t(
+      thrust::cuda::par_nosync(mr::thrust_allocator<char>(stream, mr)).on(stream.value()))
+{
+}
+
+}  // namespace rmm

cpp/src/logger.cpp

@@ -0,0 +1,46 @@
+/*
+ * Copyright (c) 2025, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <rmm/logger.hpp>
+
+#include <cstdlib>
+#include <memory>
+#include <string>
+
+namespace rmm {
+
+rapids_logger::sink_ptr default_sink()
+{
+  auto* filename = std::getenv("RMM_DEBUG_LOG_FILE");
+  if (filename != nullptr) {
+    return std::make_shared<rapids_logger::basic_file_sink_mt>(filename, true);
+  }
+  return std::make_shared<rapids_logger::stderr_sink_mt>();
+}
+
+std::string default_pattern() { return "[%6t][%H:%M:%S:%f][%-6l] %v"; }
+
+rapids_logger::logger& default_logger()
+{
+  static rapids_logger::logger logger_ = [] {
+    rapids_logger::logger logger_{"RMM", {default_sink()}};
+    logger_.set_pattern(default_pattern());
+    return logger_;
+  }();
+  return logger_;
+}
+
+}  // namespace rmm

cpp/src/prefetch.cpp

@@ -0,0 +1,34 @@
+/*
+ * Copyright (c) 2024-2025, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <rmm/prefetch.hpp>
+
+#include <cuda_runtime_api.h>
+
+namespace rmm {
+
+void prefetch(void const* ptr,
+              std::size_t size,
+              rmm::cuda_device_id device,
+              rmm::cuda_stream_view stream)
+{
+  auto result = cudaMemPrefetchAsync(ptr, size, device.value(), stream.value());
+  // cudaErrorInvalidValue is returned when non-managed memory is passed to
+  // cudaMemPrefetchAsync. We treat this as a no-op.
+  if (result != cudaErrorInvalidValue && result != cudaSuccess) { RMM_CUDA_TRY(result); }
+}
+
+}  // namespace rmm