@@ -418,7 +418,7 @@ static void acc_f32_sycl(const float *x, const float *y, float *dst,
418418}
419419
420420template <typename T>
421- static void arange_kernel (T * dst, const int k, T start, T step,
421+ static void arange_kernel (T * dst, const int k, T start, T step,
422422 const sycl::nd_item<1 > &item_ct1) {
423423 SYCL_GLOBAL_ID_LOOP (k, item_ct1) {
424424 dst[i] = start + static_cast <T>(i) * step;
@@ -639,21 +639,16 @@ static inline void dispatch_ggml_sycl_op_pad(ggml_backend_sycl_context & ctx, gg
639639 }
640640}
641641
642- // ב-namespace ggml_sycl_detail:
643642static inline void ggml_sycl_op_arange (ggml_backend_sycl_context & ctx, ggml_tensor * dst) {
644643 GGML_ASSERT (dst->type == GGML_TYPE_F32);
645-
646644 float start, stop, step;
647645 memcpy (&start, dst->op_params , sizeof (float ));
648646 memcpy (&stop, (float *) dst->op_params + 1 , sizeof (float ));
649647 memcpy (&step, (float *) dst->op_params + 2 , sizeof (float ));
650-
651648 dpct::queue_ptr stream = ctx.stream ();
652649 SYCL_CHECK (ggml_sycl_set_device (ctx.device ));
653-
654650 float * dst_ptr = (float *)dst->data ;
655- const int k = (int )ggml_nelements (dst); // הוספה חשובה!
656-
651+ const int k = (int )ggml_nelements (dst);
657652 const int num_blocks = ceil_div (k, SYCL_ARANGE_BLOCK_SIZE);
658653 stream->parallel_for (
659654 sycl::nd_range<1 >(sycl::range<1 >(num_blocks) * sycl::range<1 >(SYCL_ARANGE_BLOCK_SIZE),
@@ -1204,4 +1199,4 @@ void ggml_sycl_geglu_quick(ggml_backend_sycl_context & ctx, ggml_tensor * dst) {
12041199void ggml_sycl_arange (ggml_backend_sycl_context & ctx, ggml_tensor * dst) {
12051200 scope_op_debug_print scope_dbg_print (__func__, dst, /* num_src=*/ 0 );
12061201 ggml_sycl_detail::ggml_sycl_op_arange (ctx, dst);
1207- }
1202+ }
0 commit comments