Skip to content

Commit a33e305

Browse files
committed
Using prop.major and prop.minor, include HIP and MUSA
1 parent 02fc26c commit a33e305

File tree

1 file changed

+3
-2
lines changed

1 file changed

+3
-2
lines changed

ggml/src/ggml-cuda/ggml-cuda.cu

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -230,6 +230,8 @@ static ggml_cuda_device_info ggml_cuda_init() {
230230
cudaDeviceProp prop;
231231
CUDA_CHECK(cudaGetDeviceProperties(&prop, id));
232232

233+
is_cc121 |= prop.major == 12 && prop.minor == 1;
234+
233235
info.default_tensor_split[id] = total_vram;
234236
total_vram += prop.totalGlobalMem;
235237
info.devices[id].integrated = false; // Temporarily disabled due to issues with corrupted output (e.g. #15034)
@@ -275,8 +277,6 @@ static ggml_cuda_device_info ggml_cuda_init() {
275277
turing_devices_without_mma.push_back({ id, device_name });
276278
}
277279

278-
is_cc121 |= info.devices[id].cc == 1210;
279-
280280
#endif // defined(GGML_USE_HIP)
281281
}
282282

@@ -298,6 +298,7 @@ static ggml_cuda_device_info ggml_cuda_init() {
298298
// CUBLAS_CHECK(cublasLoggerConfigure(1, 1, 0, nullptr));
299299

300300
// Setting device scheduling strategy for iGPUs to "spinning" to avoid delays in cuda synchronize calls.
301+
// This fix is temporary, as the strategy will be the default in later drivers.
301302
if (is_cc121) {
302303
CUDA_CHECK(cudaSetDeviceFlags(cudaDeviceScheduleSpin));
303304
}

0 commit comments

Comments
 (0)