Skip to content

Commit 113e3d4

Browse files
s-Nicks-Nick
committed
Convert submit and parallel_for to use new extension in convert.cpp
1 parent 6ddef2c commit 113e3d4

File tree

1 file changed

+22
-22
lines changed

1 file changed

+22
-22
lines changed

‎ggml/src/ggml-sycl/convert.cpp

Lines changed: 22 additions & 22 deletions
Original file line numberDiff line numberDiff line change
@@ -171,9 +171,9 @@ static void dequantize_row_q4_K_sycl(const void *vx, dst_t *y, const int64_t k,
171171
dpct::has_capability_or_fail(stream->get_device(),
172172
{sycl::aspect::fp16});
173173

174-
stream->submit([&](sycl::handler &cgh) {
174+
syclex::submit(*stream,[&](sycl::handler &cgh) {
175175
sycl::local_accessor<uint8_t, 1> scale_local_acc(sycl::range<1>(12), cgh);
176-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
176+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
177177
sycl::range<3>(1, 1, 32),
178178
sycl::range<3>(1, 1, 32)),
179179
[=](sycl::nd_item<3> item_ct1) {
@@ -191,10 +191,10 @@ static void dequantize_row_q4_K_sycl_reorder(const void * vx, dst_t * y, const i
191191

192192
dpct::has_capability_or_fail(stream->get_device(), { sycl::aspect::fp16 });
193193

194-
stream->submit([&](sycl::handler & cgh) {
194+
syclex::submit(*stream,[&](sycl::handler & cgh) {
195195
sycl::local_accessor<uint8_t, 1> scale_local_acc(sycl::range<1>(12), cgh);
196196

197-
cgh.parallel_for(sycl::nd_range<1>(sycl::range<1>(global_size), sycl::range<1>(local_size)),
197+
syclex::nd_launch(cgh,sycl::nd_range<1>(sycl::range<1>(global_size), sycl::range<1>(local_size)),
198198
[=](sycl::nd_item<1> item_ct1) {
199199
dequantize_block_q4_K_reorder(vx, y, get_pointer(scale_local_acc), item_ct1, nb);
200200
});
@@ -284,8 +284,8 @@ static void dequantize_row_iq1_s_sycl(const void *vx, dst_t *y, const int64_t k,
284284
dpct::has_capability_or_fail(stream->get_device(),
285285
{sycl::aspect::fp16});
286286

287-
stream->submit([&](sycl::handler &cgh) {
288-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
287+
syclex::submit(*stream,[&](sycl::handler &cgh) {
288+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
289289
sycl::range<3>(1, 1, 32),
290290
sycl::range<3>(1, 1, 32)),
291291
[=](sycl::nd_item<3> item_ct1) {
@@ -305,8 +305,8 @@ static void dequantize_row_iq1_m_sycl(const void *vx, dst_t *y, const int64_t k,
305305
dpct::has_capability_or_fail(stream->get_device(),
306306
{sycl::aspect::fp16});
307307

308-
stream->submit([&](sycl::handler &cgh) {
309-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
308+
syclex::submit(*stream,[&](sycl::handler &cgh) {
309+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
310310
sycl::range<3>(1, 1, 32),
311311
sycl::range<3>(1, 1, 32)),
312312
[=](sycl::nd_item<3> item_ct1) {
@@ -326,8 +326,8 @@ static void dequantize_row_iq2_xxs_sycl(const void *vx, dst_t *y, const int64_t
326326
dpct::has_capability_or_fail(stream->get_device(),
327327
{sycl::aspect::fp16});
328328

329-
stream->submit([&](sycl::handler &cgh) {
330-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
329+
syclex::submit(*stream,[&](sycl::handler &cgh) {
330+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
331331
sycl::range<3>(1, 1, 32),
332332
sycl::range<3>(1, 1, 32)),
333333
[=](sycl::nd_item<3> item_ct1) {
@@ -347,8 +347,8 @@ static void dequantize_row_iq2_xs_sycl(const void *vx, dst_t *y, const int64_t k
347347
dpct::has_capability_or_fail(stream->get_device(),
348348
{sycl::aspect::fp16});
349349

350-
stream->submit([&](sycl::handler &cgh) {
351-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
350+
syclex::submit(*stream,[&](sycl::handler &cgh) {
351+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
352352
sycl::range<3>(1, 1, 32),
353353
sycl::range<3>(1, 1, 32)),
354354
[=](sycl::nd_item<3> item_ct1) {
@@ -368,8 +368,8 @@ static void dequantize_row_iq2_s_sycl(const void *vx, dst_t *y, const int64_t k,
368368
dpct::has_capability_or_fail(stream->get_device(),
369369
{sycl::aspect::fp16});
370370

371-
stream->submit([&](sycl::handler &cgh) {
372-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
371+
syclex::submit(*stream,[&](sycl::handler &cgh) {
372+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
373373
sycl::range<3>(1, 1, 32),
374374
sycl::range<3>(1, 1, 32)),
375375
[=](sycl::nd_item<3> item_ct1) {
@@ -388,8 +388,8 @@ static void dequantize_row_iq3_xxs_sycl(const void *vx, dst_t *y, const int64_t
388388
dpct::has_capability_or_fail(stream->get_device(),
389389
{sycl::aspect::fp16});
390390

391-
stream->submit([&](sycl::handler &cgh) {
392-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
391+
syclex::submit(*stream,[&](sycl::handler &cgh) {
392+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
393393
sycl::range<3>(1, 1, 32),
394394
sycl::range<3>(1, 1, 32)),
395395
[=](sycl::nd_item<3> item_ct1) {
@@ -409,8 +409,8 @@ static void dequantize_row_iq3_s_sycl(const void *vx, dst_t *y, const int64_t k,
409409
dpct::has_capability_or_fail(stream->get_device(),
410410
{sycl::aspect::fp16});
411411

412-
stream->submit([&](sycl::handler &cgh) {
413-
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
412+
syclex::submit(*stream,[&](sycl::handler &cgh) {
413+
syclex::nd_launch(cgh,sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
414414
sycl::range<3>(1, 1, 32),
415415
sycl::range<3>(1, 1, 32)),
416416
[=](sycl::nd_item<3> item_ct1) {
@@ -432,8 +432,8 @@ static void dequantize_row_iq4_xs_sycl(const void *vx, dst_t *y, const int64_t k
432432
dpct::has_capability_or_fail(stream->get_device(),
433433
{sycl::aspect::fp16});
434434

435-
stream->submit([&](sycl::handler &cgh) {
436-
cgh.parallel_for(
435+
syclex::submit(*stream,[&](sycl::handler &cgh) {
436+
syclex::nd_launch(cgh,
437437
sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
438438
sycl::range<3>(1, 1, 32),
439439
sycl::range<3>(1, 1, 32)),
@@ -453,8 +453,8 @@ static void dequantize_row_iq4_nl_sycl(const void *vx, dst_t *y, const int64_t k
453453
dpct::has_capability_or_fail(stream->get_device(),
454454
{sycl::aspect::fp16});
455455

456-
stream->submit([&](sycl::handler &cgh) {
457-
cgh.parallel_for(
456+
syclex::submit(*stream,[&](sycl::handler &cgh) {
457+
syclex::nd_launch(cgh,
458458
sycl::nd_range<3>(sycl::range<3>(1, 1, nb) *
459459
sycl::range<3>(1, 1, 32),
460460
sycl::range<3>(1, 1, 32)),

0 commit comments

Comments
 (0)