|
301 | 301 | <div class="container">
|
302 | 302 | <div class="col">
|
303 | 303 | <code><pre>
|
304 |
| -cgh.<mark>parallel_for</mark><my_kernel>(<mark>nd_range{{1024, 16}, {32, 4}}</mark>, |
305 |
| - [=](<mark>nd_item<2> item)</mark>{ |
| 304 | +cgh.<mark>parallel_for</mark><my_kernel>(<mark>sycl::nd_range{{1024, 16}, {32, 4}}</mark>, |
| 305 | + [=](<mark>sycl::nd_item<2> item)</mark>{ |
306 | 306 | // SYCL kernel function is executed
|
307 | 307 | // on a range of work-items
|
308 | 308 | });
|
|
329 | 329 | <div style="font-size: 90%; display: grid; grid-template-columns: 45% 55%; grid-template-rows: 1fr 1fr 1fr;">
|
330 | 330 | <div style="margin: auto 0; vertical-align: middle;">
|
331 | 331 | <code><pre>
|
332 |
| -cgh.parallel_for<kernel>((<mark>nd_range<1>{1024,32}</mark>, |
333 |
| - [=](<mark>nd_item<1> ndItem</mark>){ |
| 332 | +cgh.parallel_for<kernel>((<mark>sycl::nd_range<1>{1024,32}</mark>, |
| 333 | + [=](<mark>sycl::nd_item<1> ndItem</mark>){ |
334 | 334 | /* kernel function code */
|
335 | 335 | id globalId = ndItem.get_global_id();
|
336 | 336 | id localId = ndItem.get_local_id();
|
|
343 | 343 | </div>
|
344 | 344 | <div style="margin: auto 0; vertical-align: middle;">
|
345 | 345 | <code><pre>
|
346 |
| -cgh.parallel_for<kernel>(<mark>range<1>{1024}</mark>, |
347 |
| - [=](<mark>item<1> item</mark>){ |
| 346 | +cgh.parallel_for<kernel>(<mark>sycl::range<1>{1024}</mark>, |
| 347 | + [=](<mark>sycl::item<1> item</mark>){ |
348 | 348 | /* kernel function code */
|
349 | 349 | id globalId = item.get_id();
|
350 | 350 | });
|
|
356 | 356 | </div>
|
357 | 357 | <div style="margin: auto 0; vertical-align: middle;">
|
358 | 358 | <code><pre>
|
359 |
| -cgh.parallel_for<kernel>(<mark>range<1>{1024}</mark>, |
360 |
| - [=](<mark>id<1> globalId</mark>){ |
| 359 | +cgh.parallel_for<kernel>(<mark>sycl::range<1>{1024}</mark>, |
| 360 | + [=](<mark>sycl::id<1> globalId</mark>){ |
361 | 361 | /* kernel function code */
|
362 | 362 | });
|
363 | 363 | </code></pre>
|
|
455 | 455 | <div class="container">
|
456 | 456 | <div class="col-left-3">
|
457 | 457 | <code><pre>
|
458 |
| -buffer<float, 1> bufA(dA.data(), range<1>(dA.size())); |
459 |
| -buffer<float, 1> bufB(dB.data(), range<1>(dB.size())); |
460 |
| -buffer<float, 1> bufO(dO.data(), range<1>(dO.size())); |
| 458 | +sycl::buffer<float, 1> bufA(dA.data(), sycl::range<1>(dA.size())); |
| 459 | +sycl::buffer<float, 1> bufB(dB.data(), sycl::range<1>(dB.size())); |
| 460 | +sycl::buffer<float, 1> bufO(dO.data(), sycl::range<1>(dO.size())); |
461 | 461 |
|
462 |
| -gpuQueue.submit([&](handler &cgh){ |
| 462 | +gpuQueue.submit([&](sycl::handler &cgh){ |
463 | 463 | sycl::accessor inA{bufA, cgh, sycl::read_only};
|
464 | 464 | sycl::accessor inB{bufB, cgh, sycl::read_only};
|
465 | 465 | sycl::accessor out{bufO, cgh, sycl::write_only};
|
466 |
| - cgh.parallel_for<add>(range<1>(dA.size()), |
467 |
| - [=](id<1> i){ |
| 466 | + cgh.parallel_for<add>(sycl::range<1>(dA.size()), |
| 467 | + [=](sycl::id<1> i){ |
468 | 468 | <mark>out[i] = inA[i] + inB[i];</mark>
|
469 | 469 | });
|
470 | 470 | });
|
|
485 | 485 | <div class="container">
|
486 | 486 | <div class="col-left-3">
|
487 | 487 | <code><pre>
|
488 |
| -buffer<float, 1> bufA(dA.data(), range<1>(dA.size())); |
489 |
| -buffer<float, 1> bufB(dB.data(), range<1>(dB.size())); |
490 |
| -buffer<float, 1> bufO(dO.data(), range<1>(dO.size())); |
| 488 | +sycl::buffer<float, 1> bufA(dA.data(), sycl::range<1>(dA.size())); |
| 489 | +sycl::buffer<float, 1> bufB(dB.data(), sycl::range<1>(dB.size())); |
| 490 | +sycl::buffer<float, 1> bufO(dO.data(), sycl::range<1>(dO.size())); |
491 | 491 |
|
492 | 492 | gpuQueue.submit([&](handler &cgh){
|
493 | 493 | sycl::accessor inA{bufA, cgh, sycl::read_only};
|
494 | 494 | sycl::accessor inB{bufB, cgh, sycl::read_only};
|
495 | 495 | sycl::accessor out{bufO, cgh, sycl::write_only};
|
496 |
| - cgh.parallel_for<add>(rng, [=](item<3> i){ |
| 496 | + cgh.parallel_for<add>(rng, [=](sycl::item<3> i){ |
497 | 497 | <mark>auto ptrA = inA.get_pointer();</mark>
|
498 | 498 | <mark>auto ptrB = inB.get_pointer();</mark>
|
499 | 499 | <mark>auto ptrO = out.get_pointer();</mark>
|
|
0 commit comments