Fix batched impl

ggerganov · Mar 19, 2024 · 421da29 · 421da29
1 parent b80cf3b
commit 421da29
Showing 1 changed file with 11 additions and 9 deletions.
diff --git a/ggml-sycl.cpp b/ggml-sycl.cpp
@@ -15272,8 +15272,8 @@ static void ggml_sycl_mul_mat_batched_sycl(const ggml_tensor *src0,
  sycl_pool_alloc<sycl::half> dst_f16;
  char * dst_t;
 
- dpct::library_data_t cu_compute_type = dpct::library_data_t::real_float;
- dpct::library_data_t cu_data_type = dpct::library_data_t::real_float;
+ dpct::library_data_t cu_compute_type = dpct::library_data_t::real_half;
+ dpct::library_data_t cu_data_type = dpct::library_data_t::real_half;
 
  // dst strides
  size_t nbd2 = dst->nb[2];
@@ -15282,16 +15282,16 @@ static void ggml_sycl_mul_mat_batched_sycl(const ggml_tensor *src0,
  const sycl::half alpha_f16 = 1.0f;
  const sycl::half beta_f16 = 0.0f;
 
- const float alpha_f32 = 1.0f;
- const float beta_f32 = 0.0f;
-
- const void * alpha = &alpha_f32;
- const void * beta = &beta_f32;
+ const void * alpha = &alpha_f16;
+ const void * beta = &beta_f16;
 
  // TODO: Renable (dst->op_params[0] =! GGML_PREC_DEFAULT) pathway
- // oneMKL open source supports half, half, float, float: datatypes
+ // when oneMKL open source supports half, half, float, float: datatypes
+
+ dst_t = (char *) dst_f16.alloc(ne_dst);
 
- dst_t = (char *) dst_ddf;
+ nbd2 /= sizeof(float) / sizeof(sycl::half);
+ nbd3 /= sizeof(float) / sizeof(sycl::half);
 
  GGML_ASSERT(ne12 % ne02 == 0);
  GGML_ASSERT(ne13 % ne03 == 0);
@@ -15377,6 +15377,8 @@ static void ggml_sycl_mul_mat_batched_sycl(const ggml_tensor *src0,
  }
 #endif
 
+ const to_fp32_sycl_t to_fp32_sycl = ggml_get_to_fp32_sycl(GGML_TYPE_F16);
+ to_fp32_sycl(dst_f16.get(), dst_ddf, ne_dst, main_stream);
 }
 catch (sycl::exception const &exc) {
  std::cerr << exc.what() << "Exception caught at file:" << __FILE__