else
#endif
{
- ggml_sycl_pool_alloc<sycl::half> dst_f16(ctx.pool(), row_diff * src1_ncols);
-
- const sycl::half alpha_f16 = 1.0f;
- const sycl::half beta_f16 = 0.0f;
+ const float alpha = 1.0f;
+ const float beta = 0.0f;
SYCL_CHECK(CHECK_TRY_ERROR(dpct::gemm(
*stream, oneapi::mkl::transpose::trans,
oneapi::mkl::transpose::nontrans, row_diff, src1_ncols, ne10,
- &alpha_f16, src0_ptr, dpct::library_data_t::real_half, ne00,
- src1_ptr, dpct::library_data_t::real_half, ne10, &beta_f16,
- dst_f16.get(), dpct::library_data_t::real_half, ldc,
- dpct::library_data_t::real_half)));
- scope_op_debug_print scope_dbg_print(__func__, "/to_fp32_sycl", dst, /*num_src=*/2,
- " : converting dst to fp32");
- const to_fp32_sycl_t to_fp32_sycl = ggml_get_to_fp32_sycl(GGML_TYPE_F16, dst);
- to_fp32_sycl(dst_f16.get(), dst_dd_i, row_diff*src1_ncols, stream);
+ &alpha, src0_ptr, dpct::library_data_t::real_half, ne00,
+ src1_ptr, dpct::library_data_t::real_half, ne10, &beta,
+ dst_dd_i, dpct::library_data_t::real_float, ldc,
+ dpct::library_data_t::real_float)));
}
} else {
ggml_sycl_pool_alloc<float> src0_ddq_as_f32(ctx.pool());