summaryrefslogtreecommitdiff
path: root/ggml-cuda.cu
diff options
context:
space:
mode:
authorGeorgi Gerganov <ggerganov@gmail.com>2023-11-02 08:35:10 +0200
committerGitHub <noreply@github.com>2023-11-02 08:35:10 +0200
commit4d719a6d4e74b9a98e75f826f865f3153717d54b (patch)
treef27bd8dc5d99548648f7521584727d39009ae506 /ggml-cuda.cu
parent183b3fac6c28e65d23ac0230c1dd6fb84bf0154d (diff)
cuda : check if this fixes Pascal card regression (#3882)
Diffstat (limited to 'ggml-cuda.cu')
-rw-r--r--ggml-cuda.cu2
1 files changed, 1 insertions, 1 deletions
diff --git a/ggml-cuda.cu b/ggml-cuda.cu
index 57a528ed..e4629512 100644
--- a/ggml-cuda.cu
+++ b/ggml-cuda.cu
@@ -7420,7 +7420,7 @@ static void ggml_cuda_mul_mat(const ggml_tensor * src0, const ggml_tensor * src1
} else if (all_on_device && !use_tensor_cores && src0->type == GGML_TYPE_F16 && !ggml_is_contiguous(src0) && !ggml_is_transposed(src1) && src1->ne[1] == 1) {
// KQV single-batch
ggml_cuda_mul_mat_vec_nc(src0, src1, dst);
- } else if (all_on_device && src0->type == GGML_TYPE_F16 && src1->type == GGML_TYPE_F32 && !ggml_is_transposed(src0) && !ggml_is_transposed(src1)) {
+ } else if (all_on_device && use_tensor_cores && src0->type == GGML_TYPE_F16 && src1->type == GGML_TYPE_F32 && !ggml_is_transposed(src0) && !ggml_is_transposed(src1)) {
// KQ + KQV multi-batch
ggml_cuda_mul_mat_mat_batched_cublas(src0, src1, dst);
} else if (src0->type == GGML_TYPE_F32) {