Skip to content

Commit e64a40c

Browse files
v-klochkovv-klochkov
authored
[ESIMD][NFC] Replace the deprecated feature_not_supported exception (#5861)
Signed-off-by: Vyacheslav N Klochkov <[email protected]>
1 parent 02b32b5 commit e64a40c

File tree

6 files changed

+46
-44
lines changed

6 files changed

+46
-44
lines changed

‎sycl/include/sycl/ext/intel/esimd/common.hpp

Lines changed: 3 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -48,7 +48,9 @@
4848

4949
#define __ESIMD_API ESIMD_INLINE
5050

51-
#define __ESIMD_UNSUPPORTED_ON_HOST throw cl::sycl::feature_not_supported()
51+
#define __ESIMD_UNSUPPORTED_ON_HOST \
52+
throw sycl::exception(sycl::errc::feature_not_supported, \
53+
"This ESIMD feature is not supported on HOST")
5254

5355
#endif // __SYCL_DEVICE_ONLY__
5456

‎sycl/include/sycl/ext/intel/esimd/detail/atomic_intrin.hpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -16,7 +16,7 @@
1616
template <typename Ty> Ty atomic_add_fetch(Ty *ptr, Ty val) {
1717
#ifdef _WIN32
1818
// TODO: Windows will be supported soon
19-
throw cl::sycl::feature_not_supported();
19+
__ESIMD_UNSUPPORTED_ON_HOST;
2020
#else
2121
return __atomic_add_fetch(ptr, val, __ATOMIC_RELAXED);
2222
#endif

‎sycl/include/sycl/ext/intel/esimd/detail/elem_type_traits.hpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -167,7 +167,7 @@ __esimd_convertvector_to(vector_type_t<StdTy, N> Val)
167167
#else
168168
{
169169
// TODO implement for host
170-
throw sycl::feature_not_supported();
170+
__ESIMD_UNSUPPORTED_ON_HOST;
171171
}
172172
#endif // __SYCL_DEVICE_ONLY__
173173

@@ -180,7 +180,7 @@ __esimd_convertvector_from(vector_type_t<__raw_t<WrapperTy>, N> Val)
180180
#else
181181
{
182182
// TODO implement for host
183-
throw sycl::feature_not_supported();
183+
__ESIMD_UNSUPPORTED_ON_HOST;
184184
}
185185
#endif // __SYCL_DEVICE_ONLY__
186186

‎sycl/include/sycl/ext/intel/esimd/detail/memory_intrin.hpp

Lines changed: 8 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -378,7 +378,7 @@ __esimd_gather_scaled2(SurfIndAliasTy surf_ind, uint32_t global_offset,
378378
static_assert(N == 1 || N == 8 || N == 16 || N == 32);
379379
static_assert(TySizeLog2 <= 2 && Scale == 0);
380380
static_assert(std::is_integral<Ty>::value || TySizeLog2 == 2);
381-
throw cl::sycl::feature_not_supported();
381+
__ESIMD_UNSUPPORTED_ON_HOST;
382382
}
383383
#endif // __SYCL_DEVICE_ONLY__
384384

@@ -469,7 +469,7 @@ __esimd_svm_atomic0(__ESIMD_DNS::vector_type_t<uint64_t, N> addrs,
469469
;
470470
#else
471471
{
472-
throw cl::sycl::feature_not_supported();
472+
__ESIMD_UNSUPPORTED_ON_HOST;
473473
}
474474
#endif // __SYCL_DEVICE_ONLY__
475475

@@ -493,7 +493,7 @@ __esimd_svm_atomic1(__ESIMD_DNS::vector_type_t<uint64_t, N> addrs,
493493
retv[i] = atomic_add_fetch<Ty>(p, src0[i]);
494494
break;
495495
default:
496-
throw cl::sycl::feature_not_supported();
496+
__ESIMD_UNSUPPORTED_ON_HOST;
497497
}
498498
}
499499
}
@@ -512,7 +512,7 @@ __esimd_svm_atomic2(__ESIMD_DNS::vector_type_t<uint64_t, N> addrs,
512512
;
513513
#else
514514
{
515-
throw cl::sycl::feature_not_supported();
515+
__ESIMD_UNSUPPORTED_ON_HOST;
516516
}
517517
#endif // __SYCL_DEVICE_ONLY__
518518

@@ -828,12 +828,12 @@ __esimd_dword_atomic0(__ESIMD_DNS::simd_mask_storage_t<N> pred,
828828
retv[i] = atomic_add_fetch<Ty>(p, 1);
829829
break;
830830
default:
831-
throw cl::sycl::feature_not_supported();
831+
__ESIMD_UNSUPPORTED_ON_HOST;
832832
}
833833
}
834834
}
835835
} else {
836-
throw cl::sycl::feature_not_supported();
836+
__ESIMD_UNSUPPORTED_ON_HOST;
837837
}
838838
return retv;
839839
}
@@ -849,7 +849,7 @@ __esimd_dword_atomic1(__ESIMD_DNS::simd_mask_storage_t<N> pred,
849849
;
850850
#else
851851
{
852-
throw cl::sycl::feature_not_supported();
852+
__ESIMD_UNSUPPORTED_ON_HOST;
853853
}
854854
#endif // __SYCL_DEVICE_ONLY__
855855

@@ -864,7 +864,7 @@ __esimd_dword_atomic2(__ESIMD_DNS::simd_mask_storage_t<N> pred,
864864
;
865865
#else
866866
{
867-
throw cl::sycl::feature_not_supported();
867+
__ESIMD_UNSUPPORTED_ON_HOST;
868868
}
869869
#endif // __SYCL_DEVICE_ONLY__
870870

‎sycl/include/sycl/ext/intel/experimental/esimd/detail/math_intrin.hpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -668,7 +668,7 @@ __esimd_dpas(__ESIMD_DNS::vector_type_t<T0, N> src0,
668668
repeat_count, T, T1, T2, N, N1, N2>(
669669
std::addressof(src0), src1, src2);
670670
#else // __SYCL_EXPLICIT_SIMD_PLUGIN__
671-
throw cl::sycl::feature_not_supported();
671+
__ESIMD_UNSUPPORTED_ON_HOST;
672672
return __ESIMD_DNS::vector_type_t<T, N>();
673673
#endif // __SYCL_EXPLICIT_SIMD_PLUGIN__
674674
}
@@ -685,7 +685,7 @@ __esimd_dpas2(__ESIMD_DNS::vector_type_t<T1, N1> src1,
685685
repeat_count, T, T1, T2, N, N1, N2>(nullptr, src1,
686686
src2);
687687
#else // __SYCL_EXPLICIT_SIMD_PLUGIN__
688-
throw cl::sycl::feature_not_supported();
688+
__ESIMD_UNSUPPORTED_ON_HOST;
689689
return __ESIMD_DNS::vector_type_t<T, N>();
690690
#endif // __SYCL_EXPLICIT_SIMD_PLUGIN__
691691
}
@@ -698,7 +698,7 @@ inline __ESIMD_DNS::vector_type_t<T, N>
698698
__esimd_dpasw(__ESIMD_DNS::vector_type_t<T, N> src0,
699699
__ESIMD_DNS::vector_type_t<T1, N1> src1,
700700
__ESIMD_DNS::vector_type_t<T2, N2> src2) {
701-
throw cl::sycl::feature_not_supported();
701+
__ESIMD_UNSUPPORTED_ON_HOST;
702702
return __ESIMD_DNS::vector_type_t<T, N>();
703703
}
704704

@@ -709,7 +709,7 @@ template <__ESIMD_ENS::argument_type src1_precision,
709709
inline __ESIMD_DNS::vector_type_t<T, N>
710710
__esimd_dpasw2(__ESIMD_DNS::vector_type_t<T1, N1> src1,
711711
__ESIMD_DNS::vector_type_t<T2, N2> src2) {
712-
throw cl::sycl::feature_not_supported();
712+
__ESIMD_UNSUPPORTED_ON_HOST;
713713
return __ESIMD_DNS::vector_type_t<T, N>();
714714
}
715715

‎sycl/include/sycl/ext/intel/experimental/esimd/detail/memory_intrin.hpp

Lines changed: 28 additions & 28 deletions
Original file line numberDiff line numberDiff line change
@@ -69,7 +69,7 @@ __esimd_raw_sends2(uint8_t modifier, uint8_t execSize,
6969
;
7070
#else
7171
{
72-
throw cl::sycl::feature_not_supported();
72+
__ESIMD_UNSUPPORTED_ON_HOST;
7373
}
7474
#endif // __SYCL_DEVICE_ONLY__
7575

@@ -110,7 +110,7 @@ __esimd_raw_send2(uint8_t modifier, uint8_t execSize,
110110
;
111111
#else
112112
{
113-
throw cl::sycl::feature_not_supported();
113+
__ESIMD_UNSUPPORTED_ON_HOST;
114114
}
115115
#endif // __SYCL_DEVICE_ONLY__
116116

@@ -150,7 +150,7 @@ __esimd_raw_sends2_noresult(uint8_t modifier, uint8_t execSize,
150150
;
151151
#else
152152
{
153-
throw cl::sycl::feature_not_supported();
153+
__ESIMD_UNSUPPORTED_ON_HOST;
154154
}
155155
#endif // __SYCL_DEVICE_ONLY__
156156

@@ -184,7 +184,7 @@ __esimd_raw_send2_noresult(uint8_t modifier, uint8_t execSize,
184184
;
185185
#else
186186
{
187-
throw cl::sycl::feature_not_supported();
187+
__ESIMD_UNSUPPORTED_ON_HOST;
188188
}
189189
#endif // __SYCL_DEVICE_ONLY__
190190

@@ -202,7 +202,7 @@ __ESIMD_INTRIN void __esimd_nbarrier(uint8_t mode, uint8_t id,
202202
;
203203
#else // __SYCL_DEVICE_ONLY__
204204
{
205-
throw cl::sycl::feature_not_supported();
205+
__ESIMD_UNSUPPORTED_ON_HOST;
206206
}
207207
#endif // __SYCL_DEVICE_ONLY__
208208

@@ -215,7 +215,7 @@ __ESIMD_INTRIN void __esimd_nbarrier_init(uint8_t count)
215215
;
216216
#else // __SYCL_DEVICE_ONLY__
217217
{
218-
throw cl::sycl::feature_not_supported();
218+
__ESIMD_UNSUPPORTED_ON_HOST;
219219
}
220220
#endif // __SYCL_DEVICE_ONLY__
221221

@@ -242,7 +242,7 @@ __ESIMD_INTRIN void __esimd_raw_send_nbarrier_signal(
242242
;
243243
#else // __SYCL_DEVICE_ONLY__
244244
{
245-
throw cl::sycl::feature_not_supported();
245+
__ESIMD_UNSUPPORTED_ON_HOST;
246246
}
247247
#endif // __SYCL_DEVICE_ONLY__
248248

@@ -275,7 +275,7 @@ __esimd_lsc_load_slm(__ESIMD_DNS::simd_mask_storage_t<N> pred,
275275
;
276276
#else // __SYCL_DEVICE_ONLY__
277277
{
278-
throw cl::sycl::feature_not_supported();
278+
__ESIMD_UNSUPPORTED_ON_HOST;
279279
return 0;
280280
}
281281
#endif // __SYCL_DEVICE_ONLY__
@@ -313,7 +313,7 @@ __esimd_lsc_load_bti(__ESIMD_DNS::simd_mask_storage_t<N> pred,
313313
;
314314
#else // __SYCL_DEVICE_ONLY__
315315
{
316-
throw cl::sycl::feature_not_supported();
316+
__ESIMD_UNSUPPORTED_ON_HOST;
317317
return 0;
318318
}
319319
#endif // __SYCL_DEVICE_ONLY__
@@ -347,7 +347,7 @@ __esimd_lsc_load_stateless(__ESIMD_DNS::simd_mask_storage_t<N> pred,
347347
;
348348
#else // __SYCL_DEVICE_ONLY__
349349
{
350-
throw cl::sycl::feature_not_supported();
350+
__ESIMD_UNSUPPORTED_ON_HOST;
351351
return 0;
352352
}
353353
#endif // __SYCL_DEVICE_ONLY__
@@ -383,7 +383,7 @@ __esimd_lsc_prefetch_bti(__ESIMD_DNS::simd_mask_storage_t<N> pred,
383383
;
384384
#else // __SYCL_DEVICE_ONLY__
385385
{
386-
throw cl::sycl::feature_not_supported();
386+
__ESIMD_UNSUPPORTED_ON_HOST;
387387
}
388388
#endif // __SYCL_DEVICE_ONLY__
389389

@@ -414,7 +414,7 @@ __esimd_lsc_prefetch_stateless(__ESIMD_DNS::simd_mask_storage_t<N> pred,
414414
;
415415
#else // __SYCL_DEVICE_ONLY__
416416
{
417-
throw cl::sycl::feature_not_supported();
417+
__ESIMD_UNSUPPORTED_ON_HOST;
418418
}
419419
#endif // __SYCL_DEVICE_ONLY__
420420

@@ -447,7 +447,7 @@ __ESIMD_INTRIN void __esimd_lsc_store_slm(
447447
;
448448
#else // __SYCL_DEVICE_ONLY__
449449
{
450-
throw cl::sycl::feature_not_supported();
450+
__ESIMD_UNSUPPORTED_ON_HOST;
451451
}
452452
#endif // __SYCL_DEVICE_ONLY__
453453

@@ -484,7 +484,7 @@ __ESIMD_INTRIN void __esimd_lsc_store_bti(
484484
;
485485
#else // __SYCL_DEVICE_ONLY__
486486
{
487-
throw cl::sycl::feature_not_supported();
487+
__ESIMD_UNSUPPORTED_ON_HOST;
488488
}
489489
#endif // __SYCL_DEVICE_ONLY__
490490

@@ -517,7 +517,7 @@ __ESIMD_INTRIN void __esimd_lsc_store_stateless(
517517
;
518518
#else // __SYCL_DEVICE_ONLY__
519519
{
520-
throw cl::sycl::feature_not_supported();
520+
__ESIMD_UNSUPPORTED_ON_HOST;
521521
}
522522
#endif // __SYCL_DEVICE_ONLY__
523523

@@ -563,7 +563,7 @@ __esimd_lsc_load2d_stateless(__ESIMD_DNS::simd_mask_storage_t<N> Pred,
563563
;
564564
#else // __SYCL_DEVICE_ONLY__
565565
{
566-
throw cl::sycl::feature_not_supported();
566+
__ESIMD_UNSUPPORTED_ON_HOST;
567567
return 0;
568568
}
569569
#endif // __SYCL_DEVICE_ONLY__
@@ -603,7 +603,7 @@ __ESIMD_INTRIN void __esimd_lsc_prefetch2d_stateless(
603603
;
604604
#else // __SYCL_DEVICE_ONLY__
605605
{
606-
throw cl::sycl::feature_not_supported();
606+
__ESIMD_UNSUPPORTED_ON_HOST;
607607
}
608608
#endif // __SYCL_DEVICE_ONLY__
609609

@@ -649,7 +649,7 @@ __esimd_lsc_store2d_stateless(__ESIMD_DNS::simd_mask_storage_t<N> Pred,
649649
;
650650
#else // __SYCL_DEVICE_ONLY__
651651
{
652-
throw cl::sycl::feature_not_supported();
652+
__ESIMD_UNSUPPORTED_ON_HOST;
653653
}
654654
#endif // __SYCL_DEVICE_ONLY__
655655

@@ -680,7 +680,7 @@ __esimd_lsc_xatomic_slm_0(__ESIMD_DNS::simd_mask_storage_t<N> pred,
680680
;
681681
#else // __SYCL_DEVICE_ONLY__
682682
{
683-
throw cl::sycl::feature_not_supported();
683+
__ESIMD_UNSUPPORTED_ON_HOST;
684684
return 0;
685685
}
686686
#endif // __SYCL_DEVICE_ONLY__
@@ -715,7 +715,7 @@ __esimd_lsc_xatomic_slm_1(
715715
;
716716
#else // __SYCL_DEVICE_ONLY__
717717
{
718-
throw cl::sycl::feature_not_supported();
718+
__ESIMD_UNSUPPORTED_ON_HOST;
719719
return 0;
720720
}
721721
#endif // __SYCL_DEVICE_ONLY__
@@ -752,7 +752,7 @@ __esimd_lsc_xatomic_slm_2(
752752
;
753753
#else // __SYCL_DEVICE_ONLY__
754754
{
755-
throw cl::sycl::feature_not_supported();
755+
__ESIMD_UNSUPPORTED_ON_HOST;
756756
return 0;
757757
}
758758
#endif // __SYCL_DEVICE_ONLY__
@@ -787,7 +787,7 @@ __esimd_lsc_xatomic_bti_0(__ESIMD_DNS::simd_mask_storage_t<N> pred,
787787
;
788788
#else // __SYCL_DEVICE_ONLY__
789789
{
790-
throw cl::sycl::feature_not_supported();
790+
__ESIMD_UNSUPPORTED_ON_HOST;
791791
return 0;
792792
}
793793
#endif // __SYCL_DEVICE_ONLY__
@@ -825,7 +825,7 @@ __esimd_lsc_xatomic_bti_1(
825825
;
826826
#else // __SYCL_DEVICE_ONLY__
827827
{
828-
throw cl::sycl::feature_not_supported();
828+
__ESIMD_UNSUPPORTED_ON_HOST;
829829
return 0;
830830
}
831831
#endif // __SYCL_DEVICE_ONLY__
@@ -865,7 +865,7 @@ __esimd_lsc_xatomic_bti_2(
865865
;
866866
#else // __SYCL_DEVICE_ONLY__
867867
{
868-
throw cl::sycl::feature_not_supported();
868+
__ESIMD_UNSUPPORTED_ON_HOST;
869869
return 0;
870870
}
871871
#endif // __SYCL_DEVICE_ONLY__
@@ -897,7 +897,7 @@ __esimd_lsc_xatomic_stateless_0(__ESIMD_DNS::simd_mask_storage_t<N> pred,
897897
;
898898
#else // __SYCL_DEVICE_ONLY__
899899
{
900-
throw cl::sycl::feature_not_supported();
900+
__ESIMD_UNSUPPORTED_ON_HOST;
901901
return 0;
902902
}
903903
#endif // __SYCL_DEVICE_ONLY__
@@ -932,7 +932,7 @@ __esimd_lsc_xatomic_stateless_1(
932932
;
933933
#else // __SYCL_DEVICE_ONLY__
934934
{
935-
throw cl::sycl::feature_not_supported();
935+
__ESIMD_UNSUPPORTED_ON_HOST;
936936
return 0;
937937
}
938938
#endif // __SYCL_DEVICE_ONLY__
@@ -969,7 +969,7 @@ __esimd_lsc_xatomic_stateless_2(
969969
;
970970
#else // __SYCL_DEVICE_ONLY__
971971
{
972-
throw cl::sycl::feature_not_supported();
972+
__ESIMD_UNSUPPORTED_ON_HOST;
973973
return 0;
974974
}
975975
#endif // __SYCL_DEVICE_ONLY__
@@ -989,7 +989,7 @@ __ESIMD_INTRIN void __esimd_lsc_fence(__ESIMD_DNS::simd_mask_storage_t<N> pred)
989989
;
990990
#else // __SYCL_DEVICE_ONLY__
991991
{
992-
throw cl::sycl::feature_not_supported();
992+
__ESIMD_UNSUPPORTED_ON_HOST;
993993
}
994994
#endif // __SYCL_DEVICE_ONLY__
995995

0 commit comments

Comments
 (0)