diff --git a/.editorconfig b/.editorconfig
new file mode 100644
index 000000000..3e53dc366
--- /dev/null
+++ b/.editorconfig
@@ -0,0 +1,7 @@
+# editorconfig.org
+
+root = true
+
+[*]
+end_of_line = lf
+insert_final_newline = true
diff --git a/ci/computecpp.filter b/ci/computecpp.filter
index b8d8d44d3..f9ac46365 100644
--- a/ci/computecpp.filter
+++ b/ci/computecpp.filter
@@ -10,6 +10,7 @@ exceptions
 group
 hierarchical
 host_task
+id
 image
 kernel
 kernel_args
diff --git a/tests/common/common.h b/tests/common/common.h
index c71cae045..2bd0e1e5d 100644
--- a/tests/common/common.h
+++ b/tests/common/common.h
@@ -2,7 +2,7 @@
 //
 //  SYCL 2020 Conformance Test Suite
 //
-//  Copyright (c) 2020-2021 The Khronos Group Inc.
+//  Copyright (c) 2020-2022 The Khronos Group Inc.
 //  Copyright:	(c) 2017 by Codeplay Software LTD. All Rights Reserved.
 //
 *******************************************************************************/
@@ -17,10 +17,12 @@
 #include "../../util/math_vector.h"
 #include "../../util/proxy.h"
 #include "../../util/test_base.h"
-#include "../common/cts_async_handler.h"
-#include "../common/cts_selector.h"
-#include "../common/get_cts_object.h"
+
+#include "cts_async_handler.h"
+#include "cts_selector.h"
+#include "get_cts_object.h"
 #include "macros.h"
+#include "string_makers.h"
 
 #include <cinttypes>
 #include <numeric>
