Skip to content

Add cuda 12.8 support #605

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
rapids-bot merged 4 commits into from
Jan 24, 2025
Merged

Add cuda 12.8 support #605

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
4 commits
Select commit Hold shift + click to select a range
33fbe9b
Properly define block size for > sm_100
robertmaynard Oct 11, 2024
5b9e567
Allow CUDA ODR violations
robertmaynard Jan 23, 2025
49a7a85
Update copyright years
robertmaynard Jan 23, 2025
1fefd3a
Correct style issues found by CI
robertmaynard Jan 23, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
7 changes: 6 additions & 1 deletion cpp/cmake/modules/ConfigureCUDA.cmake
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
# =============================================================================
# Copyright (c) 2018-2024, NVIDIA CORPORATION.
# Copyright (c) 2018-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
Expand Down Expand Up @@ -35,6 +35,11 @@ if(CMAKE_COMPILER_IS_GNUCXX)
endif()
endif()

# Allow invalid CUDA kernels in the short term
if(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL 12.8.0)
list(APPEND CUVS_CUDA_FLAGS -static-global-template-stub=false)
endif()

if(CUDA_LOG_COMPILE_TIME)
list(APPEND CUVS_CUDA_FLAGS "--time=nvcc_compile_log.csv")
endif()
Expand Down
5 changes: 3 additions & 2 deletions cpp/src/neighbors/detail/nn_descent.cuh
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/*
* Copyright (c) 2023-2024, NVIDIA CORPORATION.
* Copyright (c) 2023-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.
Expand Down Expand Up @@ -706,7 +706,8 @@ __device__ __forceinline__ void remove_duplicates(
template <typename Index_t, typename ID_t = InternalID_t<Index_t>>
RAFT_KERNEL
#ifdef __CUDA_ARCH__
#if (__CUDA_ARCH__) == 750 || ((__CUDA_ARCH__) >= 860 && (__CUDA_ARCH__) <= 890)
#if (__CUDA_ARCH__) == 750 || ((__CUDA_ARCH__) >= 860 && (__CUDA_ARCH__) <= 890) || \
(__CUDA_ARCH__) == 1200
__launch_bounds__(BLOCK_SIZE)
#else
__launch_bounds__(BLOCK_SIZE, 4)
Expand Down
Loading