From bc615548d3f69b21a3602ca2afc37292a2139585 Mon Sep 17 00:00:00 2001 From: slaren Date: Sun, 7 Apr 2024 20:31:09 +0200 Subject: [PATCH] fix windows build --- ggml-cuda.cu | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/ggml-cuda.cu b/ggml-cuda.cu index 297fdbe13..67c04a176 100644 --- a/ggml-cuda.cu +++ b/ggml-cuda.cu @@ -1967,13 +1967,13 @@ struct mmid_row_mapping { static __global__ void k_copy_src1_to_contiguous(const char * __restrict__ src1_original, char * __restrict__ src1_contiguous, int * __restrict__ cur_src1_row, mmid_row_mapping * __restrict__ row_mapping, - const char * ids_dev, int64_t i02, size_t ids_nb1, size_t ids_nb0, + const char * __restrict ids, int64_t i02, size_t ids_nb1, size_t ids_nb0, int64_t ne11, int64_t ne10, size_t nb11, size_t nb12) { int32_t iid1 = blockIdx.x; int32_t id = blockIdx.y; - const int32_t row_id_i = *(const int32_t *) (ids_dev + iid1*ids_nb1 + id*ids_nb0); + const int32_t row_id_i = *(const int32_t *) (ids + iid1*ids_nb1 + id*ids_nb0); if (row_id_i != i02) { return; @@ -2120,7 +2120,7 @@ static void ggml_cuda_mul_mat_id(ggml_backend_cuda_context & ctx, ggml_tensor * CUDA_CHECK(cudaMemsetAsync(dev_cur_src1_row.get(), 0, sizeof(int), stream)); { - dim3 block_dims(std::min((uint)ne10, 512u)); + dim3 block_dims(std::min((unsigned int)ne10, 512u)); dim3 grid_dims(ids->ne[1], n_ids); k_copy_src1_to_contiguous<<>>( src1_original, src1_contiguous.get(), @@ -2151,7 +2151,7 @@ static void ggml_cuda_mul_mat_id(ggml_backend_cuda_context & ctx, ggml_tensor * #ifndef MMID_MEMCPY { - dim3 block_dims(std::min((uint)ne0, 512u)); + dim3 block_dims(std::min((unsigned int)ne0, 512u)); dim3 grid_dims(num_src1_rows); k_copy_dst_from_contiguous<<>>( dst_original, dst_contiguous.get(),