From 502ae393fab25f3d0a20f59d950988f9e9199cfc Mon Sep 17 00:00:00 2001 From: Miles Lubin Date: Fri, 18 Sep 2026 20:46:41 -0400 Subject: [PATCH] Enable all NVCC warnings as errors Signed-off-by: Miles Lubin --- cpp/CMakeLists.txt | 9 ++++++++- .../feasibility_jump/load_balancing.cuh | 14 ++++++++++++-- cpp/tests/CMakeLists.txt | 6 +++--- 3 files changed, 23 insertions(+), 6 deletions(-) diff --git a/cpp/CMakeLists.txt b/cpp/CMakeLists.txt index e74ea9f26a..62999e6438 100644 --- a/cpp/CMakeLists.txt +++ b/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 () diff --git a/cpp/src/mip_heuristics/feasibility_jump/load_balancing.cuh b/cpp/src/mip_heuristics/feasibility_jump/load_balancing.cuh index dfc9b3c885..52038d1a67 100644 --- a/cpp/src/mip_heuristics/feasibility_jump/load_balancing.cuh +++ b/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 -__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::climber_data_t::view_t fj) { diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt index 2462149820..41031de5c3 100644 --- a/cpp/tests/CMakeLists.txt +++ b/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}