Enable all NVCC warnings as errors

cpp/CMakeLists.txt

@@ -177,7 +177,14 @@ message("-- Host target architecture = '${CMAKE_SYSTEM_PROCESSOR}'")
 
 # make the flags global in order to propagate flags to test cmake files
 set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} --expt-relaxed-constexpr --expt-extended-lambda")
-list(APPEND CUOPT_CUDA_FLAGS -Werror=cross-execution-space-call -Wno-deprecated-declarations -Xcompiler=-Werror --default-stream=per-thread)
+list(APPEND CUOPT_CUDA_FLAGS
+        --Werror=all-warnings
+        # Unused declarations are handled in a follow-up change.
+        --diag-suppress=177
+        -Werror=cross-execution-space-call
+        -Wno-deprecated-declarations
+        -Xcompiler=-Werror
+        --default-stream=per-thread)
 if ("${CMAKE_CUDA_HOST_COMPILER}" MATCHES "clang" OR "${CMAKE_CXX_COMPILER_ID}" STREQUAL "Clang")
     list(APPEND CUOPT_CUDA_FLAGS -Xcompiler=-Wall)
 else ()

cpp/src/mip_heuristics/feasibility_jump/load_balancing.cuh

@@ -1,6 +1,6 @@
 /* clang-format off */
 /*
- * SPDX-FileCopyrightText: Copyright (c) 2024-2025, NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+ * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved.
  * SPDX-License-Identifier: Apache-2.0
  */
 /* clang-format on */
@@ -466,8 +466,18 @@ __global__ void load_balancing_mtm_compute_candidates(
 
 // Compute the scores of each candidate move of each variable
 // maximize occupancy for better results, we can afford the extra L1TEX traffic
+#if defined(__CUDA_ARCH__) && \
+  (__CUDA_ARCH__ == 750 || (__CUDA_ARCH__ >= 1100 && __CUDA_ARCH__ < 1200))
+inline constexpr int loadbalance_min_blocks_per_sm = 8;
+#elif defined(__CUDA_ARCH__) && ((__CUDA_ARCH__ >= 860 && __CUDA_ARCH__ < 900) || \
+                                 (__CUDA_ARCH__ >= 1200 && __CUDA_ARCH__ < 1300))
+inline constexpr int loadbalance_min_blocks_per_sm = 12;
+#else
+inline constexpr int loadbalance_min_blocks_per_sm = 16;
+#endif
+
 template <typename i_t, typename f_t>
-__launch_bounds__(TPB_loadbalance, 16) __global__
+__launch_bounds__(TPB_loadbalance, loadbalance_min_blocks_per_sm) __global__
   void load_balancing_mtm_compute_scores(const __grid_constant__
                                          typename fj_t<i_t, f_t>::climber_data_t::view_t fj)
 {

cpp/tests/CMakeLists.txt

@@ -17,8 +17,8 @@ if(BUILD_TESTS AND NOT SKIP_ROUTING_BUILD)
       PRIVATE
       "${CMAKE_CURRENT_SOURCE_DIR}/../src"
       "${CMAKE_CURRENT_SOURCE_DIR}"
-      "${dejavu_SOURCE_DIR}"
   )
+  target_include_directories(cuopttestutils SYSTEM PRIVATE "${dejavu_SOURCE_DIR}")
 
   # cuopttestutils compiles routing/utilities/check_constraints.cu, so it belongs to the
   # routing engine and names it directly. Linking the cuopt umbrella here would hand every
@@ -52,8 +52,8 @@ if(BUILD_TESTS AND NOT SKIP_ROUTING_BUILD)
       PRIVATE
       "${CMAKE_CURRENT_SOURCE_DIR}/../src"
       "${CMAKE_CURRENT_SOURCE_DIR}"
-      "${dejavu_SOURCE_DIR}"
   )
+  target_include_directories(cuopttestutils_static SYSTEM PRIVATE "${dejavu_SOURCE_DIR}")
 
   target_link_libraries(cuopttestutils_static
       PUBLIC
@@ -98,8 +98,8 @@ function(ConfigureTest CMAKE_TEST_NAME)
         "${papilo_SOURCE_DIR}/src"
         "${papilo_BINARY_DIR}"
         "${pslp_SOURCE_DIR}/include"
-        "${dejavu_SOURCE_DIR}"
     )
+    target_include_directories(${CMAKE_TEST_NAME} SYSTEM PRIVATE "${dejavu_SOURCE_DIR}")
     target_link_libraries(${CMAKE_TEST_NAME}
        PRIVATE
         ${_cuopt_link_target}