diff --git a/tests/common/device_eval.h b/tests/common/device_eval.h
new file mode 100644
index 000000000..96bdd7c54
--- /dev/null
+++ b/tests/common/device_eval.h
@@ -0,0 +1,36 @@
+/*******************************************************************************
+//
+//  SYCL 2020 Conformance Test Suite
+//
+//  Copyright (c) 2022 The Khronos Group Inc.
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_COMMON_DEVICE_EVAL_H
+#define __SYCLCTS_TESTS_COMMON_DEVICE_EVAL_H
+
+#include <sycl/sycl.hpp>
+
+#define DEVICE_EVAL_T(T, expr)                                      \
+  ([=] {                                                            \
+    sycl::buffer<std::decay_t<T>, 1> result_buf{1};                 \
+    sycl_cts::util::get_cts_object::queue()                         \
+        .submit([=, &result_buf](sycl::handler& cgh) {              \
+          sycl::accessor result{result_buf, cgh, sycl::write_only}; \
+          cgh.single_task([=] { result[0] = expr; });               \
+        })                                                          \
+        .wait_and_throw();                                          \
+    sycl::host_accessor acc{result_buf, sycl::read_only};           \
+    return acc[0];                                                  \
+  })()
+
+/**
+ * Evaluates a given expression on the SYCL device and returns the result.
+ *
+ * Limitations:
+ * - Operands must exist in surrounding scope ([=] capture).
+ * - No lambda expressions (requires C++20). Use DEVICE_EVAL_T instead.
+ */
+#define DEVICE_EVAL(expr) DEVICE_EVAL_T(decltype(expr), expr)
+
+#endif  // __SYCLCTS_TESTS_COMMON_DEVICE_EVAL_H
diff --git a/tests/common/string_makers.h b/tests/common/string_makers.h
new file mode 100644
index 000000000..261d1fc68
--- /dev/null
+++ b/tests/common/string_makers.h
@@ -0,0 +1,35 @@
+/*******************************************************************************
+//
+//  SYCL 2020 Conformance Test Suite
+//
+//  Copyright (c) 2022 The Khronos Group Inc.
+//
+*******************************************************************************/
+
+#ifndef __SYCLCTS_TESTS_COMMON_STRING_MAKERS_H
+#define __SYCLCTS_TESTS_COMMON_STRING_MAKERS_H
+
+#include <sstream>
+
+#include <catch2/catch_tostring.hpp>
+#include <sycl/sycl.hpp>
+
+namespace Catch {
+template <int Dimensions>
+struct StringMaker<sycl::id<Dimensions>> {
+  static std::string convert(const sycl::id<Dimensions>& id) {
+    std::stringstream ss;
+    ss << "{";
+    for (int d = 0; d < Dimensions; ++d) {
+      ss << id[d];
+      if (d != Dimensions - 1) {
+        ss << ", ";
+      }
+    }
+    ss << "}";
+    return ss.str();
+  }
+};
+}  // namespace Catch
+
+#endif  // __SYCLCTS_TESTS_COMMON_STRING_MAKERS_H
diff --git a/tests/id/id.cpp b/tests/id/id.cpp
new file mode 100644
index 000000000..89f252283
--- /dev/null
+++ b/tests/id/id.cpp
@@ -0,0 +1,481 @@
+/*******************************************************************************
+//
+//  SYCL 2020 Conformance Test Suite
+//
+//  Copyright (c) 2022 The Khronos Group Inc.
+//
+*******************************************************************************/
+
+#include <type_traits>
+#include <utility>
+
+#include <catch2/catch_template_test_macros.hpp>
+
+#include "../common/common.h"
+#include "../common/device_eval.h"
+
+using namespace sycl_cts;
+
+TEST_CASE("id provides a default constructor", "[id]") {
+  using sycl::id;
+  STATIC_CHECK(std::is_default_constructible_v<id<1>>);
+  STATIC_CHECK(std::is_default_constructible_v<id<2>>);
+  STATIC_CHECK(std::is_default_constructible_v<id<3>>);
+
+  CHECK(id<1>{} == id<1>{0});
+  CHECK(id<2>{} == id<2>{0, 0});
+  CHECK(id<3>{} == id<3>{0, 0, 0});
+
+  CHECK(DEVICE_EVAL(id<1>{}) == id<1>{0});
+  CHECK(DEVICE_EVAL(id<2>{}) == id<2>{0, 0});
+  CHECK(DEVICE_EVAL(id<3>{}) == id<3>{0, 0, 0});
+}
+
+TEST_CASE("id provides specialized constructors for each dimensionality",
+          "[id]") {
+  using sycl::id;
+
+  STATIC_CHECK(std::is_constructible_v<id<1>, size_t>);
+  STATIC_CHECK(std::is_constructible_v<id<2>, size_t, size_t>);
+  STATIC_CHECK(std::is_constructible_v<id<3>, size_t, size_t, size_t>);
+
+  const id<1> a{5};
+  CHECK(a[0] == 5);
+
+  const id<2> b{5, 8};
+  CHECK(b[0] == 5);
+  CHECK(b[1] == 8);
+
+  const id<3> c{5, 8, 3};
+  CHECK(c[0] == 5);
+  CHECK(c[1] == 8);
+  CHECK(c[2] == 3);
+
+  CHECK(DEVICE_EVAL(id<1>{5}) == id<1>{5});
+  CHECK(DEVICE_EVAL((id<2>{5, 8})) == id<2>{5, 8});
+  CHECK(DEVICE_EVAL((id<3>{5, 8, 3})) == id<3>{5, 8, 3});
+}
+
+// id h[elper] type for creating ids in templated contexts
+template <int Dimensions>
+using idh = util::get_cts_object::id<Dimensions>;
+
+// TODO SPEC: Do common by-value semantics require trivially copyable?
+// See also https://github.com/KhronosGroup/SYCL-Docs/issues/210
+TEMPLATE_TEST_CASE_SIG("id provides common by-value semantics", "[id]",
+                       ((int D), D), 1, 2, 3) {
+  using sycl::id;
+
+  SECTION("copy constructor") {
+    CHECK(std::is_trivially_copy_constructible_v<id<D>>);
+    CHECK(DEVICE_EVAL(std::is_trivially_copy_constructible_v<id<D>>));
+
+    const auto copy = [] {
+      const auto a = idh<D>::get(5, 8, 3);
+      id<D> b{a};
+      return b;
+    };
+    CHECK(copy() == idh<D>::get(5, 8, 3));
+    CHECK(DEVICE_EVAL_T(id<D>, copy()) == idh<D>::get(5, 8, 3));
+  }
+
+  SECTION("copy assignment operator") {
+    CHECK(std::is_trivially_copy_assignable_v<id<D>>);
+    CHECK(DEVICE_EVAL(std::is_trivially_copy_assignable_v<id<D>>));
+
+    const auto copy = [] {
+      const auto a = idh<D>::get(5, 8, 3);
+      id<D> b;
+      b = a;
+      return b;
+    };
+    CHECK(copy() == idh<D>::get(5, 8, 3));
+    CHECK(DEVICE_EVAL_T(id<D>, copy()) == idh<D>::get(5, 8, 3));
+  }
+
+  SECTION("destructor") {
+    CHECK(std::is_trivially_destructible_v<id<D>>);
+    CHECK(DEVICE_EVAL(std::is_trivially_destructible_v<id<D>>));
+  }
+
+  SECTION("move constructor") {
+    CHECK(std::is_trivially_move_constructible_v<id<D>>);
+    CHECK(DEVICE_EVAL(std::is_trivially_move_constructible_v<id<D>>));
+
+    const auto move = [] {
+      auto a = idh<D>::get(5, 8, 3);
+      id<D> b{std::move(a)};
+      return b;
+    };
+    CHECK(move() == idh<D>::get(5, 8, 3));
+    CHECK(DEVICE_EVAL_T(id<D>, move()) == idh<D>::get(5, 8, 3));
+  }
+
+  SECTION("move assignment operator") {
+    CHECK(std::is_trivially_move_assignable_v<id<D>>);
+    CHECK(DEVICE_EVAL(std::is_trivially_move_assignable_v<id<D>>));
+
+    const auto move = [] {
+      auto a = idh<D>::get(5, 8, 3);
+      id<D> b;
+      b = std::move(a);
+      return b;
+    };
+    CHECK(move() == idh<D>::get(5, 8, 3));
+    CHECK(DEVICE_EVAL_T(id<D>, move()) == idh<D>::get(5, 8, 3));
+  }
+
+  SECTION("equality operators") {
+    auto a1 = idh<D>::get(5, 8, 3);
+    auto a2 = idh<D>::get(5, 8, 3);
+    auto b1 = idh<D>::get(4, 8, 2);
+
+    CHECK(a1 == a1);
+    CHECK(a1 == a2);
+    CHECK(a2 == a1);
+    CHECK(b1 == b1);
+    CHECK_FALSE(a1 == b1);
+    CHECK_FALSE(b1 == a1);
+    CHECK_FALSE(a2 == b1);
+
+    CHECK(DEVICE_EVAL(a1 == a1));
+    CHECK(DEVICE_EVAL(a1 == a2));
+    CHECK(DEVICE_EVAL(a2 == a1));
+    CHECK(DEVICE_EVAL(b1 == b1));
+    CHECK_FALSE(DEVICE_EVAL(a1 == b1));
+    CHECK_FALSE(DEVICE_EVAL(b1 == a1));
+    CHECK_FALSE(DEVICE_EVAL(a2 == b1));
+
+    CHECK_FALSE(a1 != a1);
+    CHECK_FALSE(a1 != a2);
+    CHECK_FALSE(a2 != a1);
+    CHECK_FALSE(b1 != b1);
+    CHECK(a1 != b1);
+    CHECK(b1 != a1);
+    CHECK(a2 != b1);
+
+    CHECK_FALSE(DEVICE_EVAL(a1 != a1));
+    CHECK_FALSE(DEVICE_EVAL(a1 != a2));
+    CHECK_FALSE(DEVICE_EVAL(a2 != a1));
+    CHECK_FALSE(DEVICE_EVAL(b1 != b1));
+    CHECK(DEVICE_EVAL(a1 != b1));
+    CHECK(DEVICE_EVAL(b1 != a1));
+    CHECK(DEVICE_EVAL(a2 != b1));
+  }
+}
+
+TEMPLATE_TEST_CASE_SIG("id can be implicitly conversion-constructed from range",
+                       "[id]", ((int D), D), 1, 2, 3) {
+  using sycl::id;
+
+  const auto convert = [] {
+    const auto r = util::get_cts_object::range<D>::get(5, 8, 3);
+    sycl::id<D> a;
+    // Use assignment operator to trigger implicit conversion
+    a = r;
+    return a;
+  };
+
+  CHECK(convert() == idh<D>::get(5, 8, 3));
+  CHECK(DEVICE_EVAL_T(id<D>, convert()) == idh<D>::get(5, 8, 3));
+}
+
+TEMPLATE_TEST_CASE_SIG("id can be implicitly conversion-constructed from item",
+                       "[id]", ((int D), D), 1, 2, 3) {
+  using sycl::id;
+
+  auto q = util::get_cts_object::queue();
+  sycl::buffer<id<D>, 1> result_buf{1};
+  const auto r = util::get_cts_object::range<D>::get(5, 8, 3);
+  q.submit([r, &result_buf](sycl::handler& cgh) {
+     sycl::accessor result{result_buf, cgh, sycl::write_only};
+     cgh.parallel_for(r, [=](sycl::item<D> itm) {
+       if (itm.get_id() == id<D>{r} - 1) {
+         // Use assignment operator to trigger implicit conversion
+         result[0] = itm;
+       }
+     });
+   }).wait_and_throw();
+  sycl::host_accessor acc{result_buf, sycl::read_only};
+
+  CHECK(acc[0] == idh<D>::get(4, 7, 2));
+}
+
+TEMPLATE_TEST_CASE_SIG("id supports get() and operator[]", "[id]", ((int D), D),
+                       1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const size_t values[] = {5, 8, 3};
+
+  for (int i = 0; i < D; ++i) {
+    CHECK(a.get(i) == values[i]);
+    CHECK(a[i] == values[i]);
+    CHECK(DEVICE_EVAL(a.get(i)) == values[i]);
+    CHECK(DEVICE_EVAL(a[i]) == values[i]);
+  }
+
+  const auto assign_component = [](auto x, auto c, auto v) {
+    x[c] = v;
+    return x;
+  };
+
+  using sycl::id;
+
+  CHECK(assign_component(a, 0, 7) == idh<D>::get(7, 8, 3));
+  CHECK(DEVICE_EVAL_T(id<D>, assign_component(a, 0, 7)) ==
+        idh<D>::get(7, 8, 3));
+
+  if (D >= 2) {
+    CHECK(assign_component(a, 1, 9) == idh<D>::get(5, 9, 3));
+    CHECK(DEVICE_EVAL_T(id<D>, assign_component(a, 1, 9)) ==
+          idh<D>::get(5, 9, 3));
+  }
+
+  if (D == 3) {
+    CHECK(assign_component(a, 2, 11) == idh<D>::get(5, 8, 11));
+    CHECK(DEVICE_EVAL_T(id<D>, assign_component(a, 2, 11)) ==
+          idh<D>::get(5, 8, 11));
+  }
+}
+
+TEST_CASE("id can be converted to size_t if Dimensions == 1", "[id]") {
+  using sycl::id;
+  const auto convert = [] {
+    const sycl::id a{42};
+    const size_t b = a;
+    return b;
+  };
+  CHECK(convert() == 42);
+  CHECK(DEVICE_EVAL(convert()) == 42);
+}
+
+TEMPLATE_TEST_CASE_SIG(
+    "id supports various binary operators of the form `id OP id`", "[id]",
+    ((int D), D), 1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const auto b = idh<D>::get(4, 8, 2);
+
+  CHECK(a + b == idh<D>::get(9, 16, 5));
+  CHECK(a - b == idh<D>::get(1, 0, 1));
+  CHECK(a * b == idh<D>::get(20, 64, 6));
+  CHECK(a / b == idh<D>::get(1, 1, 1));
+  CHECK(a % b == idh<D>::get(1, 0, 1));
+  CHECK(a << b == idh<D>::get(80, 2048, 12));
+  CHECK(a >> b == idh<D>::get(0, 0, 0));
+  CHECK((a & b) == idh<D>::get(4, 8, 2));
+  CHECK((a | b) == idh<D>::get(5, 8, 3));
+  CHECK((a ^ b) == idh<D>::get(1, 0, 1));
+  CHECK((a && b) == idh<D>::get(1, 1, 1));
+  CHECK((a || b) == idh<D>::get(1, 1, 1));
+  CHECK((a < b) == idh<D>::get(0, 0, 0));
+  CHECK((a > b) == idh<D>::get(1, 0, 1));
+  CHECK((a <= b) == idh<D>::get(0, 1, 0));
+  CHECK((a >= b) == idh<D>::get(1, 1, 1));
+
+  CHECK(DEVICE_EVAL(a + b) == idh<D>::get(9, 16, 5));
+  CHECK(DEVICE_EVAL(a - b) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL(a * b) == idh<D>::get(20, 64, 6));
+  CHECK(DEVICE_EVAL(a / b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a % b) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL(a << b) == idh<D>::get(80, 2048, 12));
+  CHECK(DEVICE_EVAL(a >> b) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL(a & b) == idh<D>::get(4, 8, 2));
+  CHECK(DEVICE_EVAL(a | b) == idh<D>::get(5, 8, 3));
+  CHECK(DEVICE_EVAL(a ^ b) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL(a && b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a || b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a < b) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL(a > b) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL(a <= b) == idh<D>::get(0, 1, 0));
+  CHECK(DEVICE_EVAL(a >= b) == idh<D>::get(1, 1, 1));
+}
+
+TEMPLATE_TEST_CASE_SIG(
+    "id supports various binary operators of the form `id OP size_t` and "
+    "`size_t OP id`",
+    "[id]", ((int D), D), 1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const size_t b = 3;
+
+  CHECK(a + b == idh<D>::get(8, 11, 6));
+  CHECK(b + a == idh<D>::get(8, 11, 6));
+  CHECK(a - b == idh<D>::get(2, 5, 0));
+  CHECK(b - a == idh<D>::get(-2ul, -5ul, 0));
+  CHECK(a * b == idh<D>::get(15, 24, 9));
+  CHECK(b * a == idh<D>::get(15, 24, 9));
+  CHECK(a / b == idh<D>::get(1, 2, 1));
+  CHECK(b / a == idh<D>::get(0, 0, 1));
+  CHECK(a % b == idh<D>::get(2, 2, 0));
+  CHECK(b % a == idh<D>::get(3, 3, 0));
+  CHECK(a << b == idh<D>::get(40, 64, 24));
+  CHECK(b << a == idh<D>::get(96, 768, 24));
+  CHECK(a >> b == idh<D>::get(0, 1, 0));
+  CHECK(b >> a == idh<D>::get(0, 0, 0));
+  CHECK((a & b) == idh<D>::get(1, 0, 3));
+  CHECK((b & a) == idh<D>::get(1, 0, 3));
+  CHECK((a | b) == idh<D>::get(7, 11, 3));
+  CHECK((b | a) == idh<D>::get(7, 11, 3));
+  CHECK((a ^ b) == idh<D>::get(6, 11, 0));
+  CHECK((b ^ a) == idh<D>::get(6, 11, 0));
+  CHECK((a && b) == idh<D>::get(1, 1, 1));
+  CHECK((b && a) == idh<D>::get(1, 1, 1));
+  CHECK((a || b) == idh<D>::get(1, 1, 1));
+  CHECK((b || a) == idh<D>::get(1, 1, 1));
+  CHECK((a < b) == idh<D>::get(0, 0, 0));
+  CHECK((b < a) == idh<D>::get(1, 1, 0));
+  CHECK((a > b) == idh<D>::get(1, 1, 0));
+  CHECK((b > a) == idh<D>::get(0, 0, 0));
+  CHECK((a <= b) == idh<D>::get(0, 0, 1));
+  CHECK((b <= a) == idh<D>::get(1, 1, 1));
+  CHECK((a >= b) == idh<D>::get(1, 1, 1));
+  CHECK((b >= a) == idh<D>::get(0, 0, 1));
+
+  CHECK(DEVICE_EVAL(a + b) == idh<D>::get(8, 11, 6));
+  CHECK(DEVICE_EVAL(b + a) == idh<D>::get(8, 11, 6));
+  CHECK(DEVICE_EVAL(a - b) == idh<D>::get(2, 5, 0));
+  CHECK(DEVICE_EVAL(b - a) == idh<D>::get(-2ul, -5ul, 0));
+  CHECK(DEVICE_EVAL(a * b) == idh<D>::get(15, 24, 9));
+  CHECK(DEVICE_EVAL(b * a) == idh<D>::get(15, 24, 9));
+  CHECK(DEVICE_EVAL(a / b) == idh<D>::get(1, 2, 1));
+  CHECK(DEVICE_EVAL(b / a) == idh<D>::get(0, 0, 1));
+  CHECK(DEVICE_EVAL(a % b) == idh<D>::get(2, 2, 0));
+  CHECK(DEVICE_EVAL(b % a) == idh<D>::get(3, 3, 0));
+  CHECK(DEVICE_EVAL(a << b) == idh<D>::get(40, 64, 24));
+  CHECK(DEVICE_EVAL(b << a) == idh<D>::get(96, 768, 24));
+  CHECK(DEVICE_EVAL(a >> b) == idh<D>::get(0, 1, 0));
+  CHECK(DEVICE_EVAL(b >> a) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL(a & b) == idh<D>::get(1, 0, 3));
+  CHECK(DEVICE_EVAL(b & a) == idh<D>::get(1, 0, 3));
+  CHECK(DEVICE_EVAL(a | b) == idh<D>::get(7, 11, 3));
+  CHECK(DEVICE_EVAL(b | a) == idh<D>::get(7, 11, 3));
+  CHECK(DEVICE_EVAL(a ^ b) == idh<D>::get(6, 11, 0));
+  CHECK(DEVICE_EVAL(b ^ a) == idh<D>::get(6, 11, 0));
+  CHECK(DEVICE_EVAL(a && b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(b && a) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a || b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(b || a) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a < b) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL(b < a) == idh<D>::get(1, 1, 0));
+  CHECK(DEVICE_EVAL(a > b) == idh<D>::get(1, 1, 0));
+  CHECK(DEVICE_EVAL(b > a) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL(a <= b) == idh<D>::get(0, 0, 1));
+  CHECK(DEVICE_EVAL(b <= a) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(a >= b) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL(b >= a) == idh<D>::get(0, 0, 1));
+}
+
+#define COMPOUND_OP(operand_value, expr) \
+  ([=](auto x) { return expr, x; })(operand_value)
+
+TEMPLATE_TEST_CASE_SIG(
+    "id supports various compound binary operators of the form `id OP= id`",
+    "[id]", ((int D), D), 1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const auto b = idh<D>::get(4, 8, 2);
+
+  CHECK(COMPOUND_OP(a, x += b) == idh<D>::get(9, 16, 5));
+  CHECK(COMPOUND_OP(a, x -= b) == idh<D>::get(1, 0, 1));
+  CHECK(COMPOUND_OP(a, x *= b) == idh<D>::get(20, 64, 6));
+  CHECK(COMPOUND_OP(a, x /= b) == idh<D>::get(1, 1, 1));
+  CHECK(COMPOUND_OP(a, x %= b) == idh<D>::get(1, 0, 1));
+  CHECK(COMPOUND_OP(a, x <<= b) == idh<D>::get(80, 2048, 12));
+  CHECK(COMPOUND_OP(a, x >>= b) == idh<D>::get(0, 0, 0));
+  CHECK(COMPOUND_OP(a, x &= b) == idh<D>::get(4, 8, 2));
+  CHECK(COMPOUND_OP(a, x |= b) == idh<D>::get(5, 8, 3));
+  CHECK(COMPOUND_OP(a, x ^= b) == idh<D>::get(1, 0, 1));
+
+  using sycl::id;
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x += b)) == idh<D>::get(9, 16, 5));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x -= b)) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x *= b)) == idh<D>::get(20, 64, 6));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x /= b)) == idh<D>::get(1, 1, 1));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x %= b)) == idh<D>::get(1, 0, 1));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x <<= b)) ==
+        idh<D>::get(80, 2048, 12));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x >>= b)) == idh<D>::get(0, 0, 0));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x &= b)) == idh<D>::get(4, 8, 2));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x |= b)) == idh<D>::get(5, 8, 3));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x ^= b)) == idh<D>::get(1, 0, 1));
+}
+
+TEMPLATE_TEST_CASE_SIG(
+    "id supports various compound binary operators of the form `id OP= size_t`",
+    "[id]", ((int D), D), 1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const size_t b = 3;
+
+  CHECK(COMPOUND_OP(a, x += b) == idh<D>::get(8, 11, 6));
+  CHECK(COMPOUND_OP(a, x -= b) == idh<D>::get(2, 5, 0));
+  CHECK(COMPOUND_OP(a, x *= b) == idh<D>::get(15, 24, 9));
+  CHECK(COMPOUND_OP(a, x /= b) == idh<D>::get(1, 2, 1));
+  CHECK(COMPOUND_OP(a, x %= b) == idh<D>::get(2, 2, 0));
+  CHECK(COMPOUND_OP(a, x <<= b) == idh<D>::get(40, 64, 24));
+  CHECK(COMPOUND_OP(a, x >>= b) == idh<D>::get(0, 1, 0));
+  CHECK(COMPOUND_OP(a, x &= b) == idh<D>::get(1, 0, 3));
+  CHECK(COMPOUND_OP(a, x |= b) == idh<D>::get(7, 11, 3));
+  CHECK(COMPOUND_OP(a, x ^= b) == idh<D>::get(6, 11, 0));
+
+  using sycl::id;
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x += b)) == idh<D>::get(8, 11, 6));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x -= b)) == idh<D>::get(2, 5, 0));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x *= b)) == idh<D>::get(15, 24, 9));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x /= b)) == idh<D>::get(1, 2, 1));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x %= b)) == idh<D>::get(2, 2, 0));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x <<= b)) ==
+        idh<D>::get(40, 64, 24));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x >>= b)) == idh<D>::get(0, 1, 0));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x &= b)) == idh<D>::get(1, 0, 3));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x |= b)) == idh<D>::get(7, 11, 3));
+  CHECK(DEVICE_EVAL_T(id<D>, COMPOUND_OP(a, x ^= b)) == idh<D>::get(6, 11, 0));
+}
+
+#undef COMPOUND_OP
+
+TEMPLATE_TEST_CASE_SIG("id supports unary +/- operators", "[id]", ((int D), D),
+                       1, 2, 3) {
+  const auto a = idh<D>::get(5, 8, 3);
+  const auto b = idh<D>::get(-5, -8, -3);
+  CHECK(+a == a);
+  CHECK(-a == b);
+  CHECK(+b == b);
+  CHECK(-b == a);
+
+  CHECK(DEVICE_EVAL(+a) == a);
+  CHECK(DEVICE_EVAL(-a) == b);
+  CHECK(DEVICE_EVAL(+b) == b);
+  CHECK(DEVICE_EVAL(-b) == a);
+}
+
+TEMPLATE_TEST_CASE_SIG(
+    "id supports pre- and postfix increment/decrement operators", "[id]",
+    ((int D), D), 1, 2, 3) {
+#define INC_DEC_OP(operand_value, expr) \
+  ([=](auto x) { return std::pair{expr, x}; })(operand_value)
+
+  const auto a = idh<D>::get(5, 8, 3);
+  const auto b = idh<D>::get(6, 9, 4);
+  const auto c = idh<D>::get(4, 7, 2);
+
+  CHECK(INC_DEC_OP(a, ++x) == std::pair{b, b});
+  CHECK(INC_DEC_OP(a, --x) == std::pair{c, c});
+  CHECK(INC_DEC_OP(a, x++) == std::pair{a, b});
+  CHECK(INC_DEC_OP(a, x--) == std::pair{a, c});
+
+  using id_pair = std::pair<sycl::id<D>, sycl::id<D>>;
+
+  CHECK(DEVICE_EVAL_T(id_pair, INC_DEC_OP(a, ++x)) == std::pair{b, b});
+  CHECK(DEVICE_EVAL_T(id_pair, INC_DEC_OP(a, --x)) == std::pair{c, c});
+  CHECK(DEVICE_EVAL_T(id_pair, INC_DEC_OP(a, x++)) == std::pair{a, b});
+  CHECK(DEVICE_EVAL_T(id_pair, INC_DEC_OP(a, x--)) == std::pair{a, c});
+
+#undef INC_DEC_OP
+}
+
+TEST_CASE("id can deduce dimensionality from constructor parameters", "[id]") {
+  using sycl::id;
+  CHECK(std::is_same_v<decltype(id{5}), id<1>>);
+  CHECK(std::is_same_v<decltype(id{5, 8}), id<2>>);
+  CHECK(std::is_same_v<decltype(id{5, 8, 3}), id<3>>);
+  CHECK(DEVICE_EVAL((std::is_same_v<decltype(id{5}), id<1>>)));
+  CHECK(DEVICE_EVAL((std::is_same_v<decltype(id{5, 8}), id<2>>)));
+  CHECK(DEVICE_EVAL((std::is_same_v<decltype(id{5, 8, 3}), id<3>>)));
+}
diff --git a/tests/id/id_api.cpp b/tests/id/id_api.cpp
deleted file mode 100644
index c4543ad2e..000000000
--- a/tests/id/id_api.cpp
+++ /dev/null
@@ -1,305 +0,0 @@
-/*******************************************************************************
-//
-//  SYCL 2020 Conformance Test Suite
-//
-//  Copyright:	(c) 2017 by Codeplay Software LTD. All Rights Reserved.
-//
-*******************************************************************************/
-
-#include "../common/common.h"
-
-#define TEST_NAME id_api
-
-namespace id_api__ {
-using namespace sycl_cts;
-
-template <int dims>
-class test_kernel {};
-
-template <int dims>
-void test_id_kernels(
-    sycl::id<dims> id,
-    sycl::accessor<int, 1, sycl::access_mode::read_write,
-                       sycl::target::device>
-        error_ptr,
-    int m_iteration) {
-  sycl::id<dims> id_two(id * 2);
-  sycl::id<dims> id_three(id);
-  size_t integer = 16;
-  for (int j = 0; j < dims; j++) {
-    if (id_two.get(j) == 0) {
-      id_two[j] = 1;
-    }
-  }
-  const sycl::id<dims> id_two_const(id_two);
-  const sycl::id<dims> id_const(id);
-
-  // operators
-  // +=
-  INDEX_ASSIGNMENT_TESTS(+=, +, id, id_two, id_three);
-
-  // -=
-  INDEX_ASSIGNMENT_TESTS(-=, -, id, id_two, id_three);
-
-  // *=
-  INDEX_ASSIGNMENT_TESTS(*=, *, id, id_two, id_three);
-
-  // /=
-  INDEX_ASSIGNMENT_TESTS(/=, /, id, id_two, id_three);
-
-  // %=
-  INDEX_ASSIGNMENT_TESTS(%=, %, id, id_two, id_three);
-
-  // >>=
-  INDEX_ASSIGNMENT_TESTS(>>=, >>, id, id_two, id_three);
-
-  // <<=
-  INDEX_ASSIGNMENT_TESTS(<<=, <<, id, id_two, id_three);
-
-  // &=
-  INDEX_ASSIGNMENT_TESTS(&=, &, id, id_two, id_three);
-
-  // |=
-  INDEX_ASSIGNMENT_TESTS(|=, |, id, id_two, id_three);
-
-  // ^=
-  INDEX_ASSIGNMENT_TESTS(^=, ^, id, id_two, id_three);
-
-  // check id<dimensions> operatorOP(const id<dimensions> &rhs)
-
-  // *
-  INDEX_KERNEL_TEST(*, id, id_two_const, id_three);
-
-  // /
-  INDEX_KERNEL_TEST(/, id, id_two_const, id_three);
-
-  //+
-  INDEX_KERNEL_TEST(+, id, id_two_const, id_three);
-
-  //-
-  INDEX_KERNEL_TEST(-, id, id_two_const, id_three);
-
-  //%
-  INDEX_KERNEL_TEST(%, id, id_two_const, id_three);
-
-  //<<
-  INDEX_KERNEL_TEST(<<, id, id_two_const, id_three);
-
-  //>>
-  INDEX_KERNEL_TEST(>>, id, id_two_const, id_three);
-
-  //&
-  INDEX_KERNEL_TEST(&, id, id_two_const, id_three);
-
-  //|
-  INDEX_KERNEL_TEST(|, id, id_two_const, id_three);
-
-  //^
-  INDEX_KERNEL_TEST (^, id, id_two_const, id_three);
-
-  // &&
-  INDEX_KERNEL_TEST(&&, id, id_two_const, id_three);
-
-  // ||
-  INDEX_KERNEL_TEST(||, id, id_two_const, id_three);
-
-  // >
-  INDEX_KERNEL_TEST(>, id, id_two_const, id_three);
-
-  // <
-  INDEX_KERNEL_TEST(<, id, id_two_const, id_three);
-
-  // >=
-  INDEX_KERNEL_TEST(>=, id, id_two_const, id_three);
-
-  // <=
-  INDEX_KERNEL_TEST(<=, id, id_two_const, id_three);
-
-  // check == and !=
-  // ==
-  INDEX_EQ_KERNEL_TEST(==, id, id_two);
-
-  // !=
-  INDEX_EQ_KERNEL_TEST(!=, id, id_two);
-
-  // check id<dimensions> operatorOP(const size_t &rhs)
-
-  // *
-  DUAL_SIZE_INDEX_KERNEL_TEST(*, id, integer, id_three);
-
-  // +
-  DUAL_SIZE_INDEX_KERNEL_TEST(+, id, integer, id_three);
-
-  // -
-  DUAL_SIZE_INDEX_KERNEL_TEST(-, id, integer, id_three);
-
-  // /
-  DUAL_SIZE_INDEX_KERNEL_TEST(/, id, integer, id_three);
-
-  // %
-  DUAL_SIZE_INDEX_KERNEL_TEST(%, id, integer, id_three);
-
-  // <<
-  DUAL_SIZE_INDEX_KERNEL_TEST(<<, id, integer, id_three);
-
-  // >>
-  DUAL_SIZE_INDEX_KERNEL_TEST(>>, id, integer, id_three);
-
-  // |
-  DUAL_SIZE_INDEX_KERNEL_TEST(|, id, integer, id_three);
-
-  // ^
-  DUAL_SIZE_INDEX_KERNEL_TEST (^, id, integer, id_three);
-
-  // && id can only be lhs
-  INDEX_SIZE_T_KERNEL_TEST(&&, id, integer, id_three);
-
-  // || id can only be lhs
-  INDEX_SIZE_T_KERNEL_TEST(||, id, integer, id_three);
-
-  // <
-  DUAL_SIZE_INDEX_KERNEL_TEST(<, id, integer, id_three);
-
-  // >
-  DUAL_SIZE_INDEX_KERNEL_TEST(>, id, integer, id_three);
-
-  // <=
-  DUAL_SIZE_INDEX_KERNEL_TEST(<=, id, integer, id_three);
-
-  // >=
-  DUAL_SIZE_INDEX_KERNEL_TEST(>=, id, integer, id_three);
-
-  // check id<dimensions> &operatorOP(const size_t &rhs)
-
-  // +=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(+=, +, id, integer, id_three);
-
-  // -=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(-=, -, id, integer, id_three);
-
-  // *=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(*=, *, id, integer, id_three);
-
-  // /=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(/=, /, id, integer, id_three);
-
-  // %=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(%=, %, id, integer, id_three);
-
-  // >>=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(>>=, >>, id, integer, id_three);
-
-  // <<=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(<<=, <<, id, integer, id_three);
-
-  // &=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(&=, &, id, integer, id_three);
-
-  // |=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(|=, |, id, integer, id_three);
-
-  // ^=
-  INDEX_ASSIGNMENT_INTEGER_TESTS(^=, ^, id, integer, id_three);
-}
-
-template <int dims>
-class test_id {
- public:
-  // golden values
-  static const int m_x = 16;
-  static const int m_y = 32;
-  static const int m_z = 64;
-  static const int m_local = 2;
-  static const int error_size = 200;  // up to 200 possible errors
-  int m_error[error_size];
-
-  void operator()(util::logger &log, sycl::range<dims> global,
-                  sycl::range<dims> local, sycl::queue q) {
-    // for testing get()
-    for (int i = 0; i < error_size; i++) {
-      m_error[i] = 0;  // no error
-    }
-
-    {
-      sycl::buffer<int, 1> error_buffer(m_error,
-                                            sycl::range<1>(error_size));
-
-      q.submit([&](sycl::handler &cgh) {
-        auto my_range = sycl::nd_range<dims>(global, local);
-
-        auto error_ptr =
-            error_buffer.get_access<sycl::access_mode::read_write>(cgh);
-
-        auto my_kernel = ([=](sycl::nd_item<dims> item) {
-          int m_iteration = 0;
-
-          // create check table
-          sycl::id<dims> id = item.get_nd_range().get_global_range();
-
-          size_t check[] = {m_x, m_y, m_z};
-
-          for (int i = 0; i < dims; i++) {
-            if (id.get(i) > check[i] || id[i] > check[i]) {
-              // report an error
-              error_ptr[m_iteration] = __LINE__;
-              m_iteration++;
-            }
-          }
-
-          test_id_kernels<dims>(id, error_ptr,
-                                m_iteration);  // test all in the kernel
-        });
-        cgh.parallel_for<class test_kernel<dims>>(my_range, my_kernel);
-      });
-
-      q.wait_and_throw();
-    }
-    for (int i = 0; i < error_size; i++) {
-      CHECK_VALUE(log, m_error[i], 0, i);
-    }
-  }
-};
-
-/** test sycl::range::get(int index) return size_t
- */
-class TEST_NAME : public util::test_base {
- public:
-  /** return information about this test
-   */
-  void get_info(test_base::info &out) const override {
-    set_test_info(out, TOSTRING(TEST_NAME), TEST_FILE);
-  }
-
-  /** execute the test
-   */
-  void run(util::logger &log) override {
-    {
-      // use across all the dimensions
-      auto my_queue = util::get_cts_object::queue();
-      // templated approach
-      {
-        sycl::range<1> range_1d_g(test_id<1>::m_x);
-        sycl::range<2> range_2d_g(test_id<2>::m_x, test_id<2>::m_y);
-        sycl::range<3> range_3d_g(test_id<3>::m_x, test_id<3>::m_y,
-                                      test_id<3>::m_z);
-
-        sycl::range<1> range_1d_l(test_id<1>::m_local);
-        sycl::range<2> range_2d_l(test_id<2>::m_local, test_id<2>::m_local);
-        sycl::range<3> range_3d_l(test_id<3>::m_local, test_id<3>::m_local,
-                                      test_id<3>::m_local);
-
-        test_id<1> test1d;
-        test1d(log, range_1d_g, range_1d_l, my_queue);
-        test_id<2> test2d;
-        test2d(log, range_2d_g, range_2d_l, my_queue);
-        test_id<3> test3d;
-        test3d(log, range_3d_g, range_3d_l, my_queue);
-      }
-    }
-  }
-};
-
-// construction of this proxy will register the above test
-util::test_proxy<TEST_NAME> proxy;
-
-}  // namespace id_api__