diff --git a/sycl/include/CL/sycl/accessor.hpp b/sycl/include/CL/sycl/accessor.hpp
index cf13264b4fca0..328495030420b 100644
--- a/sycl/include/CL/sycl/accessor.hpp
+++ b/sycl/include/CL/sycl/accessor.hpp
@@ -1023,7 +1023,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), AdjustedDim, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1054,7 +1054,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), AdjustedDim, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1084,7 +1084,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1115,7 +1115,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1145,7 +1145,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1178,7 +1178,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1237,7 +1237,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1269,7 +1269,7 @@ class __SYCL_SPECIAL_CLASS accessor :
             getAdjustedMode(PropertyList),
             detail::getSyclObjImpl(BufferRef).get(), Dimensions, sizeof(DataT),
             BufferRef.OffsetInBytes, BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1443,7 +1443,14 @@ class __SYCL_SPECIAL_CLASS accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
+    if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
+                                BufferRef.get_range()))
+      throw sycl::invalid_object_error(
+          "accessor with requested offset and range would exceed the bounds of "
+          "the buffer",
+          PI_INVALID_VALUE);
+
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1477,7 +1484,14 @@ class __SYCL_SPECIAL_CLASS accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
+    if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
+                                BufferRef.get_range()))
+      throw sycl::invalid_object_error(
+          "accessor with requested offset and range would exceed the bounds of "
+          "the buffer",
+          PI_INVALID_VALUE);
+
     if (!IsPlaceH)
       addHostAccessorAndWait(AccessorBaseHost::impl.get());
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
@@ -1538,7 +1552,14 @@ class __SYCL_SPECIAL_CLASS accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
+    if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
+                                BufferRef.get_range()))
+      throw sycl::invalid_object_error(
+          "accessor with requested offset and range would exceed the bounds of "
+          "the buffer",
+          PI_INVALID_VALUE);
+
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1571,7 +1592,14 @@ class __SYCL_SPECIAL_CLASS accessor :
                          detail::getSyclObjImpl(BufferRef).get(), Dimensions,
                          sizeof(DataT), BufferRef.OffsetInBytes,
                          BufferRef.IsSubBuffer) {
-    checkDeviceAccessorBufferSize(BufferRef.size());
+    preScreenAccessor(BufferRef.size(), PropertyList);
+    if (BufferRef.isOutOfBounds(AccessOffset, AccessRange,
+                                BufferRef.get_range()))
+      throw sycl::invalid_object_error(
+          "accessor with requested offset and range would exceed the bounds of "
+          "the buffer",
+          PI_INVALID_VALUE);
+
     detail::associateWithHandler(CommandGroupHandler, this, AccessTarget);
     detail::constructorNotification(detail::getSyclObjImpl(BufferRef).get(),
                                     detail::AccessorBaseHost::impl.get(),
@@ -1761,12 +1789,22 @@ class __SYCL_SPECIAL_CLASS accessor :
     return getQualifiedPtr();
   }
 
-  void checkDeviceAccessorBufferSize(const size_t elemInBuffer) {
+  void preScreenAccessor(const size_t elemInBuffer,
+                         const PropertyListT &PropertyList) {
+    // check device accessor buffer size
     if (!IsHostBuf && elemInBuffer == 0)
-      throw cl::sycl::invalid_object_error(
+      throw sycl::invalid_object_error(
           "SYCL buffer size is zero. To create a device accessor, SYCL "
           "buffer size must be greater than zero.",
           PI_INVALID_VALUE);
+
+    // check that no_init property is compatible with access mode
+    if (PropertyList.template has_property<property::no_init>() &&
+        AccessMode == access::mode::read) {
+      throw sycl::invalid_object_error(
+          "accessor would cannot be both read_only and no_init",
+          PI_INVALID_VALUE);
+    }
   }
 };
 
diff --git a/sycl/include/CL/sycl/backend.hpp b/sycl/include/CL/sycl/backend.hpp
index 2954217f1ca64..1065e55d497f3 100644
--- a/sycl/include/CL/sycl/backend.hpp
+++ b/sycl/include/CL/sycl/backend.hpp
@@ -69,7 +69,8 @@ auto get_native(const SyclObjectT &Obj)
     -> backend_return_t<BackendName, SyclObjectT> {
   // TODO use SYCL 2020 exception when implemented
   if (Obj.get_backend() != BackendName) {
-    throw runtime_error("Backends mismatch", PI_INVALID_OPERATION);
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_INVALID_OPERATION);
   }
   return Obj.template get_native<BackendName>();
 }
@@ -82,7 +83,8 @@ inline backend_return_t<backend::opencl, event>
 get_native<backend::opencl, event>(const event &Obj) {
   // TODO use SYCL 2020 exception when implemented
   if (Obj.get_backend() != backend::opencl) {
-    throw runtime_error("Backends mismatch", PI_INVALID_OPERATION);
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_INVALID_OPERATION);
   }
   backend_return_t<backend::opencl, event> ReturnValue;
   for (auto const &element : Obj.getNativeVector()) {
@@ -104,7 +106,8 @@ inline backend_return_t<backend::opencl, event> get_native<
     backend::opencl, event>(const event &Obj) {
   // TODO use SYCL 2020 exception when implemented
   if (Obj.get_backend() != backend::opencl) {
-    throw runtime_error("Backends mismatch", PI_INVALID_OPERATION);
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_INVALID_OPERATION);
   }
   return reinterpret_cast<
       typename detail::interop<backend::opencl, event>::type>(Obj.getNative());
diff --git a/sycl/include/CL/sycl/buffer.hpp b/sycl/include/CL/sycl/buffer.hpp
index 52a2cdf8b75c7..54c5312a46482 100644
--- a/sycl/include/CL/sycl/buffer.hpp
+++ b/sycl/include/CL/sycl/buffer.hpp
@@ -403,6 +403,11 @@ class buffer {
       handler &commandGroupHandler, range<dimensions> accessRange,
       id<dimensions> accessOffset = {},
       const detail::code_location CodeLoc = detail::code_location::current()) {
+    if (isOutOfBounds(accessOffset, accessRange, this->Range))
+      throw cl::sycl::invalid_object_error(
+          "Requested accessor would exceed the bounds of the buffer",
+          PI_INVALID_VALUE);
+
     return accessor<T, dimensions, mode, target, access::placeholder::false_t,
                     ext::oneapi::accessor_property_list<>>(
         *this, commandGroupHandler, accessRange, accessOffset, {}, CodeLoc);
@@ -414,6 +419,11 @@ class buffer {
   get_access(
       range<dimensions> accessRange, id<dimensions> accessOffset = {},
       const detail::code_location CodeLoc = detail::code_location::current()) {
+    if (isOutOfBounds(accessOffset, accessRange, this->Range))
+      throw cl::sycl::invalid_object_error(
+          "Requested accessor would exceed the bounds of the buffer",
+          PI_INVALID_VALUE);
+
     return accessor<T, dimensions, mode, access::target::host_buffer,
                     access::placeholder::false_t,
                     ext::oneapi::accessor_property_list<>>(
@@ -499,6 +509,17 @@ class buffer {
     return impl->template get_property<propertyT>();
   }
 
+protected:
+  bool isOutOfBounds(const id<dimensions> &offset,
+                     const range<dimensions> &newRange,
+                     const range<dimensions> &parentRange) {
+    bool outOfBounds = false;
+    for (int i = 0; i < dimensions; ++i)
+      outOfBounds |= newRange[i] + offset[i] > parentRange[i];
+
+    return outOfBounds;
+  }
+
 private:
   std::shared_ptr<detail::buffer_impl> impl;
   template <class Obj>
@@ -554,16 +575,6 @@ class buffer {
     return detail::getLinearIndex(offset, range) * sizeof(Type);
   }
 
-  bool isOutOfBounds(const id<dimensions> &offset,
-                     const range<dimensions> &newRange,
-                     const range<dimensions> &parentRange) {
-    bool outOfBounds = false;
-    for (int i = 0; i < dimensions; ++i)
-      outOfBounds |= newRange[i] + offset[i] > parentRange[i];
-
-    return outOfBounds;
-  }
-
   bool isContiguousRegion(const id<1> &, const range<1> &, const range<1> &) {
     // 1D sub buffer always has contiguous region
     return true;
diff --git a/sycl/include/CL/sycl/exception.hpp b/sycl/include/CL/sycl/exception.hpp
index 276c5facdf3e4..58e2de12577a8 100644
--- a/sycl/include/CL/sycl/exception.hpp
+++ b/sycl/include/CL/sycl/exception.hpp
@@ -134,7 +134,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::runtime instead.") runtime_error
     : public exception {
 public:
-  runtime_error() = default;
+  runtime_error() : exception(make_error_code(errc::runtime)) {}
 
   runtime_error(const char *Msg, cl_int Err)
       : runtime_error(std::string(Msg), Err) {}
@@ -142,16 +142,18 @@ class __SYCL2020_DEPRECATED(
   runtime_error(const std::string &Msg, cl_int Err)
       : exception(make_error_code(errc::runtime), Msg, Err) {}
 
-protected:
   runtime_error(std::error_code ec, const std::string &Msg, const cl_int CLErr)
       : exception(ec, Msg, CLErr) {}
+
+protected:
+  runtime_error(std::error_code ec) : exception(ec) {}
 };
 
 class __SYCL2020_DEPRECATED("use sycl::exception with sycl::errc::kernel or "
                             "errc::kernel_argument instead.") kernel_error
     : public runtime_error {
 public:
-  kernel_error() = default;
+  kernel_error() : runtime_error(make_error_code(errc::kernel)) {}
 
   kernel_error(const char *Msg, cl_int Err)
       : kernel_error(std::string(Msg), Err) {}
@@ -164,7 +166,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::accessor instead.") accessor_error
     : public runtime_error {
 public:
-  accessor_error() = default;
+  accessor_error() : runtime_error(make_error_code(errc::accessor)) {}
 
   accessor_error(const char *Msg, cl_int Err)
       : accessor_error(std::string(Msg), Err) {}
@@ -177,7 +179,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::nd_range instead.") nd_range_error
     : public runtime_error {
 public:
-  nd_range_error() = default;
+  nd_range_error() : runtime_error(make_error_code(errc::nd_range)) {}
 
   nd_range_error(const char *Msg, cl_int Err)
       : nd_range_error(std::string(Msg), Err) {}
@@ -190,7 +192,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::event instead.") event_error
     : public runtime_error {
 public:
-  event_error() = default;
+  event_error() : runtime_error(make_error_code(errc::event)) {}
 
   event_error(const char *Msg, cl_int Err)
       : event_error(std::string(Msg), Err) {}
@@ -203,7 +205,8 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with a sycl::errc enum value instead.")
     invalid_parameter_error : public runtime_error {
 public:
-  invalid_parameter_error() = default;
+  invalid_parameter_error()
+      : runtime_error(make_error_code(errc::kernel_argument)) {}
 
   invalid_parameter_error(const char *Msg, cl_int Err)
       : invalid_parameter_error(std::string(Msg), Err) {}
@@ -216,7 +219,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with a sycl::errc enum value instead.") device_error
     : public exception {
 public:
-  device_error() = default;
+  device_error() : exception(make_error_code(errc::invalid)) {}
 
   device_error(const char *Msg, cl_int Err)
       : device_error(std::string(Msg), Err) {}
@@ -225,6 +228,8 @@ class __SYCL2020_DEPRECATED(
       : exception(make_error_code(errc::invalid), Msg, Err) {}
 
 protected:
+  device_error(std::error_code ec) : exception(ec) {}
+
   device_error(std::error_code ec, const std::string &Msg, const cl_int CLErr)
       : exception(ec, Msg, CLErr) {}
 };
@@ -233,7 +238,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with a sycl::errc enum value instead.")
     compile_program_error : public device_error {
 public:
-  compile_program_error() = default;
+  compile_program_error() : device_error(make_error_code(errc::build)) {}
 
   compile_program_error(const char *Msg, cl_int Err)
       : compile_program_error(std::string(Msg), Err) {}
@@ -246,7 +251,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with a sycl::errc enum value instead.")
     link_program_error : public device_error {
 public:
-  link_program_error() = default;
+  link_program_error() : device_error(make_error_code(errc::build)) {}
 
   link_program_error(const char *Msg, cl_int Err)
       : link_program_error(std::string(Msg), Err) {}
@@ -259,7 +264,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with a sycl::errc enum value instead.")
     invalid_object_error : public device_error {
 public:
-  invalid_object_error() = default;
+  invalid_object_error() : device_error(make_error_code(errc::invalid)) {}
 
   invalid_object_error(const char *Msg, cl_int Err)
       : invalid_object_error(std::string(Msg), Err) {}
@@ -272,7 +277,8 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::memory_allocation instead.")
     memory_allocation_error : public device_error {
 public:
-  memory_allocation_error() = default;
+  memory_allocation_error()
+      : device_error(make_error_code(errc::memory_allocation)) {}
 
   memory_allocation_error(const char *Msg, cl_int Err)
       : memory_allocation_error(std::string(Msg), Err) {}
@@ -285,7 +291,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::platform instead.") platform_error
     : public device_error {
 public:
-  platform_error() = default;
+  platform_error() : device_error(make_error_code(errc::platform)) {}
 
   platform_error(const char *Msg, cl_int Err)
       : platform_error(std::string(Msg), Err) {}
@@ -298,7 +304,7 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::profiling instead.") profiling_error
     : public device_error {
 public:
-  profiling_error() = default;
+  profiling_error() : device_error(make_error_code(errc::profiling)) {}
 
   profiling_error(const char *Msg, cl_int Err)
       : profiling_error(std::string(Msg), Err) {}
@@ -311,7 +317,8 @@ class __SYCL2020_DEPRECATED(
     "use sycl::exception with sycl::errc::feature_not_supported instead.")
     feature_not_supported : public device_error {
 public:
-  feature_not_supported() = default;
+  feature_not_supported()
+      : device_error(make_error_code(errc::feature_not_supported)) {}
 
   feature_not_supported(const char *Msg, cl_int Err)
       : feature_not_supported(std::string(Msg), Err) {}
diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp
index 341a46ca9679d..2bcd75b54b3cb 100644
--- a/sycl/include/CL/sycl/handler.hpp
+++ b/sycl/include/CL/sycl/handler.hpp
@@ -2414,8 +2414,11 @@ class __SYCL_EXPORT handler {
                   "Invalid source accessor mode for the copy method.");
     static_assert(isValidModeForDestinationAccessor(AccessMode_Dst),
                   "Invalid destination accessor mode for the copy method.");
-    assert(Dst.get_size() >= Src.get_size() &&
-           "The destination accessor does not fit the copied memory.");
+    if (Dst.get_size() < Src.get_size())
+      throw sycl::invalid_object_error(
+          "The destination accessor size is too small to copy the memory into.",
+          CL_INVALID_OPERATION);
+
     if (copyAccToAccHelper(Src, Dst))
       return;
     setType(detail::CG::CopyAccToAcc);
diff --git a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
index 19e2f4d38bdc6..8fab908380242 100644
--- a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
+++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp
@@ -193,7 +193,8 @@ inline auto get_native<backend::ext_oneapi_level_zero>(
                         kernel_bundle<bundle_state::executable>> {
   // TODO use SYCL 2020 exception when implemented
   if (Obj.get_backend() != backend::ext_oneapi_level_zero)
-    throw runtime_error("Backends mismatch", PI_INVALID_OPERATION);
+    throw runtime_error(errc::backend_mismatch, "Backends mismatch",
+                        PI_INVALID_OPERATION);
 
   return Obj.template getNative<backend::ext_oneapi_level_zero>();
 }
diff --git a/sycl/include/sycl/ext/oneapi/reduction.hpp b/sycl/include/sycl/ext/oneapi/reduction.hpp
index 13e1c312fcc89..2944c8832275b 100644
--- a/sycl/include/sycl/ext/oneapi/reduction.hpp
+++ b/sycl/include/sycl/ext/oneapi/reduction.hpp
@@ -426,7 +426,8 @@ class reduction_impl : private reduction_impl_base {
         MIdentity(getIdentity()), InitializeToIdentity(InitializeToIdentity) {
     associateWithHandler(CGH);
     if (Buffer.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
@@ -438,7 +439,8 @@ class reduction_impl : private reduction_impl_base {
       : MRWAcc(new rw_accessor_type(Acc)), MIdentity(getIdentity()),
         InitializeToIdentity(false) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
@@ -450,7 +452,8 @@ class reduction_impl : private reduction_impl_base {
       : MDWAcc(new dw_accessor_type(Acc)), MIdentity(getIdentity()),
         InitializeToIdentity(true) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
@@ -467,7 +470,8 @@ class reduction_impl : private reduction_impl_base {
         MIdentity(getIdentity()), InitializeToIdentity(InitializeToIdentity) {
     associateWithHandler(CGH);
     if (Buffer.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
     // For now the implementation ignores the identity value given by user
     // when the implementation knows the identity.
@@ -491,7 +495,8 @@ class reduction_impl : private reduction_impl_base {
       : MRWAcc(new rw_accessor_type(Acc)), MIdentity(getIdentity()),
         InitializeToIdentity(false) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
     // For now the implementation ignores the identity value given by user
     // when the implementation knows the identity.
@@ -515,7 +520,8 @@ class reduction_impl : private reduction_impl_base {
       : MDWAcc(new dw_accessor_type(Acc)), MIdentity(getIdentity()),
         InitializeToIdentity(true) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
     // For now the implementation ignores the identity value given by user
     // when the implementation knows the identity.
@@ -542,7 +548,8 @@ class reduction_impl : private reduction_impl_base {
         MBinaryOp(BOp), InitializeToIdentity(InitializeToIdentity) {
     associateWithHandler(CGH);
     if (Buffer.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
@@ -554,7 +561,8 @@ class reduction_impl : private reduction_impl_base {
       : MRWAcc(new rw_accessor_type(Acc)), MIdentity(Identity), MBinaryOp(BOp),
         InitializeToIdentity(false) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
@@ -566,7 +574,8 @@ class reduction_impl : private reduction_impl_base {
       : MDWAcc(new dw_accessor_type(Acc)), MIdentity(Identity), MBinaryOp(BOp),
         InitializeToIdentity(true) {
     if (Acc.size() != 1)
-      throw sycl::runtime_error("Reduction variable must be a scalar.",
+      throw sycl::runtime_error(errc::invalid,
+                                "Reduction variable must be a scalar.",
                                 PI_INVALID_VALUE);
   }
 
diff --git a/sycl/source/backend.cpp b/sycl/source/backend.cpp
index 691bdbf0ab91c..59c341c5c5c3c 100644
--- a/sycl/source/backend.cpp
+++ b/sycl/source/backend.cpp
@@ -145,27 +145,29 @@ make_kernel_bundle(pi_native_handle NativeHandle, const context &TargetContext,
     switch (BinaryType) {
     case (PI_PROGRAM_BINARY_TYPE_NONE):
       if (State == bundle_state::object)
-        Plugin.call<PiApiKind::piProgramCompile>(
+        Plugin.call<errc::build, PiApiKind::piProgramCompile>(
             PiProgram, 1, &Dev, nullptr, 0, nullptr, nullptr, nullptr, nullptr);
       else if (State == bundle_state::executable)
-        Plugin.call<PiApiKind::piProgramBuild>(PiProgram, 1, &Dev, nullptr,
-                                               nullptr, nullptr);
+        Plugin.call<errc::build, PiApiKind::piProgramBuild>(
+            PiProgram, 1, &Dev, nullptr, nullptr, nullptr);
       break;
     case (PI_PROGRAM_BINARY_TYPE_COMPILED_OBJECT):
     case (PI_PROGRAM_BINARY_TYPE_LIBRARY):
       if (State == bundle_state::input)
         // TODO SYCL2020 exception
-        throw sycl::runtime_error("Program and kernel_bundle state mismatch",
+        throw sycl::runtime_error(errc::invalid,
+                                  "Program and kernel_bundle state mismatch",
                                   PI_INVALID_VALUE);
       if (State == bundle_state::executable)
-        Plugin.call<PiApiKind::piProgramLink>(ContextImpl->getHandleRef(), 1,
-                                              &Dev, nullptr, 1, &PiProgram,
-                                              nullptr, nullptr, &PiProgram);
+        Plugin.call<errc::build, PiApiKind::piProgramLink>(
+            ContextImpl->getHandleRef(), 1, &Dev, nullptr, 1, &PiProgram,
+            nullptr, nullptr, &PiProgram);
       break;
     case (PI_PROGRAM_BINARY_TYPE_EXECUTABLE):
       if (State == bundle_state::input || State == bundle_state::object)
         // TODO SYCL2020 exception
-        throw sycl::runtime_error("Program and kernel_bundle state mismatch",
+        throw sycl::runtime_error(errc::invalid,
+                                  "Program and kernel_bundle state mismatch",
                                   PI_INVALID_VALUE);
       break;
     }
diff --git a/sycl/source/detail/device_info.hpp b/sycl/source/detail/device_info.hpp
index b858a34bc0201..4ae7c7c13e483 100644
--- a/sycl/source/detail/device_info.hpp
+++ b/sycl/source/detail/device_info.hpp
@@ -1083,7 +1083,7 @@ inline bool get_device_info_host<info::device::preferred_interop_user_sync>() {
 
 template <> inline device get_device_info_host<info::device::parent_device>() {
   // TODO: implement host device partitioning
-  throw runtime_error(
+  throw invalid_object_error(
       "Partitioning to subdevices of the host device is not implemented yet",
       PI_INVALID_DEVICE);
 }
diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp
index 7ba084f315890..5dbcd463e560c 100644
--- a/sycl/source/detail/program_manager/program_manager.cpp
+++ b/sycl/source/detail/program_manager/program_manager.cpp
@@ -594,11 +594,9 @@ ProgramManager::getOrCreateKernel(OSModuleHandle M,
   auto BuildF = [&Program, &KernelName, &ContextImpl] {
     PiKernelT *Result = nullptr;
 
-    // TODO need some user-friendly error/exception
-    // instead of currently obscure one
     const detail::plugin &Plugin = ContextImpl->getPlugin();
-    Plugin.call<PiApiKind::piKernelCreate>(Program, KernelName.c_str(),
-                                           &Result);
+    Plugin.call<errc::kernel_not_supported, PiApiKind::piKernelCreate>(
+        Program, KernelName.c_str(), &Result);
 
     // Some PI Plugins (like OpenCL) require this call to enable USM
     // For others, PI will turn this into a NOP.
diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp
index a2d24846ba555..9b45d9820f949 100644
--- a/sycl/source/detail/queue_impl.hpp
+++ b/sycl/source/detail/queue_impl.hpp
@@ -109,7 +109,7 @@ class queue_impl {
                             "discard_events and enable_profiling.");
     }
     if (!Context->hasDevice(Device))
-      throw cl::sycl::invalid_parameter_error(
+      throw cl::sycl::invalid_object_error(
           "Queue cannot be constructed with the given context and device "
           "as the context does not contain the given device.",
           PI_INVALID_DEVICE);
diff --git a/sycl/source/detail/scheduler/scheduler_helpers.cpp b/sycl/source/detail/scheduler/scheduler_helpers.cpp
index 141fddd0dac2c..d3233117283dc 100644
--- a/sycl/source/detail/scheduler/scheduler_helpers.cpp
+++ b/sycl/source/detail/scheduler/scheduler_helpers.cpp
@@ -42,10 +42,10 @@ void initStream(StreamImplPtr Stream, QueueImplPtr Queue) {
     auto FlushBufAcc =
         StrBufs->FlushBuf.get_access<access::mode::discard_write,
                                      access::target::host_buffer>(
-            cgh, range<1>(FlushBufSize), id<1>(0));
+            cgh, range<1>(1), id<1>(0));
     cgh.host_task([=] {
       char *FlushBufPtr = FlushBufAcc.get_pointer();
-      std::memset(FlushBufPtr, 0, FlushBufAcc.get_size());
+      std::memset(FlushBufPtr, 0, FlushBufSize);
     });
   });
 }
diff --git a/sycl/test/basic_tests/offset-accessor-get_pointer.cpp b/sycl/test/basic_tests/offset-accessor-get_pointer.cpp
index 11779aeac410f..c2624e39abcbc 100644
--- a/sycl/test/basic_tests/offset-accessor-get_pointer.cpp
+++ b/sycl/test/basic_tests/offset-accessor-get_pointer.cpp
@@ -36,7 +36,7 @@ void test_across_ranges() {
     buffer<int, 3> buf_3D(v3.data(), range_3D);
 
     myQueue.submit([&](handler &cgh) {
-      auto acc_1D = buf_1D.get_access<r_w>(cgh, {count}, {10});
+      auto acc_1D = buf_1D.get_access<r_w>(cgh, {2}, {10});
       auto acc_2D = buf_2D.get_access<r_w>(cgh, {2, 2}, {1, 1});
       auto acc_3D = buf_3D.get_access<r_w>(cgh, {2, 2, 2}, {1, 1, 1});
       cgh.single_task<class task>([=] {
@@ -54,7 +54,7 @@ void test_across_ranges() {
     });
     myQueue.wait();
     // now host access - we offset by one more than the device test
-    auto acc_1D = buf_1D.get_access<r_w>({count}, {11});
+    auto acc_1D = buf_1D.get_access<r_w>({2}, {11});
     auto acc_2D = buf_2D.get_access<r_w>({2, 2}, {1, 2});
     auto acc_3D = buf_3D.get_access<r_w>({2, 2, 2}, {1, 1, 2});
     acc_1D.get_pointer()[1] = 4; // s.b. offset 1