14 #ifndef __SYCL_DEVICE_ONLY__
37 std::enable_if_t<IsValidCoordType<T>::value, T>
39 return Coords * Range[0];
43 std::enable_if_t<IsValidCoordType<T>::value,
vec<T, 2>>
45 return {Coords.x() * Range[0], Coords.y() * Range[1]};
49 std::enable_if_t<IsValidCoordType<T>::value,
vec<T, 4>>
51 return {Coords.x() * Range[0], Coords.y() * Range[1], Coords.z() * Range[2],
61 return {
static_cast<float>(Coords), 0.5f, 0.5f, 0.f};
65 std::enable_if_t<IsValidCoordType<T>::value, float4>
67 return {
static_cast<float>(Coords.x()),
static_cast<float>(Coords.y()), 0.5f,
72 std::enable_if_t<IsValidCoordType<T>::value, float4>
74 return {
static_cast<float>(Coords.x()),
static_cast<float>(Coords.y()),
75 static_cast<float>(Coords.z()), 0.f};
82 std::enable_if_t<std::is_integral_v<T>,
size_t>
84 return Coords * ElementSize;
88 std::enable_if_t<std::is_integral_v<T>,
size_t>
90 const uint8_t ElementSize) {
91 return Coords.x() * ElementSize + Coords.y() * ImgPitch[0];
95 std::enable_if_t<std::is_integral_v<T>,
size_t>
97 const uint8_t ElementSize) {
98 return Coords.x() * ElementSize + Coords.y() * ImgPitch[0] +
99 Coords.z() * ImgPitch[1];
125 template <
typename T>
130 switch (ChannelOrder) {
131 case image_channel_order::a:
134 case image_channel_order::r:
135 case image_channel_order::rx:
139 case image_channel_order::intensity:
145 case image_channel_order::luminance:
151 case image_channel_order::rg:
152 case image_channel_order::rgx:
157 case image_channel_order::ra:
161 case image_channel_order::rgb:
162 case image_channel_order::rgbx:
163 if (ChannelType == image_channel_type::unorm_short_565 ||
164 ChannelType == image_channel_type::unorm_short_555 ||
165 ChannelType == image_channel_type::unorm_int_101010) {
174 case image_channel_order::rgba:
175 case image_channel_order::ext_oneapi_srgba:
181 case image_channel_order::argb:
187 case image_channel_order::bgra:
193 case image_channel_order::abgr:
208 template <
typename T>
214 switch (ChannelOrder) {
215 case image_channel_order::a:
218 case image_channel_order::r:
219 case image_channel_order::rx:
220 case image_channel_order::intensity:
221 case image_channel_order::luminance:
224 case image_channel_order::rg:
225 case image_channel_order::rgx:
229 case image_channel_order::ra:
233 case image_channel_order::rgb:
234 case image_channel_order::rgbx:
235 if (ChannelType == image_channel_type::unorm_short_565 ||
236 ChannelType == image_channel_type::unorm_short_555 ||
237 ChannelType == image_channel_type::unorm_int_101010) {
245 case image_channel_order::rgba:
246 case image_channel_order::ext_oneapi_srgba:
252 case image_channel_order::argb:
258 case image_channel_order::bgra:
264 case image_channel_order::abgr:
280 template <
typename ChannelType>
285 switch (ImageChannelType) {
286 case image_channel_type::unsigned_int8:
287 case image_channel_type::unsigned_int16:
288 case image_channel_type::unsigned_int32:
289 RetData = PixelData.template convert<std::uint32_t>();
295 throw sycl::invalid_parameter_error(
296 "Datatype of read data - cl_uint4 is incompatible with the "
297 "image_channel_type of the image.",
298 PI_ERROR_INVALID_VALUE);
302 template <
typename ChannelType>
306 switch (ImageChannelType) {
307 case image_channel_type::signed_int8:
308 case image_channel_type::signed_int16:
309 case image_channel_type::signed_int32:
310 RetData = PixelData.template convert<std::int32_t>();
316 throw sycl::invalid_parameter_error(
317 "Datatype of read data - cl_int4 is incompatible with "
319 "image_channel_type of the image.",
320 PI_ERROR_INVALID_VALUE);
324 template <
typename ChannelType>
329 switch (ImageChannelType) {
330 case image_channel_type::snorm_int8:
332 RetData = (PixelData.template convert<float>()) / 127.0f;
333 RetData = sycl::fmax(RetData, -1);
335 case image_channel_type::snorm_int16:
337 RetData = (PixelData.template convert<float>()) / 32767.0f;
338 RetData = sycl::fmax(RetData, -1);
340 case image_channel_type::unorm_int8:
342 RetData = (PixelData.template convert<float>()) / 255.0f;
344 case image_channel_type::unorm_int16:
346 RetData = (PixelData.template convert<float>()) / 65535.0f;
348 case image_channel_type::unorm_short_565: {
354 ushort4 Temp(PixelData.x());
355 ushort4 MaskBits(0xF800 , 0x07E0 ,
357 ushort4 ShiftBits(11, 5, 0, 0);
358 float4 DivisorToNormalise(31.0f, 63.0f, 31.0f, 1);
359 Temp = (Temp & MaskBits) >> ShiftBits;
360 RetData = (Temp.template convert<float>()) / DivisorToNormalise;
363 case image_channel_type::unorm_short_555: {
370 ushort4 Temp(PixelData.x());
371 ushort4 MaskBits(0x7C00 , 0x03E0 ,
373 ushort4 ShiftBits(10, 5, 0, 0);
374 Temp = (Temp & MaskBits) >> ShiftBits;
375 RetData = (Temp.template convert<float>()) / 31.0f;
378 case image_channel_type::unorm_int_101010: {
381 uint4 Temp(PixelData.x());
382 uint4 MaskBits(0x3FF00000 , 0x000FFC00 ,
383 0x000003FF , 0x00000000);
384 uint4 ShiftBits(20, 10, 0, 0);
385 Temp = (Temp & MaskBits) >> ShiftBits;
386 RetData = (Temp.template convert<float>()) / 1023.0f;
389 case image_channel_type::signed_int8:
390 case image_channel_type::signed_int16:
391 case image_channel_type::signed_int32:
392 case image_channel_type::unsigned_int8:
393 case image_channel_type::unsigned_int16:
394 case image_channel_type::unsigned_int32:
398 throw sycl::invalid_parameter_error(
399 "Datatype of read data - cl_float4 is incompatible with the "
400 "image_channel_type of the image.",
401 PI_ERROR_INVALID_VALUE);
402 case image_channel_type::fp16:
405 RetData = PixelData.template convert<float>();
407 case image_channel_type::fp32:
408 RetData = PixelData.template convert<float>();
413 template <
typename ChannelType>
418 switch (ImageChannelType) {
419 case image_channel_type::snorm_int8:
421 RetDataFloat = (PixelData.template convert<float>()) / 127.0f;
422 RetDataFloat = sycl::fmax(RetDataFloat, -1);
424 case image_channel_type::snorm_int16:
426 RetDataFloat = (PixelData.template convert<float>()) / 32767.0f;
427 RetDataFloat = sycl::fmax(RetDataFloat, -1);
429 case image_channel_type::unorm_int8:
431 RetDataFloat = (PixelData.template convert<float>()) / 255.0f;
433 case image_channel_type::unorm_int16:
435 RetDataFloat = (PixelData.template convert<float>()) / 65535.0f;
437 case image_channel_type::unorm_short_565:
438 case image_channel_type::unorm_short_555:
439 case image_channel_type::unorm_int_101010:
441 throw sycl::feature_not_supported(
442 "Currently unsupported datatype conversion from image_channel_type "
444 PI_ERROR_INVALID_OPERATION);
445 case image_channel_type::signed_int8:
446 case image_channel_type::signed_int16:
447 case image_channel_type::signed_int32:
448 case image_channel_type::unsigned_int8:
449 case image_channel_type::unsigned_int16:
450 case image_channel_type::unsigned_int32:
454 throw sycl::invalid_parameter_error(
455 "Datatype to read- cl_half4 is incompatible with the "
456 "image_channel_type of the image.",
457 PI_ERROR_INVALID_VALUE);
458 case image_channel_type::fp16:
459 RetData = PixelData.template convert<half>();
461 case image_channel_type::fp32:
462 throw sycl::invalid_parameter_error(
463 "Datatype to read - cl_half4 is incompatible with the "
464 "image_channel_type of the image.",
465 PI_ERROR_INVALID_VALUE);
467 RetData = RetDataFloat.template convert<half>();
476 template <
typename ChannelType>
480 switch (ImageChannelType) {
481 case image_channel_type::unsigned_int8: {
483 std::uint32_t MinVal = min_v<std::uint8_t>();
484 std::uint32_t MaxVal = max_v<std::uint8_t>();
485 uint4 PixelData = sycl::clamp(WriteData, MinVal, MaxVal);
486 return PixelData.convert<ChannelType>();
488 case image_channel_type::unsigned_int16: {
490 std::uint32_t MinVal = min_v<std::uint16_t>();
491 std::uint32_t MaxVal = max_v<std::uint16_t>();
492 uint4 PixelData = sycl::clamp(WriteData, MinVal, MaxVal);
493 return PixelData.convert<ChannelType>();
495 case image_channel_type::unsigned_int32:
497 return WriteData.convert<ChannelType>();
502 throw sycl::invalid_parameter_error(
503 "Datatype of data to write - cl_uint4 is incompatible with the "
504 "image_channel_type of the image.",
505 PI_ERROR_INVALID_VALUE);
509 template <
typename ChannelType>
514 switch (ImageChannelType) {
515 case image_channel_type::signed_int8: {
517 std::int32_t MinVal = min_v<std::int8_t>();
518 std::int32_t MaxVal = max_v<std::int8_t>();
519 int4 PixelData = sycl::clamp(WriteData, MinVal, MaxVal);
520 return PixelData.convert<ChannelType>();
522 case image_channel_type::signed_int16: {
524 std::int32_t MinVal = min_v<std::int16_t>();
525 std::int32_t MaxVal = max_v<std::int16_t>();
526 int4 PixelData = sycl::clamp(WriteData, MinVal, MaxVal);
527 return PixelData.convert<ChannelType>();
529 case image_channel_type::signed_int32:
530 return WriteData.convert<ChannelType>();
535 throw sycl::invalid_parameter_error(
536 "Datatype of data to write - cl_int4 is incompatible with the "
537 "image_channel_type of the image.",
538 PI_ERROR_INVALID_VALUE);
542 template <
typename ChannelType>
544 float4 Temp = WriteData * MulFactor;
545 int4 TempInInt = Temp.convert<int, rounding_mode::rte>();
546 int4 TempInIntSaturated =
547 sycl::clamp(TempInInt, min_v<ChannelType>(), max_v<ChannelType>());
548 return TempInIntSaturated.convert<ChannelType>();
551 template <
typename ChannelType>
558 switch (ImageChannelType) {
559 case image_channel_type::snorm_int8:
561 return processFloatDataToPixel<ChannelType>(WriteData, 127.0f);
562 case image_channel_type::snorm_int16:
564 return processFloatDataToPixel<ChannelType>(WriteData, 32767.0f);
565 case image_channel_type::unorm_int8:
567 return processFloatDataToPixel<ChannelType>(WriteData, 255.0f);
568 case image_channel_type::unorm_int16:
570 return processFloatDataToPixel<ChannelType>(WriteData, 65535.0f);
571 case image_channel_type::unorm_short_565:
573 throw sycl::feature_not_supported(
574 "Currently unsupported datatype conversion from image_channel_type "
576 PI_ERROR_INVALID_OPERATION);
577 case image_channel_type::unorm_short_555:
584 processFloatDataToPixel<std::uint16_t>(WriteData, 32.0f);
585 PixelData = sycl::min(PixelData,
static_cast<ChannelType
>(0x1f));
592 (PixelData.x() << 10) | (PixelData.y() << 5) | PixelData.z();
593 return PixelData.convert<ChannelType>();
595 case image_channel_type::unorm_int_101010:
601 processFloatDataToPixel<std::uint32_t>(WriteData, 1023.0f);
602 PixelData = sycl::min(PixelData,
static_cast<ChannelType
>(0x3ff));
604 (PixelData.x() << 20) | (PixelData.y() << 10) | PixelData.z();
605 return PixelData.convert<ChannelType>();
607 case image_channel_type::signed_int8:
608 case image_channel_type::signed_int16:
609 case image_channel_type::signed_int32:
610 case image_channel_type::unsigned_int8:
611 case image_channel_type::unsigned_int16:
612 case image_channel_type::unsigned_int32:
616 throw sycl::invalid_parameter_error(
617 "Datatype of data to write - cl_float4 is incompatible with the "
618 "image_channel_type of the image.",
619 PI_ERROR_INVALID_VALUE);
620 case image_channel_type::fp16:
623 return WriteData.convert<ChannelType>();
624 case image_channel_type::fp32:
625 return WriteData.convert<ChannelType>();
629 template <
typename ChannelType>
633 float4 WriteDataFloat = WriteData.convert<
float>();
634 switch (ImageChannelType) {
635 case image_channel_type::snorm_int8:
637 return processFloatDataToPixel<ChannelType>(WriteDataFloat, 127.0f);
638 case image_channel_type::snorm_int16:
640 return processFloatDataToPixel<ChannelType>(WriteDataFloat, 32767.0f);
641 case image_channel_type::unorm_int8:
643 return processFloatDataToPixel<ChannelType>(WriteDataFloat, 255.0f);
644 case image_channel_type::unorm_int16:
646 return processFloatDataToPixel<ChannelType>(WriteDataFloat, 65535.0f);
647 case image_channel_type::unorm_short_565:
648 case image_channel_type::unorm_short_555:
649 case image_channel_type::unorm_int_101010:
651 throw sycl::feature_not_supported(
652 "Currently unsupported datatype conversion from image_channel_type "
654 PI_ERROR_INVALID_OPERATION);
655 case image_channel_type::signed_int8:
656 case image_channel_type::signed_int16:
657 case image_channel_type::signed_int32:
658 case image_channel_type::unsigned_int8:
659 case image_channel_type::unsigned_int16:
660 case image_channel_type::unsigned_int32:
664 throw sycl::invalid_parameter_error(
665 "Datatype of data to write - cl_float4 is incompatible with the "
666 "image_channel_type of the image.",
667 PI_ERROR_INVALID_VALUE);
668 case image_channel_type::fp16:
669 return WriteData.convert<ChannelType>();
670 case image_channel_type::fp32:
671 throw sycl::invalid_parameter_error(
672 "Datatype of data to write - cl_float4 is incompatible with the "
673 "image_channel_type of the image.",
674 PI_ERROR_INVALID_VALUE);
692 template <
typename CoordT,
typename WriteDataT>
694 id<3> ImgPitch, uint8_t ElementSize,
698 auto Ptr =
static_cast<unsigned char *
>(BasePtr) +
701 switch (ImgChannelType) {
702 case image_channel_type::snorm_int8:
703 writePixel(convertWriteData<std::int8_t>(Color, ImgChannelType),
704 reinterpret_cast<std::int8_t *
>(Ptr), ImgChannelOrder,
707 case image_channel_type::snorm_int16:
708 writePixel(convertWriteData<std::int16_t>(Color, ImgChannelType),
709 reinterpret_cast<std::int16_t *
>(Ptr), ImgChannelOrder,
712 case image_channel_type::unorm_int8:
713 writePixel(convertWriteData<std::uint8_t>(Color, ImgChannelType),
714 reinterpret_cast<std::uint8_t *
>(Ptr), ImgChannelOrder,
717 case image_channel_type::unorm_int16:
718 writePixel(convertWriteData<std::uint16_t>(Color, ImgChannelType),
719 reinterpret_cast<std::uint16_t *
>(Ptr), ImgChannelOrder,
722 case image_channel_type::unorm_short_565:
723 writePixel(convertWriteData<short>(Color, ImgChannelType),
724 reinterpret_cast<short *
>(Ptr), ImgChannelOrder, ImgChannelType);
726 case image_channel_type::unorm_short_555:
727 writePixel(convertWriteData<short>(Color, ImgChannelType),
728 reinterpret_cast<short *
>(Ptr), ImgChannelOrder, ImgChannelType);
730 case image_channel_type::unorm_int_101010:
731 writePixel(convertWriteData<std::uint32_t>(Color, ImgChannelType),
732 reinterpret_cast<std::uint32_t *
>(Ptr), ImgChannelOrder,
735 case image_channel_type::signed_int8:
736 writePixel(convertWriteData<std::int8_t>(Color, ImgChannelType),
737 reinterpret_cast<std::int8_t *
>(Ptr), ImgChannelOrder,
740 case image_channel_type::signed_int16:
741 writePixel(convertWriteData<std::int16_t>(Color, ImgChannelType),
742 reinterpret_cast<std::int16_t *
>(Ptr), ImgChannelOrder,
745 case image_channel_type::signed_int32:
746 writePixel(convertWriteData<std::int32_t>(Color, ImgChannelType),
747 reinterpret_cast<std::int32_t *
>(Ptr), ImgChannelOrder,
750 case image_channel_type::unsigned_int8:
751 writePixel(convertWriteData<std::uint8_t>(Color, ImgChannelType),
752 reinterpret_cast<std::uint8_t *
>(Ptr), ImgChannelOrder,
755 case image_channel_type::unsigned_int16:
756 writePixel(convertWriteData<std::uint16_t>(Color, ImgChannelType),
757 reinterpret_cast<std::uint16_t *
>(Ptr), ImgChannelOrder,
760 case image_channel_type::unsigned_int32:
761 writePixel(convertWriteData<std::uint32_t>(Color, ImgChannelType),
762 reinterpret_cast<std::uint32_t *
>(Ptr), ImgChannelOrder,
765 case image_channel_type::fp16:
769 convertWriteData<half>(Color, ImgChannelType),
770 reinterpret_cast<half *
>(Ptr), ImgChannelOrder, ImgChannelType);
772 case image_channel_type::fp32:
773 writePixel(convertWriteData<float>(Color, ImgChannelType),
774 reinterpret_cast<float *
>(Ptr), ImgChannelOrder, ImgChannelType);
793 template <
typename DataT>
797 const uint8_t ElementSize) {
799 auto Ptr =
static_cast<unsigned char *
>(BasePtr) +
804 switch (ImageChannelType) {
807 case image_channel_type::snorm_int8:
808 convertReadData<std::int8_t>(
readPixel(
reinterpret_cast<std::int8_t *
>(Ptr),
809 ImageChannelOrder, ImageChannelType),
810 image_channel_type::snorm_int8, Color);
812 case image_channel_type::snorm_int16:
813 convertReadData<std::int16_t>(
814 readPixel(
reinterpret_cast<std::int16_t *
>(Ptr), ImageChannelOrder,
816 image_channel_type::snorm_int16, Color);
818 case image_channel_type::unorm_int8:
819 convertReadData<std::uint8_t>(
820 readPixel(
reinterpret_cast<std::uint8_t *
>(Ptr), ImageChannelOrder,
822 image_channel_type::unorm_int8, Color);
824 case image_channel_type::unorm_int16:
825 convertReadData<std::uint16_t>(
826 readPixel(
reinterpret_cast<std::uint16_t *
>(Ptr), ImageChannelOrder,
828 image_channel_type::unorm_int16, Color);
830 case image_channel_type::unorm_short_565:
831 convertReadData<std::uint16_t>(
832 readPixel(
reinterpret_cast<std::uint16_t *
>(Ptr), ImageChannelOrder,
834 image_channel_type::unorm_short_565, Color);
836 case image_channel_type::unorm_short_555:
837 convertReadData<std::uint16_t>(
838 readPixel(
reinterpret_cast<std::uint16_t *
>(Ptr), ImageChannelOrder,
840 image_channel_type::unorm_short_555, Color);
842 case image_channel_type::unorm_int_101010:
843 convertReadData<std::uint32_t>(
844 readPixel(
reinterpret_cast<std::uint32_t *
>(Ptr), ImageChannelOrder,
846 image_channel_type::unorm_int_101010, Color);
848 case image_channel_type::signed_int8:
849 convertReadData<std::int8_t>(
readPixel(
reinterpret_cast<std::int8_t *
>(Ptr),
850 ImageChannelOrder, ImageChannelType),
851 image_channel_type::signed_int8, Color);
853 case image_channel_type::signed_int16:
854 convertReadData<std::int16_t>(
855 readPixel(
reinterpret_cast<std::int16_t *
>(Ptr), ImageChannelOrder,
857 image_channel_type::signed_int16, Color);
859 case image_channel_type::signed_int32:
860 convertReadData<std::int32_t>(
861 readPixel(
reinterpret_cast<std::int32_t *
>(Ptr), ImageChannelOrder,
863 image_channel_type::signed_int32, Color);
865 case image_channel_type::unsigned_int8:
866 convertReadData<std::uint8_t>(
867 readPixel(
reinterpret_cast<std::uint8_t *
>(Ptr), ImageChannelOrder,
869 image_channel_type::unsigned_int8, Color);
871 case image_channel_type::unsigned_int16:
872 convertReadData<std::uint16_t>(
873 readPixel(
reinterpret_cast<std::uint16_t *
>(Ptr), ImageChannelOrder,
875 image_channel_type::unsigned_int16, Color);
877 case image_channel_type::unsigned_int32:
878 convertReadData<std::uint32_t>(
879 readPixel(
reinterpret_cast<std::uint32_t *
>(Ptr), ImageChannelOrder,
881 image_channel_type::unsigned_int32, Color);
883 case image_channel_type::fp16:
884 convertReadData<half>(
readPixel(
reinterpret_cast<half *
>(Ptr),
885 ImageChannelOrder, ImageChannelType),
886 image_channel_type::fp16, Color);
888 case image_channel_type::fp32:
889 convertReadData<float>(
readPixel(
reinterpret_cast<float *
>(Ptr),
890 ImageChannelOrder, ImageChannelType),
891 image_channel_type::fp32, Color);
900 template <
typename DataT>
905 const uint8_t ElementSize) {
911 RetData = ReadPixelData<DataT>(PixelCoord, ImgPitch, ImgChannelType,
912 ImgChannelOrder, BasePtr, ElementSize);
922 template <
typename DataT>
928 void *BasePtr,
const uint8_t ElementSize) {
929 std::int32_t i0 = CoordValues.s0(), j0 = CoordValues.s1(),
930 k0 = CoordValues.s2(), i1 = CoordValues.s4(),
931 j1 = CoordValues.s5(), k1 = CoordValues.s6();
933 auto getColorInFloat = [&](int4 V) {
935 getColor<DataT>(V, SmplAddrMode, ImgRange, ImgPitch, ImgChannelType,
936 ImgChannelOrder, BasePtr, ElementSize);
937 return Res.template convert<float>();
941 float4 Ci0j0k0 = getColorInFloat(int4{i0, j0, k0, 0});
943 float4 Ci1j0k0 = getColorInFloat(int4{i1, j0, k0, 0});
945 float4 Ci0j1k0 = getColorInFloat(int4{i0, j1, k0, 0});
947 float4 Ci1j1k0 = getColorInFloat(int4{i1, j1, k0, 0});
949 float4 Ci0j0k1 = getColorInFloat(int4{i0, j0, k1, 0});
951 float4 Ci1j0k1 = getColorInFloat(int4{i1, j0, k1, 0});
953 float4 Ci0j1k1 = getColorInFloat(int4{i0, j1, k1, 0});
955 float4 Ci1j1k1 = getColorInFloat(int4{i1, j1, k1, 0});
961 Ci0j0k0 = (1 - a) * (1 - b) * (1 - c) * Ci0j0k0;
962 Ci1j0k0 = a * (1 - b) * (1 - c) * Ci1j0k0;
963 Ci0j1k0 = (1 - a) * b * (1 - c) * Ci0j1k0;
964 Ci1j1k0 = a * b * (1 - c) * Ci1j1k0;
965 Ci0j0k1 = (1 - a) * (1 - b) * c * Ci0j0k1;
966 Ci1j0k1 = a * (1 - b) * c * Ci1j0k1;
967 Ci0j1k1 = (1 - a) * b * c * Ci0j1k1;
968 Ci1j1k1 = a * b * c * Ci1j1k1;
970 float4 RetData = Ci0j0k0 + Ci1j0k0 + Ci0j1k0 + Ci1j1k0 + Ci0j0k1 + Ci1j0k1 +
1021 template <
typename CoordT,
typename DataT>
1027 void *BasePtr, uint8_t ElementSize) {
1030 float4 FloatCoorduvw;
1036 switch (SmplNormMode) {
1037 case coordinate_normalization_mode::unnormalized:
1038 switch (SmplAddrMode) {
1039 case addressing_mode::mirrored_repeat:
1040 case addressing_mode::repeat:
1041 throw sycl::feature_not_supported(
1042 "Sampler used with unsupported configuration of "
1043 "mirrored_repeat/repeat filtering mode with unnormalized "
1045 PI_ERROR_INVALID_OPERATION);
1046 case addressing_mode::clamp_to_edge:
1048 case addressing_mode::none:
1054 case coordinate_normalization_mode::normalized:
1055 switch (SmplAddrMode) {
1056 case addressing_mode::mirrored_repeat:
1057 case addressing_mode::repeat:
1063 case addressing_mode::clamp_to_edge:
1065 case addressing_mode::none:
1084 switch (SmplFiltMode) {
1085 case filtering_mode::nearest: {
1094 getColor<DataT>(PixelCoord, SmplAddrMode, ImgRange, ImgPitch,
1095 ImgChannelType, ImgChannelOrder, BasePtr, ElementSize);
1098 case filtering_mode::linear: {
1108 RetData = ReadPixelDataLinearFiltMode<DataT>(
1109 CoordValues, Retabc, SmplAddrMode, ImgRange, ImgPitch, ImgChannelType,
1110 ImgChannelOrder, BasePtr, ElementSize);
1120 template <
typename CoordT,
typename DataT>
1126 void *BasePtr, uint8_t ElementSize) {
1129 Smpl.get_coordinate_normalization_mode();
1133 return imageReadSamplerHostImpl<CoordT, DataT>(
1134 Coords, SmplNormMode, SmplAddrMode, SmplFiltMode, ImgRange, ImgPitch,
1135 ImgChannelType, ImgChannelOrder, BasePtr, ElementSize);
1139 template <
typename CoordT,
typename DataT>
1145 void *BasePtr, uint8_t ElementSize) {
1150 return imageReadSamplerHostImpl<CoordT, DataT>(
1151 Coords, SmplNormMode, SmplAddrMode, SmplFiltMode, ImgRange, ImgPitch,
1152 ImgChannelType, ImgChannelOrder, BasePtr, ElementSize);