// REQUIRES: any-device
// RUN: %clangxx -fsycl %s -o %t.out
// RUN: %t.out

#include <sycl/sycl.hpp>

#include <cassert>
#include <iostream>
#include <type_traits>

int main() {
  // TODO: uncomment property once it is implemented. now all sycl::queue
  // objects are in-order due to liboffload limitation. Test is intended to
  // check in-order execution.
  sycl::queue Q{/*sycl::property::queue::in_order()*/};
  auto Dev = Q.get_device();
  auto Ctx = Q.get_context();
  constexpr int N = 8;

  auto A = static_cast<int *>(sycl::malloc_shared(N * sizeof(int), Dev, Ctx));

  for (int i = 0; i < N; ++i) {
    A[i] = 1;
  }

  Q.parallel_for<class IntRange>(N, [=](auto i) {
    static_assert(std::is_same<decltype(i), sycl::item<1>>::value,
                  "lambda arg type is unexpected");
    A[i]++;
  });

  Q.parallel_for<class InitRange>({N}, [=](auto i) {
    static_assert(std::is_same<decltype(i), sycl::item<1>>::value,
                  "lambda arg type is unexpected");
    A[i]++;
  });

  Q.parallel_for<class InitRange2D>({4, 2}, [=](auto i) {
    static_assert(std::is_same<decltype(i), sycl::item<2>>::value,
                  "lambda arg type is unexpected");
    A[i.get_linear_id()]++;
  });

  Q.parallel_for<class InitRange3D>({2, 2, 2}, [=](auto i) {
    static_assert(std::is_same<decltype(i), sycl::item<3>>::value,
                  "lambda arg type is unexpected");
    A[i.get_linear_id()]++;
  });

  sycl::nd_range<1> NDR(sycl::range<1>{N}, sycl::range<1>{2});
  Q.parallel_for<class NdRange1D>(NDR, [=](auto nd_i) {
    static_assert(std::is_same<decltype(nd_i), sycl::nd_item<1>>::value,
                  "lambda arg type is unexpected");
    A[nd_i.get_global_id()]++;
  });

  sycl::nd_range<2> NDR2D(sycl::range<2>{4, 2}, sycl::range<2>{2, 1});
  Q.parallel_for<class NdRange2D>(NDR2D, [=](auto nd_i) {
    static_assert(std::is_same<decltype(nd_i), sycl::nd_item<2>>::value,
                  "lambda arg type is unexpected");
    A[nd_i.get_global_linear_id()]++;
  });

  sycl::nd_range<3> NDR3D(sycl::range<3>{2, 2, 2}, sycl::range<3>{1, 2, 2});
  Q.parallel_for<class NdRange3D>(NDR3D, [=](auto nd_i) {
    static_assert(std::is_same<decltype(nd_i), sycl::nd_item<3>>::value,
                  "lambda arg type is unexpected");
    A[nd_i.get_global_linear_id()]++;
  });

  Q.wait();

  bool Fail{};
  for (int i = 0; i < N; i++) {
    Fail |= !(A[i] == 8);
  }
  sycl::free(A, Ctx);
  return Fail;
}