diff --git a/tests/unit_tests/dft/include/compute_inplace.hpp b/tests/unit_tests/dft/include/compute_inplace.hpp
new file mode 100644
index 000000000..a78da23dd
--- /dev/null
+++ b/tests/unit_tests/dft/include/compute_inplace.hpp
@@ -0,0 +1,161 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_COMPUTE_INPLACE_HPP
+#define ONEMKL_COMPUTE_INPLACE_HPP
+
+#include "compute_tester.hpp"
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_in_place_buffer() {
+    if (!init(MemoryAccessModel::buffer)) {
+        return test_skipped;
+    }
+
+    descriptor_t descriptor{ size };
+    descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                         oneapi::mkl::dft::config_value::INPLACE);
+
+    const size_t container_size =
+        domain == oneapi::mkl::dft::domain::REAL ? conjugate_even_size : size;
+
+    std::vector<FwdInputType> inout_host(container_size, static_cast<FwdInputType>(0));
+    std::copy(input.cbegin(), input.cend(), inout_host.begin());
+    sycl::buffer<FwdInputType, 1> inout_buf{ inout_host.data(), sycl::range<1>(container_size) };
+
+    commit_descriptor(descriptor, sycl_queue);
+
+    try {
+        oneapi::mkl::dft::compute_forward<descriptor_t, FwdInputType>(descriptor, inout_buf);
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    if constexpr (domain == oneapi::mkl::dft::domain::REAL) {
+        std::vector<FwdInputType> out_host_ref_conjugate =
+            std::vector<FwdInputType>(conjugate_even_size);
+        for (int i = 0; i < out_host_ref_conjugate.size(); i += 2) {
+            out_host_ref_conjugate[i] = out_host_ref[i / 2].real();
+            out_host_ref_conjugate[i + 1] = out_host_ref[i / 2].imag();
+        }
+        auto acc_host = inout_buf.template get_host_access();
+        EXPECT_TRUE(check_equal_vector(acc_host.get_pointer(), out_host_ref_conjugate.data(),
+                                       inout_host.size(), abs_error_margin, rel_error_margin, std::cout));
+    }
+    else {
+        auto acc_host = inout_buf.template get_host_access();
+        EXPECT_TRUE(check_equal_vector(acc_host.get_pointer(), out_host_ref.data(),
+                                       inout_host.size(), abs_error_margin, rel_error_margin, std::cout));
+    }
+
+    descriptor_t descriptor_back{ size };
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                              oneapi::mkl::dft::config_value::INPLACE);
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+    commit_descriptor(descriptor_back, sycl_queue);
+
+    try {
+        oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                           FwdInputType>(descriptor_back, inout_buf);
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    {
+        auto acc_host = inout_buf.template get_host_access();
+        EXPECT_TRUE(check_equal_vector(acc_host.get_pointer(), input.data(), input.size(),
+                                       abs_error_margin, rel_error_margin, std::cout));
+    }
+    return !::testing::Test::HasFailure();
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_in_place_USM() {
+    if (!init(MemoryAccessModel::usm)) {
+        return test_skipped;
+    }
+
+    descriptor_t descriptor{ size };
+    descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                         oneapi::mkl::dft::config_value::INPLACE);
+    commit_descriptor(descriptor, sycl_queue);
+
+    const size_t container_size =
+        domain == oneapi::mkl::dft::domain::REAL ? conjugate_even_size : size;
+
+    auto ua_input = usm_allocator_t<FwdInputType>(cxt, *dev);
+
+    std::vector<FwdInputType, decltype(ua_input)> inout(container_size, ua_input);
+    std::copy(input.begin(), input.end(), inout.begin());
+
+    try {
+        std::vector<sycl::event> dependencies;
+        sycl::event done = oneapi::mkl::dft::compute_forward<descriptor_t, FwdInputType>(
+            descriptor, inout.data(), dependencies);
+        done.wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    if constexpr (domain == oneapi::mkl::dft::domain::REAL) {
+        std::vector<FwdInputType> out_host_ref_conjugate =
+            std::vector<FwdInputType>(conjugate_even_size);
+        for (int i = 0; i < out_host_ref_conjugate.size(); i += 2) {
+            out_host_ref_conjugate[i] = out_host_ref[i / 2].real();
+            out_host_ref_conjugate[i + 1] = out_host_ref[i / 2].imag();
+        }
+        EXPECT_TRUE(check_equal_vector(inout.data(), out_host_ref_conjugate.data(), inout.size(),
+                                       abs_error_margin, rel_error_margin, std::cout));
+    }
+    else {
+        EXPECT_TRUE(check_equal_vector(inout.data(), out_host_ref.data(), inout.size(),
+                                       abs_error_margin, rel_error_margin, std::cout));
+    }
+
+    descriptor_t descriptor_back{ size };
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                              oneapi::mkl::dft::config_value::INPLACE);
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+    commit_descriptor(descriptor_back, sycl_queue);
+
+    try {
+        std::vector<sycl::event> dependencies;
+        sycl::event done =
+            oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                               FwdInputType>(descriptor_back, inout.data());
+        done.wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    EXPECT_TRUE(
+        check_equal_vector(inout.data(), input.data(), input.size(), abs_error_margin, rel_error_margin, std::cout));
+
+    return !::testing::Test::HasFailure();
+}
+
+#endif //ONEMKL_COMPUTE_INPLACE_HPP
diff --git a/tests/unit_tests/dft/include/compute_inplace_real_real.hpp b/tests/unit_tests/dft/include/compute_inplace_real_real.hpp
new file mode 100644
index 000000000..3b9878932
--- /dev/null
+++ b/tests/unit_tests/dft/include/compute_inplace_real_real.hpp
@@ -0,0 +1,127 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_COMPUTE_INPLACE_REAL_REAL_HPP
+#define ONEMKL_COMPUTE_INPLACE_REAL_REAL_HPP
+
+#include "compute_tester.hpp"
+
+/* Test is not implemented because currently there are no available dft implementations.
+ * These are stubs to make sure that dft::oneapi::mkl::unimplemented exception is thrown */
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_in_place_real_real_USM() {
+    if (!init(MemoryAccessModel::usm)) {
+        return test_skipped;
+    }
+
+    try {
+        descriptor_t descriptor{ size };
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::INPLACE);
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+        commit_descriptor(descriptor, sycl_queue);
+
+        auto ua_input = usm_allocator_t<PrecisionType>(cxt, *dev);
+
+        std::vector<PrecisionType, decltype(ua_input)> inout_re(size, ua_input);
+        std::vector<PrecisionType, decltype(ua_input)> inout_im(size, ua_input);
+        std::copy(input_re.begin(), input_re.end(), inout_re.begin());
+        std::copy(input_im.begin(), input_im.end(), inout_im.begin());
+
+        std::vector<sycl::event> dependencies;
+        sycl::event done = oneapi::mkl::dft::compute_forward<descriptor_t, PrecisionType>(
+            descriptor, inout_re.data(), inout_im.data(), dependencies);
+        done.wait();
+
+        descriptor_t descriptor_back{ size };
+
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                                  oneapi::mkl::dft::config_value::INPLACE);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                                  oneapi::mkl::dft::config_value::REAL_REAL);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+        commit_descriptor(descriptor_back, sycl_queue);
+
+        done =
+            oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                               PrecisionType>(descriptor_back, inout_re.data(),
+                                                              inout_im.data(), dependencies);
+        done.wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    /* Once implementations exist, results will need to be verified */
+    EXPECT_TRUE(false);
+
+    return !::testing::Test::HasFailure();
+}
+
+/* Test is not implemented because currently there are no available dft implementations.
+ * These are stubs to make sure that dft::oneapi::mkl::unimplemented exception is thrown */
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_in_place_real_real_buffer() {
+    if (!init(MemoryAccessModel::buffer)) {
+        return test_skipped;
+    }
+
+    try {
+        descriptor_t descriptor{ size };
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::INPLACE);
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+        commit_descriptor(descriptor, sycl_queue);
+
+        sycl::buffer<PrecisionType, 1> inout_re_buf{ input_re.data(), sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> inout_im_buf{ input_im.data(), sycl::range<1>(size) };
+
+        oneapi::mkl::dft::compute_forward<descriptor_t, PrecisionType>(descriptor, inout_re_buf,
+                                                                       inout_im_buf);
+
+        descriptor_t descriptor_back{ size };
+
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                                  oneapi::mkl::dft::config_value::INPLACE);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                                  oneapi::mkl::dft::config_value::REAL_REAL);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+        commit_descriptor(descriptor_back, sycl_queue);
+
+        oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                           PrecisionType>(descriptor_back, inout_re_buf,
+                                                          inout_im_buf);
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    /* Once implementations exist, results will need to be verified */
+    EXPECT_TRUE(false);
+
+    return !::testing::Test::HasFailure();
+}
+
+#endif //ONEMKL_COMPUTE_INPLACE_REAL_REAL_HPP
diff --git a/tests/unit_tests/dft/include/compute_out_of_place.hpp b/tests/unit_tests/dft/include/compute_out_of_place.hpp
new file mode 100644
index 000000000..2e9f005ec
--- /dev/null
+++ b/tests/unit_tests/dft/include/compute_out_of_place.hpp
@@ -0,0 +1,140 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_COMPUTE_OUT_OF_PLACE_HPP
+#define ONEMKL_COMPUTE_OUT_OF_PLACE_HPP
+
+#include "compute_tester.hpp"
+
+/* Note: There is no implementation for Domain Real */
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_out_of_place_buffer() {
+    if (!init(MemoryAccessModel::buffer)) {
+        return test_skipped;
+    }
+
+    const size_t bwd_size = domain == oneapi::mkl::dft::domain::REAL ? (size / 2) + 1 : size;
+
+    descriptor_t descriptor{ size };
+    descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                         oneapi::mkl::dft::config_value::NOT_INPLACE);
+    commit_descriptor(descriptor, sycl_queue);
+
+    descriptor_t descriptor_back{ size };
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                              oneapi::mkl::dft::config_value::NOT_INPLACE);
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+    commit_descriptor(descriptor_back, sycl_queue);
+
+    std::vector<FwdInputType> fwd_data(input);
+    std::vector<FwdOutputType> bwd_data(bwd_size, 0);
+
+    {
+        sycl::buffer<FwdInputType, 1> fwd_buf{ fwd_data };
+        sycl::buffer<FwdOutputType, 1> bwd_buf{ bwd_data };
+
+        try {
+            oneapi::mkl::dft::compute_forward<descriptor_t, FwdInputType, FwdOutputType>(
+                descriptor, fwd_buf, bwd_buf);
+        }
+        catch (oneapi::mkl::unimplemented &e) {
+            std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+            return test_skipped;
+        }
+
+        {
+            auto acc_bwd = bwd_buf.template get_host_access();
+            EXPECT_TRUE(check_equal_vector(acc_bwd.get_pointer(), out_host_ref.data(),
+                                           bwd_data.size(), abs_error_margin, rel_error_margin,
+                                           std::cout));
+        }
+
+        try {
+            oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                               FwdOutputType, FwdInputType>(descriptor_back,
+                                                                            bwd_buf, fwd_buf);
+        }
+        catch (oneapi::mkl::unimplemented &e) {
+            std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+            return test_skipped;
+        }
+    }
+
+    EXPECT_TRUE(check_equal_vector(fwd_data.data(), input.data(), input.size(), abs_error_margin,
+                                   rel_error_margin, std::cout));
+    return !::testing::Test::HasFailure();
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_out_of_place_USM() {
+    if (!init(MemoryAccessModel::usm)) {
+        return test_skipped;
+    }
+    const std::vector<sycl::event> no_dependencies;
+
+    const size_t bwd_size = domain == oneapi::mkl::dft::domain::REAL ? (size / 2) + 1 : size;
+
+    descriptor_t descriptor{ size };
+    descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                         oneapi::mkl::dft::config_value::NOT_INPLACE);
+    commit_descriptor(descriptor, sycl_queue);
+
+    descriptor_t descriptor_back{ size };
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                              oneapi::mkl::dft::config_value::NOT_INPLACE);
+    descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+    commit_descriptor(descriptor_back, sycl_queue);
+
+    auto ua_input = usm_allocator_t<FwdInputType>(cxt, *dev);
+    auto ua_output = usm_allocator_t<FwdOutputType>(cxt, *dev);
+
+    std::vector<FwdInputType, decltype(ua_input)> fwd(input.begin(), input.end(), ua_input);
+    std::vector<FwdOutputType, decltype(ua_output)> bwd(bwd_size, ua_output);
+
+    try {
+        oneapi::mkl::dft::compute_forward<descriptor_t, FwdInputType, FwdOutputType>(
+            descriptor, fwd.data(), bwd.data(), no_dependencies)
+            .wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    EXPECT_TRUE(check_equal_vector(bwd.data(), out_host_ref.data(), bwd.size(), abs_error_margin,
+                                   rel_error_margin, std::cout));
+
+    try {
+        oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                           FwdOutputType, FwdInputType>(descriptor_back, bwd.data(),
+                                                                        fwd.data(), no_dependencies)
+            .wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    EXPECT_TRUE(check_equal_vector(fwd.data(), input.data(), input.size(), abs_error_margin,
+                                   rel_error_margin, std::cout));
+
+    return !::testing::Test::HasFailure();
+}
+
+#endif //ONEMKL_COMPUTE_OUT_OF_PLACE_HPP
diff --git a/tests/unit_tests/dft/include/compute_out_of_place_real_real.hpp b/tests/unit_tests/dft/include/compute_out_of_place_real_real.hpp
new file mode 100644
index 000000000..e3148cdb9
--- /dev/null
+++ b/tests/unit_tests/dft/include/compute_out_of_place_real_real.hpp
@@ -0,0 +1,139 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_COMPUTE_OUT_OF_PLACE_REAL_REAL_HPP
+#define ONEMKL_COMPUTE_OUT_OF_PLACE_REAL_REAL_HPP
+
+#include "compute_tester.hpp"
+
+/* Test is not implemented because currently there are no available dft implementations.
+ * These are stubs to make sure that dft::oneapi::mkl::unimplemented exception is thrown */
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_out_of_place_real_real_USM() {
+    if (!init(MemoryAccessModel::usm)) {
+        return test_skipped;
+    }
+
+    try {
+        descriptor_t descriptor{ size };
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::NOT_INPLACE);
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+        commit_descriptor(descriptor, sycl_queue);
+
+        auto ua_input = usm_allocator_t<PrecisionType>(cxt, *dev);
+        auto ua_output = usm_allocator_t<PrecisionType>(cxt, *dev);
+
+        std::vector<PrecisionType, decltype(ua_input)> in_re(size, ua_input);
+        std::vector<PrecisionType, decltype(ua_input)> in_im(size, ua_input);
+        std::vector<PrecisionType, decltype(ua_output)> out_re(size, ua_output);
+        std::vector<PrecisionType, decltype(ua_output)> out_im(size, ua_output);
+        std::vector<PrecisionType, decltype(ua_input)> out_back_re(size, ua_input);
+        std::vector<PrecisionType, decltype(ua_input)> out_back_im(size, ua_input);
+
+        std::copy(input_re.begin(), input_re.end(), in_re.begin());
+        std::copy(input_im.begin(), input_im.end(), in_im.begin());
+
+        std::vector<sycl::event> dependencies;
+        sycl::event done =
+            oneapi::mkl::dft::compute_forward<descriptor_t, PrecisionType, PrecisionType>(
+                descriptor, in_re.data(), in_im.data(), out_re.data(), out_im.data(), dependencies);
+        done.wait();
+
+        descriptor_t descriptor_back{ size };
+
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                                  oneapi::mkl::dft::config_value::NOT_INPLACE);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                                  oneapi::mkl::dft::config_value::REAL_REAL);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+        commit_descriptor(descriptor_back, sycl_queue);
+
+        done =
+            oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                               PrecisionType, PrecisionType>(
+                descriptor_back, out_re.data(), out_im.data(), out_back_re.data(),
+                out_back_im.data());
+        done.wait();
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    /* Once implementations exist, results will need to be verified */
+    EXPECT_TRUE(false);
+
+    return !::testing::Test::HasFailure();
+}
+
+/* Test is not implemented because currently there are no available dft implementations.
+ * These are stubs to make sure that dft::oneapi::mkl::unimplemented exception is thrown */
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int DFT_Test<precision, domain>::test_out_of_place_real_real_buffer() {
+    if (!init(MemoryAccessModel::buffer)) {
+        return test_skipped;
+    }
+
+    try {
+        descriptor_t descriptor{ size };
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::NOT_INPLACE);
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+        commit_descriptor(descriptor, sycl_queue);
+
+        sycl::buffer<PrecisionType, 1> in_dev_re{ input_re.data(), sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> in_dev_im{ input_im.data(), sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> out_dev_re{ sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> out_dev_im{ sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> out_back_dev_re{ sycl::range<1>(size) };
+        sycl::buffer<PrecisionType, 1> out_back_dev_im{ sycl::range<1>(size) };
+
+        oneapi::mkl::dft::compute_forward<descriptor_t, PrecisionType, PrecisionType>(
+            descriptor, in_dev_re, in_dev_im, out_dev_re, out_dev_im);
+
+        descriptor_t descriptor_back{ size };
+
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                                  oneapi::mkl::dft::config_value::NOT_INPLACE);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                                  oneapi::mkl::dft::config_value::REAL_REAL);
+        descriptor_back.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE, (1.0 / size));
+        commit_descriptor(descriptor_back, sycl_queue);
+
+        oneapi::mkl::dft::compute_backward<std::remove_reference_t<decltype(descriptor_back)>,
+                                           PrecisionType, PrecisionType>(
+            descriptor_back, out_dev_re, out_dev_im, out_back_dev_re, out_back_dev_im);
+    }
+    catch (oneapi::mkl::unimplemented &e) {
+        std::cout << "Skipping test because: \"" << e.what() << "\"" << std::endl;
+        return test_skipped;
+    }
+
+    /* Once implementations exist, results will need to be verified */
+    EXPECT_TRUE(false);
+
+    return !::testing::Test::HasFailure();
+}
+
+#endif //ONEMKL_COMPUTE_OUT_OF_PLACE_REAL_REAL_HPP
diff --git a/tests/unit_tests/dft/include/compute_tester.hpp b/tests/unit_tests/dft/include/compute_tester.hpp
new file mode 100644
index 000000000..3e3f293e5
--- /dev/null
+++ b/tests/unit_tests/dft/include/compute_tester.hpp
@@ -0,0 +1,134 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_COMPUTE_TESTER_HPP
+#define ONEMKL_COMPUTE_TESTER_HPP
+
+#include <algorithm>
+
+#if __has_include(<sycl/sycl.hpp>)
+#include <sycl/sycl.hpp>
+#else
+#include <CL/sycl.hpp>
+#endif
+#include "oneapi/mkl.hpp"
+#include "test_helper.hpp"
+#include "test_common.hpp"
+#include "reference_dft.hpp"
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+struct DFT_Test {
+    using descriptor_t = oneapi::mkl::dft::descriptor<precision, domain>;
+
+    template <typename ElemT>
+    using usm_allocator_t = sycl::usm_allocator<ElemT, sycl::usm::alloc::shared, 64>;
+
+    using PrecisionType =
+        typename std::conditional_t<precision == oneapi::mkl::dft::precision::SINGLE, float,
+                                    double>;
+
+    using FwdInputType = typename std::conditional_t<domain == oneapi::mkl::dft::domain::REAL,
+                                                     PrecisionType, std::complex<PrecisionType>>;
+    using FwdOutputType = std::complex<PrecisionType>;
+
+    enum class MemoryAccessModel { buffer, usm };
+
+    const std::int64_t size;
+    const std::int64_t conjugate_even_size;
+    double abs_error_margin;
+    double rel_error_margin;
+
+    sycl::device *dev;
+    sycl::queue sycl_queue;
+    sycl::context cxt;
+
+    std::vector<FwdInputType> input;
+    std::vector<PrecisionType> input_re;
+    std::vector<PrecisionType> input_im;
+    std::vector<FwdOutputType> out_host_ref;
+
+    DFT_Test(sycl::device *dev, std::int64_t size)
+            : size{ static_cast<std::int64_t>(size) },
+              conjugate_even_size{ 2 * (size / 2 + 1) },
+              abs_error_margin{0},
+              rel_error_margin{0},
+              dev{ dev },
+              sycl_queue{ *dev, exception_handler },
+              cxt{ sycl_queue.get_context() } {
+        input = std::vector<FwdInputType>(size);
+        input_re = std::vector<PrecisionType>(size);
+        input_im = std::vector<PrecisionType>(size);
+
+        // out_host_ref contains redundant information for domain::REAL
+        // tests. This simplifies the test implementation, but increases
+        // storage and computational requirements. There is scope for
+        // improvement here if test performance becomes an issue.
+        out_host_ref = std::vector<FwdOutputType>(size);
+        rand_vector(input, size);
+
+        if constexpr (domain == oneapi::mkl::dft::domain::REAL) {
+            for (int i = 0; i < input.size(); ++i) {
+                input_re[i] = { input[i] };
+                input_im[i] = 0;
+            }
+        }
+        else {
+            for (int i = 0; i < input.size(); ++i) {
+                input_re[i] = { input[i].real() };
+                input_im[i] = { input[i].imag() };
+            }
+        }
+    }
+
+    bool skip_test(MemoryAccessModel type) {
+        if constexpr (precision == oneapi::mkl::dft::precision::DOUBLE) {
+            if (!sycl_queue.get_device().has(sycl::aspect::fp64)) {
+                std::cout << "Device does not support double precision." << std::endl;
+                return true;
+            }
+        }
+        if (type == MemoryAccessModel::usm &&
+            !sycl_queue.get_device().has(sycl::aspect::usm_shared_allocations)) {
+            std::cout << "Device does not support usm shared allocations." << std::endl;
+            return true;
+        }
+        return false;
+    }
+
+    bool init(MemoryAccessModel type) {
+        reference_forward_dft<FwdInputType, FwdOutputType>(input, out_host_ref);
+        auto max_norm_ref = *std::max_element(std::begin(out_host_ref), std::end(out_host_ref),
+          [](const FwdOutputType& a, const FwdOutputType& b) { return std::abs(a) < std::abs(b); });
+        // Heuristic for the average-case error margins
+        abs_error_margin = std::abs(max_norm_ref) * std::log2((double)size);
+        rel_error_margin = 5.0 * std::log2((double)size);
+        return !skip_test(type);
+    }
+
+    int test_in_place_buffer();
+    int test_in_place_real_real_buffer();
+    int test_out_of_place_buffer();
+    int test_out_of_place_real_real_buffer();
+    int test_in_place_USM();
+    int test_in_place_real_real_USM();
+    int test_out_of_place_USM();
+    int test_out_of_place_real_real_USM();
+};
+
+#endif //ONEMKL_COMPUTE_TESTER_HPP
diff --git a/tests/unit_tests/dft/include/reference_dft.hpp b/tests/unit_tests/dft/include/reference_dft.hpp
new file mode 100644
index 000000000..232696241
--- /dev/null
+++ b/tests/unit_tests/dft/include/reference_dft.hpp
@@ -0,0 +1,67 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_REFERENCE_DFT_HPP
+#define ONEMKL_REFERENCE_DFT_HPP
+
+#include <cmath>
+#include <complex>
+#include <vector>
+
+#include "test_common.hpp"
+
+/** Naive DFT implementation for reference.
+ *  
+ * Directly compute a single 1D forward DFT of the form:
+ * for k in range(0, N):
+ *   out[k] = sum( exp(2 pi k n im / N) * in[n] for n in range(0, N) )
+ * where N is the size of the input / output arrays. The input may be
+ * real or complex, but the output must be complex. Unit strides are used
+ * with no offset.
+ * 
+ * @tparam TypeIn The forward data type.
+ * @tparam TypeOut The transformed (backward) data type. Written to. Must be 
+ * complex.
+ * @param in The input forward data.
+ * @param out Where to write the output data.
+**/
+template <typename TypeIn, typename TypeOut>
+void reference_forward_dft(const std::vector<TypeIn> &in, std::vector<TypeOut> &out) {
+    if (in.size() != out.size()) {
+        throw std::invalid_argument("Input and output vectors must be of equal size.");
+    }
+    using ref_t = long double; /* Do the calculations using long double */
+    static_assert(is_complex<TypeOut>(), "Output type of DFT must be complex");
+
+    const ref_t TWOPI = 2.0L * 3.141592653589793238462643383279502884197L;
+
+    const size_t N = out.size();
+    for (std::size_t k = 0; k < N; ++k) {
+        std::complex<ref_t> out_temp = 0;
+        const auto partial_expo = (static_cast<ref_t>(k) * TWOPI) / static_cast<ref_t>(N);
+        for (std::size_t n = 0; n < N; ++n) {
+            const auto expo = static_cast<ref_t>(n) * partial_expo;
+            out_temp += static_cast<std::complex<ref_t>>(in[n]) *
+                        std::complex<ref_t>{ std::cos(expo), -std::sin(expo) };
+        }
+        out[k] = static_cast<TypeOut>(out_temp);
+    }
+}
+
+#endif //ONEMKL_REFERENCE_DFT_HPP
diff --git a/tests/unit_tests/dft/include/test_common.hpp b/tests/unit_tests/dft/include/test_common.hpp
new file mode 100644
index 000000000..c0474148b
--- /dev/null
+++ b/tests/unit_tests/dft/include/test_common.hpp
@@ -0,0 +1,167 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#ifndef ONEMKL_TEST_COMMON_HPP
+#define ONEMKL_TEST_COMMON_HPP
+
+#include <cstdint>
+#include <cstdlib>
+#include <iostream>
+#include <limits>
+#include <vector>
+
+#if __has_include(<sycl/sycl.hpp>)
+#include <sycl/sycl.hpp>
+#else
+#include <CL/sycl.hpp>
+#endif
+
+template <typename T>
+struct complex_info {
+    using real_type = T;
+    static const bool is_complex = false;
+};
+
+template <typename T>
+struct complex_info<std::complex<T>> {
+    using real_type = T;
+    static const bool is_complex = true;
+};
+
+template <typename T>
+constexpr bool is_complex() {
+    return complex_info<T>::is_complex;
+}
+
+template <typename fp>
+bool check_equal(fp x, fp x_ref, double abs_error_mag, double rel_error_mag, std::ostream &out) {
+    using fp_real = typename complex_info<fp>::real_type;
+    static_assert(std::is_floating_point_v<fp_real>,
+                  "Expected floating-point real or complex type.");
+
+    const fp_real epsilon = []() {
+        if constexpr (sizeof(double) == sizeof(long double) && std::is_same_v<fp_real, double>) {
+            // The reference DFT uses long double to maintain accuracy
+            // when this isn't possible, lower the accuracy requirements
+            return 1e-12;
+        }
+        else {
+            return std::numeric_limits<fp_real>::epsilon();
+        }
+    }();
+    const fp_real abs_bound = abs_error_mag * epsilon;
+    const fp_real rel_bound = rel_error_mag * epsilon;
+
+    const auto aerr = std::abs(x - x_ref);
+    const auto rerr = aerr / std::abs(x_ref);
+    const bool ok = (rerr <= rel_bound) || (aerr <= abs_bound);
+    if (!ok) {
+        out << "Mismatching results: actual = " << x << " vs. reference = " << x_ref << "\n";
+        out << " relative error = " << rerr
+            << " absolute error = " << aerr
+            << " relative bound = " << rel_bound
+            << " absolute bound = " << abs_bound
+            << "\n";
+    }
+    return ok;
+}
+
+template <typename vec1, typename vec2>
+bool check_equal_vector(vec1 &&v, vec2 &&v_ref, int n, double abs_error_mag, double rel_error_mag, std::ostream &out) {
+    constexpr int max_print = 20;
+    int count = 0;
+    bool good = true;
+
+    for (std::size_t i = 0; i < n; ++i) {
+        if (!check_equal(v[i], v_ref[i], abs_error_mag, rel_error_mag, out)) {
+            out << " at index i =" << i << "\n";
+            good = false;
+            ++count;
+            if (count > max_print) {
+                return good;
+            }
+        }
+    }
+
+    return good;
+}
+
+// Random initialization.
+template <typename t>
+inline t rand_scalar() {
+    if constexpr (std::is_same_v<t, int32_t>) {
+        return std::rand() % 256 - 128;
+    }
+    else if constexpr (std::is_floating_point_v<t>) {
+        return t(std::rand()) / t(RAND_MAX) - t(0.5);
+    }
+    else {
+        static_assert(complex_info<t>::is_complex, "unexpect type in rand_scalar");
+        using fp = typename complex_info<t>::real_type;
+        return t(rand_scalar<fp>(), rand_scalar<fp>());
+    }
+}
+
+template <typename vec>
+void rand_vector(vec &v, int n) {
+    using fp = typename vec::value_type;
+    v.resize(n);
+    for (int i = 0; i < n; i++) {
+        v[i] = rand_scalar<fp>();
+    }
+}
+
+// Catch asynchronous exceptions.
+auto exception_handler = [](sycl::exception_list exceptions) {
+    for (std::exception_ptr const &e : exceptions) {
+        try {
+            std::rethrow_exception(e);
+        }
+        catch (sycl::exception e) {
+            std::cout << "Caught asynchronous SYCL exception:\n" << e.what() << "\n";
+            print_error_code(e);
+        }
+    }
+};
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+void commit_descriptor(oneapi::mkl::dft::descriptor<precision, domain> &descriptor,
+                       sycl::queue queue) {
+#ifdef CALL_RT_API
+    descriptor.commit(queue);
+#else
+    TEST_RUN_CT_SELECT_NO_ARGS(queue, descriptor.commit);
+#endif
+}
+
+class DimensionsDeviceNamePrint {
+public:
+    std::string operator()(
+        testing::TestParamInfo<std::tuple<sycl::device *, std::int64_t>> dev) const {
+        std::string size = "size_" + std::to_string(std::get<1>(dev.param));
+        std::string dev_name = std::get<0>(dev.param)->get_info<sycl::info::device::name>();
+        for (std::string::size_type i = 0; i < dev_name.size(); ++i) {
+            if (!isalnum(dev_name[i]))
+                dev_name[i] = '_';
+        }
+        return size.append("_").append(dev_name);
+    }
+};
+
+#endif //ONEMKL_TEST_COMMON_HPP
diff --git a/tests/unit_tests/dft/source/CMakeLists.txt b/tests/unit_tests/dft/source/CMakeLists.txt
new file mode 100644
index 000000000..82d5ab72b
--- /dev/null
+++ b/tests/unit_tests/dft/source/CMakeLists.txt
@@ -0,0 +1,59 @@
+#===============================================================================
+# Copyright 2022 Intel Corporation
+#
+# Licensed under the Apache License, Version 2.0 (the "License");
+# you may not use this file except in compliance with the License.
+# You may obtain a copy of the License at
+#
+# http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing,
+# software distributed under the License is distributed on an "AS IS" BASIS,
+# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+# See the License for the specific language governing permissions
+# and limitations under the License.
+#
+#
+# SPDX-License-Identifier: Apache-2.0
+#===============================================================================
+
+set(DFT_SOURCES "compute_tests.cpp" "descriptor_tests.cpp")
+
+if (BUILD_SHARED_LIBS)
+    add_library(dft_source_rt OBJECT ${DFT_SOURCES})
+    target_compile_options(dft_source_rt PRIVATE -DCALL_RT_API -DNOMINMAX)
+    target_include_directories(dft_source_rt
+            PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../include
+            PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../../include
+            PUBLIC ${PROJECT_SOURCE_DIR}/include
+            PUBLIC ${PROJECT_SOURCE_DIR}/deps/googletest/include
+            PUBLIC ${CMAKE_BINARY_DIR}/bin
+            PUBLIC ${CBLAS_INCLUDE}
+            )
+    if (USE_ADD_SYCL_TO_TARGET_INTEGRATION)
+        add_sycl_to_target(TARGET dft_source_rt SOURCES ${DFT_SOURCES})
+    else ()
+        target_link_libraries(dft_source_rt PUBLIC ONEMKL::SYCL::SYCL)
+    endif ()
+endif ()
+
+add_library(dft_source_ct OBJECT ${DFT_SOURCES})
+target_compile_options(dft_source_ct PRIVATE -DNOMINMAX)
+target_include_directories(dft_source_ct
+        PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../include
+        PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/../../include
+        PUBLIC ${PROJECT_SOURCE_DIR}/include
+        PUBLIC ${PROJECT_SOURCE_DIR}/deps/googletest/include
+        PUBLIC ${CMAKE_BINARY_DIR}/bin
+        PUBLIC ${CBLAS_INCLUDE}
+        )
+if (USE_ADD_SYCL_TO_TARGET_INTEGRATION)
+    add_sycl_to_target(TARGET dft_source_ct SOURCES ${DFT_SOURCES})
+    target_link_libraries(dft_source_ct PUBLIC onemkl)
+else ()
+    target_link_libraries(dft_source_ct PUBLIC
+            onemkl
+            ONEMKL::SYCL::SYCL
+            )
+endif ()
+
diff --git a/tests/unit_tests/dft/source/compute_tests.cpp b/tests/unit_tests/dft/source/compute_tests.cpp
new file mode 100644
index 000000000..dee9333d3
--- /dev/null
+++ b/tests/unit_tests/dft/source/compute_tests.cpp
@@ -0,0 +1,282 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#include <iostream>
+#include <vector>
+
+#if __has_include(<sycl/sycl.hpp>)
+#include <sycl/sycl.hpp>
+#else
+#include <CL/sycl.hpp>
+#endif
+
+#include "test_helper.hpp"
+#include "test_common.hpp"
+#include <gtest/gtest.h>
+
+#include "compute_inplace.hpp"
+#include "compute_inplace_real_real.hpp"
+#include "compute_out_of_place.hpp"
+#include "compute_out_of_place_real_real.hpp"
+
+extern std::vector<sycl::device *> devices;
+
+namespace {
+
+class ComputeTests : public ::testing::TestWithParam<std::tuple<sycl::device *, std::int64_t>> {};
+
+std::vector<std::int64_t> lengths{ 8, 21, 128 };
+
+/* test_in_place_buffer() */
+TEST_P(ComputeTests, RealSinglePrecisionInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_buffer());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_buffer());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_buffer());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_buffer());
+}
+
+/* test_in_place_real_real_buffer() */
+TEST_P(ComputeTests, RealSinglePrecisionInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_buffer());
+}
+
+/* test_out_of_place_buffer() */
+TEST_P(ComputeTests, RealSinglePrecisionNotInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_buffer());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionNotInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_buffer());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionNotInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_buffer());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionNotInPlaceBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_buffer());
+}
+
+/* test_out_of_place_real_real_buffer */
+TEST_P(ComputeTests, RealSinglePrecisionNotInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionNotInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionNotInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_buffer());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionNotInPlaceRealRealBuffer) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_buffer());
+}
+
+/* test_in_place_USM */
+TEST_P(ComputeTests, RealSinglePrecisionInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_USM());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_USM());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_USM());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_USM());
+}
+
+/* test_in_place_real_real_USM */
+TEST_P(ComputeTests, RealSinglePrecisionInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_in_place_real_real_USM());
+}
+
+/* test_out_of_place_USM */
+TEST_P(ComputeTests, RealSinglePrecisionNotInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_USM());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionNotInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_USM());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionNotInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_USM());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionNotInPlaceUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_USM());
+}
+
+/* test_out_of_place_real_real_USM */
+TEST_P(ComputeTests, RealSinglePrecisionNotInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, RealDoublePrecisionNotInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, ComplexSinglePrecisionNotInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_USM());
+}
+
+TEST_P(ComputeTests, ComplexDoublePrecisionNotInPlaceRealRealUSM) {
+    auto test = DFT_Test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>{
+        std::get<0>(GetParam()), std::get<1>(GetParam())
+    };
+    EXPECT_TRUEORSKIP(test.test_out_of_place_real_real_USM());
+}
+
+INSTANTIATE_TEST_SUITE_P(ComputeTestSuite, ComputeTests,
+                         ::testing::Combine(testing::ValuesIn(devices), testing::ValuesIn(lengths)),
+                         ::DimensionsDeviceNamePrint());
+
+} // anonymous namespace
diff --git a/tests/unit_tests/dft/source/descriptor_tests.cpp b/tests/unit_tests/dft/source/descriptor_tests.cpp
new file mode 100644
index 000000000..7bd955b2c
--- /dev/null
+++ b/tests/unit_tests/dft/source/descriptor_tests.cpp
@@ -0,0 +1,452 @@
+/***************************************************************************
+*  Copyright (C) Codeplay Software Limited
+*  Licensed under the Apache License, Version 2.0 (the "License");
+*  you may not use this file except in compliance with the License.
+*  You may obtain a copy of the License at
+*
+*      http://www.apache.org/licenses/LICENSE-2.0
+*
+*  For your convenience, a copy of the License has been included in this
+*  repository.
+*
+*  Unless required by applicable law or agreed to in writing, software
+*  distributed under the License is distributed on an "AS IS" BASIS,
+*  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+*  See the License for the specific language governing permissions and
+*  limitations under the License.
+*
+**************************************************************************/
+
+#include <iostream>
+#include <vector>
+
+#if __has_include(<sycl/sycl.hpp>)
+#include <sycl/sycl.hpp>
+#else
+#include <CL/sycl.hpp>
+#endif
+
+#include "test_helper.hpp"
+#include "test_common.hpp"
+#include <gtest/gtest.h>
+
+extern std::vector<sycl::device*> devices;
+
+namespace {
+
+constexpr std::int64_t default_1d_lengths = 4;
+const std::vector<std::int64_t> default_3d_lengths{ 124, 5, 3 };
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+inline void set_and_get_lengths(sycl::queue& sycl_queue) {
+    /* Negative Testing */
+    {
+        oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_3d_lengths };
+        EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::LENGTHS, nullptr),
+                     oneapi::mkl::invalid_argument);
+    }
+
+    /* 1D */
+    {
+        oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_1d_lengths };
+
+        std::int64_t lengths_value{ 0 };
+        std::int64_t new_lengths{ 2345 };
+        std::int64_t dimensions_before_set{ 0 };
+        std::int64_t dimensions_after_set{ 0 };
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::LENGTHS, &lengths_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimensions_before_set);
+        EXPECT_EQ(default_1d_lengths, lengths_value);
+        EXPECT_EQ(dimensions_before_set, 1);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::LENGTHS, new_lengths);
+        descriptor.get_value(oneapi::mkl::dft::config_param::LENGTHS, &lengths_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimensions_after_set);
+        EXPECT_EQ(new_lengths, lengths_value);
+        EXPECT_EQ(dimensions_before_set, dimensions_after_set);
+
+        commit_descriptor(descriptor, sycl_queue);
+    }
+
+    /* >= 2D */
+    {
+        const std::int64_t dimensions = 3;
+
+        oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_3d_lengths };
+
+        std::vector<std::int64_t> lengths_value(3);
+        std::vector<std::int64_t> new_lengths{ 1, 2, 7 };
+        std::int64_t dimensions_before_set{ 0 };
+        std::int64_t dimensions_after_set{ 0 };
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::LENGTHS, lengths_value.data());
+        descriptor.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimensions_before_set);
+
+        EXPECT_EQ(default_3d_lengths, lengths_value);
+        EXPECT_EQ(dimensions, dimensions_before_set);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::LENGTHS, new_lengths.data());
+        descriptor.get_value(oneapi::mkl::dft::config_param::LENGTHS, lengths_value.data());
+        descriptor.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimensions_after_set);
+
+        EXPECT_EQ(new_lengths, lengths_value);
+        EXPECT_EQ(dimensions_before_set, dimensions_after_set);
+    }
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+inline void set_and_get_strides(sycl::queue& sycl_queue) {
+    oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_3d_lengths };
+
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::INPUT_STRIDES, nullptr),
+                 oneapi::mkl::invalid_argument);
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::OUTPUT_STRIDES, nullptr),
+                 oneapi::mkl::invalid_argument);
+
+    constexpr std::int64_t strides_size = 4;
+    const std::int64_t default_stride_d1 = default_3d_lengths[2] * default_3d_lengths[1];
+    const std::int64_t default_stride_d2 = default_3d_lengths[2];
+    const std::int64_t default_stride_d3 = 1;
+
+    std::vector<std::int64_t> default_strides_value{ 0, default_stride_d1, default_stride_d2,
+                                                     default_stride_d3 };
+
+    std::vector<std::int64_t> input_strides_value;
+    std::vector<std::int64_t> output_strides_value;
+    if constexpr (domain == oneapi::mkl::dft::domain::COMPLEX) {
+        input_strides_value = { 50, default_stride_d1 * 2, default_stride_d2 * 2,
+                                default_stride_d3 * 2 };
+        output_strides_value = { 50, default_stride_d1 * 2, default_stride_d2 * 2,
+                                 default_stride_d3 * 2 };
+    }
+    else {
+        input_strides_value = { 0, default_3d_lengths[1] * (default_3d_lengths[2] / 2 + 1) * 2,
+                                (default_3d_lengths[2] / 2 + 1) * 2, 1 };
+        output_strides_value = { 0, default_3d_lengths[1] * (default_3d_lengths[2] / 2 + 1),
+                                 (default_3d_lengths[2] / 2 + 1), 1 };
+    }
+
+    std::vector<std::int64_t> input_strides_before_set(strides_size);
+    std::vector<std::int64_t> input_strides_after_set(strides_size);
+
+    descriptor.get_value(oneapi::mkl::dft::config_param::INPUT_STRIDES,
+                         input_strides_before_set.data());
+    EXPECT_EQ(default_strides_value, input_strides_before_set);
+    descriptor.set_value(oneapi::mkl::dft::config_param::INPUT_STRIDES, input_strides_value.data());
+    descriptor.get_value(oneapi::mkl::dft::config_param::INPUT_STRIDES,
+                         input_strides_after_set.data());
+    EXPECT_EQ(input_strides_value, input_strides_after_set);
+
+    std::vector<std::int64_t> output_strides_before_set(strides_size);
+    std::vector<std::int64_t> output_strides_after_set(strides_size);
+    descriptor.get_value(oneapi::mkl::dft::config_param::OUTPUT_STRIDES,
+                         output_strides_before_set.data());
+    EXPECT_EQ(default_strides_value, output_strides_before_set);
+    descriptor.set_value(oneapi::mkl::dft::config_param::OUTPUT_STRIDES,
+                         output_strides_value.data());
+    descriptor.get_value(oneapi::mkl::dft::config_param::OUTPUT_STRIDES,
+                         output_strides_after_set.data());
+    EXPECT_EQ(output_strides_value, output_strides_after_set);
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+inline void set_and_get_values(sycl::queue& sycl_queue) {
+    oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_1d_lengths };
+
+    using Precision_Type =
+        typename std::conditional_t<precision == oneapi::mkl::dft::precision::SINGLE, float,
+                                    double>;
+
+    {
+        Precision_Type forward_scale_set_value{ 143.5 };
+        Precision_Type forward_scale_before_set;
+        Precision_Type forward_scale_after_set;
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::FORWARD_SCALE,
+                             &forward_scale_before_set);
+        EXPECT_EQ(1.0, forward_scale_before_set);
+        descriptor.set_value(oneapi::mkl::dft::config_param::FORWARD_SCALE,
+                             forward_scale_set_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::FORWARD_SCALE,
+                             &forward_scale_after_set);
+        EXPECT_EQ(forward_scale_set_value, forward_scale_after_set);
+    }
+
+    {
+        Precision_Type backward_scale_set_value{ 143.5 };
+        Precision_Type backward_scale_before_set;
+        Precision_Type backward_scale_after_set;
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE,
+                             &backward_scale_before_set);
+        EXPECT_EQ(1.0, backward_scale_before_set);
+        descriptor.set_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE,
+                             backward_scale_set_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::BACKWARD_SCALE,
+                             &backward_scale_after_set);
+        EXPECT_EQ(backward_scale_set_value, backward_scale_after_set);
+    }
+
+    {
+        std::int64_t n_transforms_set_value{ 12 };
+        std::int64_t n_transforms_before_set;
+        std::int64_t n_transforms_after_set;
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::NUMBER_OF_TRANSFORMS,
+                             &n_transforms_before_set);
+        EXPECT_EQ(1, n_transforms_before_set);
+        descriptor.set_value(oneapi::mkl::dft::config_param::NUMBER_OF_TRANSFORMS,
+                             n_transforms_set_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::NUMBER_OF_TRANSFORMS,
+                             &n_transforms_after_set);
+        EXPECT_EQ(n_transforms_set_value, n_transforms_after_set);
+    }
+
+    {
+        std::int64_t fwd_distance_set_value{ 12 };
+        std::int64_t fwd_distance_before_set;
+        std::int64_t fwd_distance_after_set;
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::FWD_DISTANCE,
+                             &fwd_distance_before_set);
+        EXPECT_EQ(1, fwd_distance_before_set);
+        descriptor.set_value(oneapi::mkl::dft::config_param::FWD_DISTANCE, fwd_distance_set_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::FWD_DISTANCE, &fwd_distance_after_set);
+        EXPECT_EQ(fwd_distance_set_value, fwd_distance_after_set);
+
+        std::int64_t bwd_distance_set_value{ domain == oneapi::mkl::dft::domain::REAL
+                                                 ? (fwd_distance_set_value / 2) + 1
+                                                 : fwd_distance_set_value };
+        std::int64_t bwd_distance_before_set;
+        std::int64_t bwd_distance_after_set;
+
+        descriptor.get_value(oneapi::mkl::dft::config_param::BWD_DISTANCE,
+                             &bwd_distance_before_set);
+        EXPECT_EQ(1, bwd_distance_before_set);
+        descriptor.set_value(oneapi::mkl::dft::config_param::BWD_DISTANCE, bwd_distance_set_value);
+        descriptor.get_value(oneapi::mkl::dft::config_param::BWD_DISTANCE, &bwd_distance_after_set);
+        EXPECT_EQ(bwd_distance_set_value, bwd_distance_after_set);
+    }
+
+    {
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::PLACEMENT, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::INPLACE, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::NOT_INPLACE);
+        descriptor.get_value(oneapi::mkl::dft::config_param::PLACEMENT, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::NOT_INPLACE, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PLACEMENT,
+                             oneapi::mkl::dft::config_value::INPLACE);
+        descriptor.get_value(oneapi::mkl::dft::config_param::PLACEMENT, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::INPLACE, value);
+    }
+
+    {
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::COMPLEX_COMPLEX, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+        descriptor.get_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::REAL_REAL, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE,
+                             oneapi::mkl::dft::config_value::COMPLEX_COMPLEX);
+        descriptor.get_value(oneapi::mkl::dft::config_param::COMPLEX_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::COMPLEX_COMPLEX, value);
+    }
+
+    {
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::CONJUGATE_EVEN_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::COMPLEX_COMPLEX, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::CONJUGATE_EVEN_STORAGE,
+                             oneapi::mkl::dft::config_value::COMPLEX_COMPLEX);
+
+        value = oneapi::mkl::dft::config_value::COMMITTED; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::CONJUGATE_EVEN_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::COMPLEX_COMPLEX, value);
+    }
+
+    {
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::REAL_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::REAL_REAL, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::REAL_STORAGE,
+                             oneapi::mkl::dft::config_value::REAL_REAL);
+
+        value = oneapi::mkl::dft::config_value::COMMITTED; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::REAL_STORAGE, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::REAL_REAL, value);
+    }
+
+    {
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::ORDERING, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::ORDERED, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::ORDERING,
+                             oneapi::mkl::dft::config_value::BACKWARD_SCRAMBLED);
+        descriptor.get_value(oneapi::mkl::dft::config_param::ORDERING, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::BACKWARD_SCRAMBLED, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::ORDERING,
+                             oneapi::mkl::dft::config_value::ORDERED);
+        descriptor.get_value(oneapi::mkl::dft::config_param::ORDERING, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::ORDERED, value);
+    }
+
+    {
+        bool value = true;
+        descriptor.get_value(oneapi::mkl::dft::config_param::TRANSPOSE, &value);
+        EXPECT_EQ(false, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::TRANSPOSE, true);
+        descriptor.get_value(oneapi::mkl::dft::config_param::TRANSPOSE, &value);
+        EXPECT_EQ(true, value);
+        /* Set value to false again because transpose is not implemented and will fail on commit
+         * when using the MKLGPU backend */
+        descriptor.set_value(oneapi::mkl::dft::config_param::TRANSPOSE, false);
+    }
+
+    {
+        /* Only value currently supported for PACKED_FORMAT is the config_value::CCE_FORMAT */
+        oneapi::mkl::dft::config_value value{
+            oneapi::mkl::dft::config_value::COMMITTED
+        }; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::PACKED_FORMAT, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::CCE_FORMAT, value);
+
+        descriptor.set_value(oneapi::mkl::dft::config_param::PACKED_FORMAT,
+                             oneapi::mkl::dft::config_value::CCE_FORMAT);
+
+        value = oneapi::mkl::dft::config_value::COMMITTED; // Initialize with invalid value
+        descriptor.get_value(oneapi::mkl::dft::config_param::PACKED_FORMAT, &value);
+        EXPECT_EQ(oneapi::mkl::dft::config_value::CCE_FORMAT, value);
+    }
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+inline void get_readonly_values(sycl::queue& sycl_queue) {
+    oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_1d_lengths };
+
+    oneapi::mkl::dft::domain domain_value;
+    descriptor.get_value(oneapi::mkl::dft::config_param::FORWARD_DOMAIN, &domain_value);
+    EXPECT_EQ(domain_value, domain);
+
+    oneapi::mkl::dft::precision precision_value;
+    descriptor.get_value(oneapi::mkl::dft::config_param::PRECISION, &precision_value);
+    EXPECT_EQ(precision_value, precision);
+
+    std::int64_t dimension_value;
+    descriptor.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimension_value);
+    EXPECT_EQ(dimension_value, 1);
+
+    oneapi::mkl::dft::descriptor<precision, domain> descriptor3D{ default_3d_lengths };
+    descriptor3D.get_value(oneapi::mkl::dft::config_param::DIMENSION, &dimension_value);
+    EXPECT_EQ(dimension_value, 3);
+
+    oneapi::mkl::dft::config_value commit_status;
+    descriptor.get_value(oneapi::mkl::dft::config_param::COMMIT_STATUS, &commit_status);
+    EXPECT_EQ(commit_status, oneapi::mkl::dft::config_value::UNCOMMITTED);
+
+    commit_descriptor(descriptor, sycl_queue);
+    descriptor.get_value(oneapi::mkl::dft::config_param::COMMIT_STATUS, &commit_status);
+    EXPECT_EQ(commit_status, oneapi::mkl::dft::config_value::COMMITTED);
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+inline void set_readonly_values(sycl::queue& sycl_queue) {
+    oneapi::mkl::dft::descriptor<precision, domain> descriptor{ default_1d_lengths };
+
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::FORWARD_DOMAIN,
+                                      oneapi::mkl::dft::domain::REAL),
+                 oneapi::mkl::invalid_argument);
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::FORWARD_DOMAIN,
+                                      oneapi::mkl::dft::domain::COMPLEX),
+                 oneapi::mkl::invalid_argument);
+
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::PRECISION,
+                                      oneapi::mkl::dft::precision::SINGLE),
+                 oneapi::mkl::invalid_argument);
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::PRECISION,
+                                      oneapi::mkl::dft::precision::DOUBLE),
+                 oneapi::mkl::invalid_argument);
+
+    std::int64_t set_dimension{ 3 };
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::DIMENSION, set_dimension),
+                 oneapi::mkl::invalid_argument);
+
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::COMMIT_STATUS,
+                                      oneapi::mkl::dft::config_value::COMMITTED),
+                 oneapi::mkl::invalid_argument);
+    EXPECT_THROW(descriptor.set_value(oneapi::mkl::dft::config_param::COMMIT_STATUS,
+                                      oneapi::mkl::dft::config_value::UNCOMMITTED),
+                 oneapi::mkl::invalid_argument);
+
+    commit_descriptor(descriptor, sycl_queue);
+}
+
+template <oneapi::mkl::dft::precision precision, oneapi::mkl::dft::domain domain>
+int test(sycl::device* dev) {
+    sycl::queue sycl_queue(*dev, exception_handler);
+
+    if constexpr (precision == oneapi::mkl::dft::precision::DOUBLE) {
+        if (!sycl_queue.get_device().has(sycl::aspect::fp64)) {
+            std::cout << "Device does not support double precision." << std::endl;
+            return test_skipped;
+        }
+    }
+
+    set_and_get_lengths<precision, domain>(sycl_queue);
+    set_and_get_strides<precision, domain>(sycl_queue);
+    set_and_get_values<precision, domain>(sycl_queue);
+    get_readonly_values<precision, domain>(sycl_queue);
+    set_readonly_values<precision, domain>(sycl_queue);
+
+    return !::testing::Test::HasFailure();
+}
+
+class DescriptorTests : public ::testing::TestWithParam<sycl::device*> {};
+
+TEST_P(DescriptorTests, DescriptorTestsRealSingle) {
+    EXPECT_TRUEORSKIP(
+        (test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::REAL>(GetParam())));
+}
+
+TEST_P(DescriptorTests, DescriptorTestsRealDouble) {
+    EXPECT_TRUEORSKIP(
+        (test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::REAL>(GetParam())));
+}
+
+TEST_P(DescriptorTests, DescriptorTestsComplexSingle) {
+    EXPECT_TRUEORSKIP(
+        (test<oneapi::mkl::dft::precision::SINGLE, oneapi::mkl::dft::domain::COMPLEX>(GetParam())));
+}
+
+TEST_P(DescriptorTests, DescriptorTestsComplexDouble) {
+    EXPECT_TRUEORSKIP(
+        (test<oneapi::mkl::dft::precision::DOUBLE, oneapi::mkl::dft::domain::COMPLEX>(GetParam())));
+}
+
+INSTANTIATE_TEST_SUITE_P(DescriptorTestSuite, DescriptorTests, testing::ValuesIn(devices),
+                         ::DeviceNamePrint());
+
+} // anonymous namespace
diff --git a/tests/unit_tests/include/test_helper.hpp b/tests/unit_tests/include/test_helper.hpp
index 56e625af2..a7bd82f6c 100644
--- a/tests/unit_tests/include/test_helper.hpp
+++ b/tests/unit_tests/include/test_helper.hpp
@@ -55,6 +55,8 @@
 
 #if defined(ENABLE_MKLCPU_BACKEND) || defined(ENABLE_NETLIB_BACKEND)
 #ifdef ENABLE_MKLCPU_BACKEND
