Warnings are errors (#299)

This PR fixes current RAFT C++/CUDA compilation warnings and turns on -Wall to treat warnings as errors.

Fixes #225
Fixes #289

Authors:
  - Mark Harris (https://github.com/harrism)

Approvers:
  - Brad Rees (https://github.com/BradReesWork)
  - Corey J. Nolet (https://github.com/cjnolet)

URL: #299

cpp/cmake/modules/ConfigureCUDA.cmake

@@ -26,8 +26,8 @@ endif()
 list(APPEND RAFT_CUDA_FLAGS --expt-extended-lambda --expt-relaxed-constexpr)
 
 # set warnings as errors
-# list(APPEND RAFT_CUDA_FLAGS -Werror=cross-execution-space-call)
-# list(APPEND RAFT_CUDA_FLAGS -Xcompiler=-Wall,-Werror,-Wno-error=deprecated-declarations)
+list(APPEND RAFT_CUDA_FLAGS -Werror=cross-execution-space-call)
+list(APPEND RAFT_CUDA_FLAGS -Xcompiler=-Wall,-Werror,-Wno-error=deprecated-declarations)
 
 # Option to enable line info in CUDA device compilation to allow introspection when profiling / memchecking
 if(CUDA_ENABLE_LINEINFO)

cpp/include/raft/distance/pairwise_distance_base.cuh

@@ -20,6 +20,8 @@
 #include <raft/linalg/norm.cuh>
 #include <raft/vectorized.cuh>
 
+#include <cstddef>
+
 namespace raft {
 namespace distance {
 
@@ -312,20 +314,20 @@ __global__ __launch_bounds__(
 }
 
 template <typename P, typename IdxT, typename T>
-dim3 launchConfigGenerator(IdxT m, IdxT n, size_t sMemSize, T func) {
+dim3 launchConfigGenerator(IdxT m, IdxT n, std::size_t sMemSize, T func) {
   const auto numSMs = raft::getMultiProcessorCount();
   int numBlocksPerSm = 0;
   dim3 grid;
 
   CUDA_CHECK(cudaOccupancyMaxActiveBlocksPerMultiprocessor(
     &numBlocksPerSm, func, P::Nthreads, sMemSize));
-  int minGridSize = numSMs * numBlocksPerSm;
-  int yChunks = raft::ceildiv<int>(m, P::Mblk);
-  int xChunks = raft::ceildiv<int>(n, P::Nblk);
+  std::size_t minGridSize = numSMs * numBlocksPerSm;
+  std::size_t yChunks = raft::ceildiv<int>(m, P::Mblk);
+  std::size_t xChunks = raft::ceildiv<int>(n, P::Nblk);
   grid.y = yChunks > minGridSize ? minGridSize : yChunks;
   grid.x = (minGridSize - grid.y) <= 0 ? 1 : xChunks;
   if (grid.x != 1) {
-    int i = 1;
+    std::size_t i = 1;
     while (grid.y * i < minGridSize) {
       i++;
     }
@@ -336,4 +338,4 @@ dim3 launchConfigGenerator(IdxT m, IdxT n, size_t sMemSize, T func) {
 }
 
 };  // namespace distance
-};  // namespace raft
+};  // namespace raft

cpp/include/raft/lap/lap_functions.cuh

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020, NVIDIA CORPORATION.
+ * Copyright (c) 2020-2021, NVIDIA CORPORATION.
  * Copyright 2020 KETAN DATE & RAKESH NAGI
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
@@ -24,19 +24,17 @@
  */
 #pragma once
 
-#include <cuda.h>
-#include <cuda_runtime.h>
-#include <device_launch_parameters.h>
-#include <thrust/device_ptr.h>
-#include <thrust/reduce.h>
-#include <thrust/scan.h>
 #include "d_structs.h"
 
 #include <raft/cudart_utils.h>
 #include <raft/handle.hpp>
+#include <raft/lap/lap_kernels.cuh>
 #include <raft/mr/device/buffer.hpp>
 
-#include <raft/lap/lap_kernels.cuh>
+#include <thrust/reduce.h>
+#include <thrust/scan.h>
+
+#include <cstddef>
 
 namespace raft {
 namespace lap {

cpp/include/raft/lap/lap_kernels.cuh

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020, NVIDIA CORPORATION.
+ * Copyright (c) 2020-2021, NVIDIA CORPORATION.
  * Copyright 2020 KETAN DATE & RAKESH NAGI
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
@@ -24,18 +24,15 @@
  */
 #pragma once
 
-#include <cuda.h>
-#include <cuda_runtime.h>
-#include <device_launch_parameters.h>
-#include <thrust/device_ptr.h>
-#include <thrust/reduce.h>
-#include <thrust/scan.h>
 #include "d_structs.h"
 
 #include <raft/cudart_utils.h>
 #include <raft/handle.hpp>
 #include <raft/mr/device/buffer.hpp>
 
+#include <thrust/for_each.h>
+
+#include <cstddef>
 namespace raft {
 namespace lap {
 namespace detail {

cpp/include/raft/mr/allocator.hpp

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2019-2020, NVIDIA CORPORATION.
+ * Copyright (c) 2019-2021, NVIDIA CORPORATION.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -16,7 +16,9 @@
 
 #pragma once
 
-#include <cuda_runtime.h>
+#include <cuda_runtime_api.h>
+
+#include <cstddef>
 
 namespace raft {
 namespace mr {

cpp/include/raft/mr/buffer_base.hpp

@@ -16,8 +16,11 @@
 
 #pragma once
 
-#include <cuda_runtime.h>
 #include <raft/cudart_utils.h>
+
+#include <cuda_runtime.h>
+
+#include <cstddef>
 #include <memory>
 #include <utility>
 

cpp/include/raft/mr/device/allocator.hpp

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2019-2020, NVIDIA CORPORATION.
+ * Copyright (c) 2019-2021, NVIDIA CORPORATION.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -16,10 +16,12 @@
 
 #pragma once
 
-#include <cstddef>
 #include <raft/mr/allocator.hpp>
+
 #include <rmm/mr/device/per_device_resource.hpp>
 
+#include <cstddef>
+
 namespace raft {
 namespace mr {
 namespace device {

cpp/include/raft/mr/host/allocator.hpp

@@ -16,12 +16,13 @@
 
 #pragma once
 
-#include <cstddef>
-
-#include <cuda_runtime.h>
 #include <raft/cudart_utils.h>
 #include <raft/mr/allocator.hpp>
 
+#include <cuda_runtime.h>
+
+#include <cstddef>
+
 namespace raft {
 namespace mr {
 namespace host {

cpp/include/raft/sparse/hierarchy/detail/agglomerative.cuh

@@ -20,13 +20,16 @@
 #include <raft/cuda_utils.cuh>
 #include <raft/handle.hpp>
 #include <raft/mr/device/buffer.hpp>
+
 #include <rmm/device_uvector.hpp>
 #include <rmm/exec_policy.hpp>
 
 #include <thrust/device_ptr.h>
 #include <thrust/execution_policy.h>
 #include <thrust/sort.h>
 
+#include <cstddef>
+
 namespace raft {
 
 namespace hierarchy {
@@ -97,8 +100,8 @@ class UnionFind {
 template <typename value_idx, typename value_t>
 void build_dendrogram_host(const handle_t &handle, const value_idx *rows,
                            const value_idx *cols, const value_t *data,
-                           size_t nnz, value_idx *children, value_t *out_delta,
-                           value_idx *out_size) {
+                           std::size_t nnz, value_idx *children,
+                           value_t *out_delta, value_idx *out_size) {
   auto d_alloc = handle.get_device_allocator();
   auto stream = handle.get_stream();
 
@@ -120,7 +123,7 @@ void build_dendrogram_host(const handle_t &handle, const value_idx *rows,
 
   UnionFind<value_idx, value_t> U(nnz + 1);
 
-  for (value_idx i = 0; i < nnz; i++) {
+  for (std::size_t i = 0; i < nnz; i++) {
     value_idx a = mst_src_h[i];
     value_idx b = mst_dst_h[i];
     value_t delta = mst_weights_h[i];
@@ -167,7 +170,7 @@ __global__ void write_levels_kernel(const value_idx *children,
  */
 template <typename value_idx>
 __global__ void inherit_labels(const value_idx *children,
-                               const value_idx *levels, size_t n_leaves,
+                               const value_idx *levels, std::size_t n_leaves,
                                value_idx *labels, int cut_level,
                                value_idx n_vertices) {
   value_idx tid = blockDim.x * blockIdx.x + threadIdx.x;
@@ -222,8 +225,8 @@ struct init_label_roots {
  */
 template <typename value_idx, int tpb = 256>
 void extract_flattened_clusters(const raft::handle_t &handle, value_idx *labels,
-                                const value_idx *children, size_t n_clusters,
-                                size_t n_leaves) {
+                                const value_idx *children,
+                                std::size_t n_clusters, std::size_t n_leaves) {
   auto d_alloc = handle.get_device_allocator();
   auto stream = handle.get_stream();
   auto thrust_policy = rmm::exec_policy(rmm::cuda_stream_view{stream});
@@ -241,7 +244,7 @@ void extract_flattened_clusters(const raft::handle_t &handle, value_idx *labels,
      *        out for each of the children
      */
 
-    size_t n_edges = (n_leaves - 1) * 2;
+    auto n_edges = (n_leaves - 1) * 2;
 
     thrust::device_ptr<const value_idx> d_ptr =
       thrust::device_pointer_cast(children);
@@ -250,7 +253,9 @@ void extract_flattened_clusters(const raft::handle_t &handle, value_idx *labels,
 
     // Prevent potential infinite loop from labeling disconnected
     // connectivities graph.
-    RAFT_EXPECTS(n_vertices == (n_leaves - 1) * 2,
+    RAFT_EXPECTS(n_leaves > 0, "n_leaves must be positive");
+    RAFT_EXPECTS(static_cast<std::size_t>(n_vertices) ==
+                   static_cast<std::size_t>((n_leaves - 1) * 2),
                  "Multiple components found in MST or MST is invalid. "
                  "Cannot find single-linkage solution.");
 

cpp/include/raft/sparse/op/sort.h

@@ -16,25 +16,22 @@
 
 #pragma once
 
-#include <cusparse_v2.h>
-
 #include <raft/cudart_utils.h>
 #include <raft/sparse/cusparse_wrappers.h>
+#include <raft/sparse/utils.h>
 #include <raft/cuda_utils.cuh>
 #include <raft/mr/device/allocator.hpp>
 #include <raft/mr/device/buffer.hpp>
+#include <raft/sparse/coo.cuh>
 
 #include <thrust/device_ptr.h>
 #include <thrust/scan.h>
 
+#include <cusparse_v2.h>
+
 #include <cuda_runtime.h>
-#include <stdio.h>
 
 #include <algorithm>
-#include <iostream>
-
-#include <raft/sparse/utils.h>
-#include <raft/sparse/coo.cuh>
 
 namespace raft {
 namespace sparse {
@@ -106,8 +103,6 @@ void coo_sort(COO<T> *const in,
 template <typename value_idx, typename value_t>
 void coo_sort_by_weight(value_idx *rows, value_idx *cols, value_t *data,
                         value_idx nnz, cudaStream_t stream) {
-  thrust::device_ptr<value_idx> t_rows = thrust::device_pointer_cast(rows);
-  thrust::device_ptr<value_idx> t_cols = thrust::device_pointer_cast(cols);
   thrust::device_ptr<value_t> t_data = thrust::device_pointer_cast(data);
 
   auto first = thrust::make_zip_iterator(thrust::make_tuple(rows, cols));
@@ -117,4 +112,4 @@ void coo_sort_by_weight(value_idx *rows, value_idx *cols, value_t *data,
 }
 };  // namespace op
 };  // end NAMESPACE sparse
-};  // end NAMESPACE raft
+};  // end NAMESPACE raft

cpp/test/eigen_solvers.cu

@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020, NVIDIA CORPORATION.
+ * Copyright (c) 2020-2021, NVIDIA CORPORATION.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -14,12 +14,14 @@
  * limitations under the License.
  */
 
+#include <raft/handle.hpp>
+#include <raft/spectral/partition.hpp>
+
 #include <gtest/gtest.h>
+
+#include <cstddef>
 #include <iostream>
 #include <memory>
-#include <raft/handle.hpp>
-
-#include <raft/spectral/partition.hpp>
 
 namespace raft {
 
@@ -37,8 +39,6 @@ TEST(Raft, EigenSolvers) {
   value_type* vs{nullptr};
   index_type nnz = 0;
   index_type nrows = 0;
-  auto stream = h.get_stream();
-  auto t_exe_pol = thrust::cuda::par.on(stream);
 
   sparse_matrix_t<index_type, value_type> sm1{h, ro, ci, vs, nrows, nnz};
   ASSERT_EQ(nullptr, sm1.row_offsets_);
@@ -53,7 +53,7 @@ TEST(Raft, EigenSolvers) {
   //
   value_type* eigvals{nullptr};
   value_type* eigvecs{nullptr};
-  unsigned long long seed{100110021003};
+  std::uint64_t seed{100110021003};
 
   eigen_solver_config_t<index_type, value_type> cfg{
     neigvs, maxiter, restart_iter, tol, reorthog, seed};