Skip to content

support CUDA 13.1, SM 12.0 #3696

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

Draft
ktdq wants to merge 7 commits into
base: master
Choose a base branch
from
Draft

support CUDA 13.1, SM 12.0 #3696

Show file tree
Hide file tree
Changes from 4 commits
Commits
Show all changes
7 commits
Select commit Hold shift + click to select a range
186342d
Allow multi-digit CUDA compute capabilities
ktdq Dec 5, 2025
a177500
Merge pull request #1 from ktdq/codex/fix-arrayfire-for-cuda-13
ktdq Dec 5, 2025
6fcbc28
Handle optional cuFFT enums and use queried clock rate
ktdq Dec 6, 2025
1b5b130
Update CMakeLists.txt
ktdq Dec 7, 2025
5d00837
Merge branch 'master' into codex/fix-arrayfire-for-cuda-13-9tn2hx
ktdq Dec 7, 2025
5f487cc
Merge pull request #2 from ktdq/codex/fix-arrayfire-for-cuda-13-9tn2hx
ktdq Dec 7, 2025
9d3ea85
support more cuda arch
ktdq Dec 7, 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
2 changes: 1 addition & 1 deletion src/backend/cuda/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -859,7 +859,7 @@ if(AF_INSTALL_STANDALONE)

if(WIN32 OR NOT AF_WITH_STATIC_CUDA_NUMERIC_LIBS)
if(CUDA_VERSION_MAJOR VERSION_EQUAL 13)
afcu_collect_libs(cufft LIB_MAJOR 12 LIB_MINOR 0)
afcu_collect_libs(cufft LIB_MAJOR 12 LIB_MINOR 1)
elseif(CUDA_VERSION_MAJOR VERSION_EQUAL 12)
afcu_collect_libs(cufft LIB_MAJOR 11 LIB_MINOR 3)
elseif(CUDA_VERSION_MAJOR VERSION_EQUAL 11)
Expand Down
6 changes: 6 additions & 0 deletions src/backend/cuda/cufft.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -38,19 +38,25 @@ const char *_cufftGetResultString(cufftResult res) {

case CUFFT_UNALIGNED_DATA: return "cuFFT: unaligned data (deprecated)";

#ifdef CUFFT_INCOMPLETE_PARAMETER_LIST
case CUFFT_INCOMPLETE_PARAMETER_LIST:
return "cuFFT: call is missing parameters";
#endif

case CUFFT_INVALID_DEVICE:
return "cuFFT: plan execution different than plan creation";

#ifdef CUFFT_PARSE_ERROR
case CUFFT_PARSE_ERROR: return "cuFFT: plan parse error";
#endif

case CUFFT_NO_WORKSPACE: return "cuFFT: no workspace provided";

case CUFFT_NOT_IMPLEMENTED: return "cuFFT: not implemented";

#ifdef CUFFT_LICENSE_ERROR
case CUFFT_LICENSE_ERROR: return "cuFFT: license error";
#endif

#if CUDA_VERSION >= 8000
case CUFFT_NOT_SUPPORTED: return "cuFFT: not supported";
Expand Down
6 changes: 5 additions & 1 deletion src/backend/cuda/device_manager.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -598,9 +598,13 @@ DeviceManager::DeviceManager()
AF_TRACE("Unsuppored device: {}", dev.prop.name);
continue;
} else {
int clock_rate_khz = 0;
CUDA_CHECK(
cudaDeviceGetAttribute(&clock_rate_khz, cudaDevAttrClockRate,
i));
dev.flops = static_cast<size_t>(dev.prop.multiProcessorCount) *
compute2cores(dev.prop.major, dev.prop.minor) *
dev.prop.clockRate;
static_cast<size_t>(clock_rate_khz);
dev.nativeId = i;
AF_TRACE(
"Found device: {} (sm_{}{}) ({:0.3} GB | ~{} GFLOPs | {} "
Expand Down
2 changes: 1 addition & 1 deletion src/backend/cuda/kernel/regions.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -341,7 +341,7 @@ __global__ static void update_equiv(arrayfire::cuda::Param<T> equiv_map,
}

template<typename T>
struct clamp_to_one : public thrust::unary_function<T, T> {
struct clamp_to_one {
__host__ __device__ T operator()(const T& in) const {
return (in >= (T)1) ? (T)1 : in;
}
Expand Down