+#define TEST_RUN_INTELCPU_SELECT_NO_ARGS(q, func) \
+    func(oneapi::mkl::backend_selector<oneapi::mkl::backend::mklcpu>{ q })
 #define TEST_RUN_INTELCPU_SELECT(q, func, ...) \
     func(oneapi::mkl::backend_selector<oneapi::mkl::backend::mklcpu>{ q }, __VA_ARGS__)
 #else
@@ -62,13 +64,17 @@
     func(oneapi::mkl::backend_selector<oneapi::mkl::backend::netlib>{ q }, __VA_ARGS__)
 #endif
 #else
+#define TEST_RUN_INTELCPU_SELECT_NO_ARGS(q, func)
 #define TEST_RUN_INTELCPU_SELECT(q, func, ...)
 #endif
 
 #ifdef ENABLE_MKLGPU_BACKEND
+#define TEST_RUN_INTELGPU_SELECT_NO_ARGS(q, func) \
+    func(oneapi::mkl::backend_selector<oneapi::mkl::backend::mklgpu>{ q })
 #define TEST_RUN_INTELGPU_SELECT(q, func, ...) \
     func(oneapi::mkl::backend_selector<oneapi::mkl::backend::mklgpu>{ q }, __VA_ARGS__)
 #else
+#define TEST_RUN_INTELGPU_SELECT_NO_ARGS(q, func)
 #define TEST_RUN_INTELGPU_SELECT(q, func, ...)
 #endif
 
@@ -119,6 +125,20 @@
 #define CHECK_HOST_OR_CPU(q) q.is_host() || q.get_device().is_cpu()
 #endif
 
+#define TEST_RUN_CT_SELECT_NO_ARGS(q, func)                                \
+    do {                                                                   \
+        if (CHECK_HOST_OR_CPU(q))   {                                      \
+            TEST_RUN_INTELCPU_SELECT_NO_ARGS(q, func);                     \
+        }                                                                  \
+        else if (q.get_device().is_gpu()) {                                \
+            unsigned int vendor_id = static_cast<unsigned int>(            \
+                q.get_device().get_info<sycl::info::device::vendor_id>()); \
+            if (vendor_id == INTEL_ID) {                                   \
+                TEST_RUN_INTELGPU_SELECT_NO_ARGS(q, func);                 \
+            }                                                              \
+        }                                                                  \
+    } while (0);
+
 #define TEST_RUN_CT_SELECT(q, func, ...)                                   \
     do {                                                                   \
         if (CHECK_HOST_OR_CPU(q))                                          \