diff --git a/common/matrix/dense_kernels.hpp.inc b/common/matrix/dense_kernels.hpp.inc
index c26203115fe..a9947ad891f 100644
--- a/common/matrix/dense_kernels.hpp.inc
+++ b/common/matrix/dense_kernels.hpp.inc
@@ -539,4 +539,48 @@ __global__ __launch_bounds__(default_block_size) void outplace_absolute_dense(
 }
 
 
+template <typename ValueType, typename ComplexType>
+__global__ __launch_bounds__(default_block_size) void make_complex(
+    size_type num_rows, size_type num_cols, const ValueType *__restrict__ in,
+    size_type stride_in, ComplexType *__restrict__ out, size_type stride_out)
+{
+    const auto tidx = thread::get_thread_id_flat();
+    auto row = tidx / num_cols;
+    auto col = tidx % num_cols;
+    if (row < num_rows) {
+        out[row * stride_out + col] = in[row * stride_in + col];
+    }
+}
+
+
+template <typename ValueType>
+__global__ __launch_bounds__(default_block_size) void get_real(
+    size_type num_rows, size_type num_cols, const ValueType *__restrict__ in,
+    size_type stride_in, remove_complex<ValueType> *__restrict__ out,
+    size_type stride_out)
+{
+    const auto tidx = thread::get_thread_id_flat();
+    auto row = tidx / num_cols;
+    auto col = tidx % num_cols;
+    if (row < num_rows) {
+        out[row * stride_out + col] = real(in[row * stride_in + col]);
+    }
+}
+
+
+template <typename ValueType>
+__global__ __launch_bounds__(default_block_size) void get_imag(
+    size_type num_rows, size_type num_cols, const ValueType *__restrict__ in,
+    size_type stride_in, remove_complex<ValueType> *__restrict__ out,
+    size_type stride_out)
+{
+    const auto tidx = thread::get_thread_id_flat();
+    auto row = tidx / num_cols;
+    auto col = tidx % num_cols;
+    if (row < num_rows) {
+        out[row * stride_out + col] = imag(in[row * stride_in + col]);
+    }
+}
+
+
 }  // namespace kernel
diff --git a/core/device_hooks/common_kernels.inc.cpp b/core/device_hooks/common_kernels.inc.cpp
index a981372c630..34992284498 100644
--- a/core/device_hooks/common_kernels.inc.cpp
+++ b/core/device_hooks/common_kernels.inc.cpp
@@ -256,6 +256,21 @@ GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL(ValueType)
 GKO_NOT_COMPILED(GKO_HOOK_MODULE);
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
+template <typename ValueType>
+GKO_DECLARE_MAKE_COMPLEX_KERNEL(ValueType)
+GKO_NOT_COMPILED(GKO_HOOK_MODULE);
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+template <typename ValueType>
+GKO_DECLARE_GET_REAL_KERNEL(ValueType)
+GKO_NOT_COMPILED(GKO_HOOK_MODULE);
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+template <typename ValueType>
+GKO_DECLARE_GET_IMAG_KERNEL(ValueType)
+GKO_NOT_COMPILED(GKO_HOOK_MODULE);
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
 
 }  // namespace dense
 
diff --git a/core/matrix/dense.cpp b/core/matrix/dense.cpp
index 2e0b663966f..cb350c31628 100644
--- a/core/matrix/dense.cpp
+++ b/core/matrix/dense.cpp
@@ -88,6 +88,9 @@ GKO_REGISTER_OPERATION(convert_to_sparsity_csr, dense::convert_to_sparsity_csr);
 GKO_REGISTER_OPERATION(extract_diagonal, dense::extract_diagonal);
 GKO_REGISTER_OPERATION(inplace_absolute_dense, dense::inplace_absolute_dense);
 GKO_REGISTER_OPERATION(outplace_absolute_dense, dense::outplace_absolute_dense);
+GKO_REGISTER_OPERATION(make_complex, dense::make_complex);
+GKO_REGISTER_OPERATION(get_real, dense::get_real);
+GKO_REGISTER_OPERATION(get_imag, dense::get_imag);
 
 
 }  // namespace dense
@@ -795,6 +798,84 @@ Dense<ValueType>::compute_absolute() const
 }
 
 
+template <typename ValueType>
+std::unique_ptr<typename Dense<ValueType>::complex_type>
+Dense<ValueType>::make_complex() const
+{
+    auto exec = this->get_executor();
+
+    auto complex_dense = complex_type::create(exec, this->get_size());
+
+    exec->run(dense::make_make_complex(this, complex_dense.get()));
+
+    return complex_dense;
+}
+
+
+template <typename ValueType>
+void Dense<ValueType>::make_complex(Dense<to_complex<ValueType>> *result) const
+{
+    auto exec = this->get_executor();
+
+    GKO_ASSERT_EQUAL_DIMENSIONS(this, result);
+
+    exec->run(dense::make_make_complex(
+        this, make_temporary_clone(exec, result).get()));
+}
+
+
+template <typename ValueType>
+std::unique_ptr<typename Dense<ValueType>::absolute_type>
+Dense<ValueType>::get_real() const
+{
+    auto exec = this->get_executor();
+
+    auto real_dense = absolute_type::create(exec, this->get_size());
+
+    exec->run(dense::make_get_real(this, real_dense.get()));
+
+    return real_dense;
+}
+
+
+template <typename ValueType>
+void Dense<ValueType>::get_real(Dense<remove_complex<ValueType>> *result) const
+{
+    auto exec = this->get_executor();
+
+    GKO_ASSERT_EQUAL_DIMENSIONS(this, result);
+
+    exec->run(
+        dense::make_get_real(this, make_temporary_clone(exec, result).get()));
+}
+
+
+template <typename ValueType>
+std::unique_ptr<typename Dense<ValueType>::absolute_type>
+Dense<ValueType>::get_imag() const
+{
+    auto exec = this->get_executor();
+
+    auto imag_dense = absolute_type::create(exec, this->get_size());
+
+    exec->run(dense::make_get_imag(this, imag_dense.get()));
+
+    return imag_dense;
+}
+
+
+template <typename ValueType>
+void Dense<ValueType>::get_imag(Dense<remove_complex<ValueType>> *result) const
+{
+    auto exec = this->get_executor();
+
+    GKO_ASSERT_EQUAL_DIMENSIONS(this, result);
+
+    exec->run(
+        dense::make_get_imag(this, make_temporary_clone(exec, result).get()));
+}
+
+
 #define GKO_DECLARE_DENSE_MATRIX(_type) class Dense<_type>
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_MATRIX);
 
diff --git a/core/matrix/dense_kernels.hpp b/core/matrix/dense_kernels.hpp
index 1f9bdae1cc2..e99d455aa77 100644
--- a/core/matrix/dense_kernels.hpp
+++ b/core/matrix/dense_kernels.hpp
@@ -182,6 +182,21 @@ namespace kernels {
         const matrix::Dense<_vtype> *source,               \
         matrix::Dense<remove_complex<_vtype>> *result)
 
+#define GKO_DECLARE_MAKE_COMPLEX_KERNEL(_vtype)                    \
+    void make_complex(std::shared_ptr<const DefaultExecutor> exec, \
+                      const matrix::Dense<_vtype> *source,         \
+                      matrix::Dense<to_complex<_vtype>> *result)
+
+#define GKO_DECLARE_GET_REAL_KERNEL(_vtype)                    \
+    void get_real(std::shared_ptr<const DefaultExecutor> exec, \
+                  const matrix::Dense<_vtype> *source,         \
+                  matrix::Dense<remove_complex<_vtype>> *result)
+
+#define GKO_DECLARE_GET_IMAG_KERNEL(_vtype)                    \
+    void get_imag(std::shared_ptr<const DefaultExecutor> exec, \
+                  const matrix::Dense<_vtype> *source,         \
+                  matrix::Dense<remove_complex<_vtype>> *result)
+
 
 #define GKO_DECLARE_ALL_AS_TEMPLATES                                        \
     template <typename ValueType>                                           \
@@ -235,7 +250,13 @@ namespace kernels {
     template <typename ValueType>                                           \
     GKO_DECLARE_INPLACE_ABSOLUTE_DENSE_KERNEL(ValueType);                   \
     template <typename ValueType>                                           \
-    GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL(ValueType)
+    GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL(ValueType);                  \
+    template <typename ValueType>                                           \
+    GKO_DECLARE_MAKE_COMPLEX_KERNEL(ValueType);                             \
+    template <typename ValueType>                                           \
+    GKO_DECLARE_GET_REAL_KERNEL(ValueType);                                 \
+    template <typename ValueType>                                           \
+    GKO_DECLARE_GET_IMAG_KERNEL(ValueType)
 
 
 namespace omp {
diff --git a/cuda/matrix/dense_kernels.cu b/cuda/matrix/dense_kernels.cu
index 72d557e7dd1..83804b43083 100644
--- a/cuda/matrix/dense_kernels.cu
+++ b/cuda/matrix/dense_kernels.cu
@@ -734,6 +734,57 @@ void outplace_absolute_dense(std::shared_ptr<const CudaExecutor> exec,
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
 
+template <typename ValueType>
+void make_complex(std::shared_ptr<const CudaExecutor> exec,
+                  const matrix::Dense<ValueType> *source,
+                  matrix::Dense<to_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    kernel::make_complex<<<grid_dim, default_block_size>>>(
+        dim[0], dim[1], as_cuda_type(source->get_const_values()),
+        source->get_stride(), as_cuda_type(result->get_values()),
+        result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+
+template <typename ValueType>
+void get_real(std::shared_ptr<const CudaExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    kernel::get_real<<<grid_dim, default_block_size>>>(
+        dim[0], dim[1], as_cuda_type(source->get_const_values()),
+        source->get_stride(), as_cuda_type(result->get_values()),
+        result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+
+template <typename ValueType>
+void get_imag(std::shared_ptr<const CudaExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    kernel::get_imag<<<grid_dim, default_block_size>>>(
+        dim[0], dim[1], as_cuda_type(source->get_const_values()),
+        source->get_stride(), as_cuda_type(result->get_values()),
+        result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
+
 }  // namespace dense
 }  // namespace cuda
 }  // namespace kernels
diff --git a/cuda/test/matrix/dense_kernels.cpp b/cuda/test/matrix/dense_kernels.cpp
index 1dee85ba561..823aabe9b2b 100644
--- a/cuda/test/matrix/dense_kernels.cpp
+++ b/cuda/test/matrix/dense_kernels.cpp
@@ -643,4 +643,76 @@ TEST_F(Dense, OutplaceAbsoluteMatrixIsEquivalentToRef)
 }
 
 
+TEST_F(Dense, MakeComplexIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = x->make_complex();
+    auto dcomplex_x = dx->make_complex();
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, MakeComplexWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = ComplexMtx::create(ref, x->get_size());
+    x->make_complex(complex_x.get());
+    auto dcomplex_x = ComplexMtx::create(cuda, x->get_size());
+    dx->make_complex(dcomplex_x.get());
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, GetRealIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = x->get_real();
+    auto dreal_x = dx->get_real();
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetRealWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = Mtx::create(ref, x->get_size());
+    x->get_real(real_x.get());
+    auto dreal_x = Mtx::create(cuda, dx->get_size());
+    dx->get_real(dreal_x.get());
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetImagIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = x->get_imag();
+    auto dimag_x = dx->get_imag();
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
+TEST_F(Dense, GetImagWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = Mtx::create(ref, x->get_size());
+    x->get_imag(imag_x.get());
+    auto dimag_x = Mtx::create(cuda, dx->get_size());
+    dx->get_imag(dimag_x.get());
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
 }  // namespace
diff --git a/dpcpp/matrix/dense_kernels.dp.cpp b/dpcpp/matrix/dense_kernels.dp.cpp
index e84519e8192..fac85ba3a14 100644
--- a/dpcpp/matrix/dense_kernels.dp.cpp
+++ b/dpcpp/matrix/dense_kernels.dp.cpp
@@ -308,6 +308,33 @@ void outplace_absolute_dense(std::shared_ptr<const DpcppExecutor> exec,
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
 
+template <typename ValueType>
+void make_complex(std::shared_ptr<const DpcppExecutor> exec,
+                  const matrix::Dense<ValueType> *source,
+                  matrix::Dense<to_complex<ValueType>> *result)
+    GKO_NOT_IMPLEMENTED;
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+
+template <typename ValueType>
+void get_real(std::shared_ptr<const DpcppExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+    GKO_NOT_IMPLEMENTED;
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+
+template <typename ValueType>
+void get_imag(std::shared_ptr<const DpcppExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+    GKO_NOT_IMPLEMENTED;
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
+
 }  // namespace dense
 }  // namespace dpcpp
 }  // namespace kernels
diff --git a/hip/matrix/dense_kernels.hip.cpp b/hip/matrix/dense_kernels.hip.cpp
index fa732e1e86a..e5de9f7df83 100644
--- a/hip/matrix/dense_kernels.hip.cpp
+++ b/hip/matrix/dense_kernels.hip.cpp
@@ -763,6 +763,60 @@ void outplace_absolute_dense(std::shared_ptr<const HipExecutor> exec,
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
 
+template <typename ValueType>
+void make_complex(std::shared_ptr<const HipExecutor> exec,
+                  const matrix::Dense<ValueType> *source,
+                  matrix::Dense<to_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    hipLaunchKernelGGL(kernel::make_complex, dim3(grid_dim),
+                       dim3(default_block_size), 0, 0, dim[0], dim[1],
+                       as_hip_type(source->get_const_values()),
+                       source->get_stride(), as_hip_type(result->get_values()),
+                       result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+
+template <typename ValueType>
+void get_real(std::shared_ptr<const HipExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    hipLaunchKernelGGL(kernel::get_real, dim3(grid_dim),
+                       dim3(default_block_size), 0, 0, dim[0], dim[1],
+                       as_hip_type(source->get_const_values()),
+                       source->get_stride(), as_hip_type(result->get_values()),
+                       result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+
+template <typename ValueType>
+void get_imag(std::shared_ptr<const HipExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    const dim3 grid_dim = ceildiv(dim[0] * dim[1], default_block_size);
+
+    hipLaunchKernelGGL(kernel::get_imag, dim3(grid_dim),
+                       dim3(default_block_size), 0, 0, dim[0], dim[1],
+                       as_hip_type(source->get_const_values()),
+                       source->get_stride(), as_hip_type(result->get_values()),
+                       result->get_stride());
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
+
 }  // namespace dense
 }  // namespace hip
 }  // namespace kernels
diff --git a/hip/test/matrix/dense_kernels.hip.cpp b/hip/test/matrix/dense_kernels.hip.cpp
index fb40a3819fa..a4f4b4e58f1 100644
--- a/hip/test/matrix/dense_kernels.hip.cpp
+++ b/hip/test/matrix/dense_kernels.hip.cpp
@@ -626,4 +626,76 @@ TEST_F(Dense, OutplaceAbsoluteMatrixIsEquivalentToRef)
 }
 
 
+TEST_F(Dense, MakeComplexIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = x->make_complex();
+    auto dcomplex_x = dx->make_complex();
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, MakeComplexWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = ComplexMtx::create(ref, x->get_size());
+    x->make_complex(complex_x.get());
+    auto dcomplex_x = ComplexMtx::create(hip, x->get_size());
+    dx->make_complex(dcomplex_x.get());
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, GetRealIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = x->get_real();
+    auto dreal_x = dx->get_real();
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetRealWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = Mtx::create(ref, x->get_size());
+    x->get_real(real_x.get());
+    auto dreal_x = Mtx::create(hip, dx->get_size());
+    dx->get_real(dreal_x.get());
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetImagIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = x->get_imag();
+    auto dimag_x = dx->get_imag();
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
+TEST_F(Dense, GetImagWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = Mtx::create(ref, x->get_size());
+    x->get_imag(imag_x.get());
+    auto dimag_x = Mtx::create(hip, dx->get_size());
+    dx->get_imag(dimag_x.get());
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
 }  // namespace
diff --git a/include/ginkgo/core/matrix/dense.hpp b/include/ginkgo/core/matrix/dense.hpp
index 65113ec204b..7bb439afc34 100644
--- a/include/ginkgo/core/matrix/dense.hpp
+++ b/include/ginkgo/core/matrix/dense.hpp
@@ -142,6 +142,7 @@ class Dense
     using mat_data = gko::matrix_data<ValueType, int64>;
     using mat_data32 = gko::matrix_data<ValueType, int32>;
     using absolute_type = remove_complex<Dense>;
+    using complex_type = to_complex<Dense>;
 
     using row_major_range = gko::range<gko::accessor::row_major<ValueType, 2>>;
 
@@ -255,6 +256,42 @@ class Dense
 
     void compute_absolute_inplace() override;
 
+    /**
+     * Creates a complex copy of the original matrix. If the original matrix
+     * was real, the imaginary part of the result will be zero.
+     */
+    std::unique_ptr<complex_type> make_complex() const;
+
+    /**
+     * Writes a complex copy of the original matrix to a given complex matrix.
+     * If the original matrix was real, the imaginary part of the result will
+     * be zero.
+     */
+    void make_complex(Dense<to_complex<ValueType>> *result) const;
+
+    /**
+     * Creates a new real matrix and extracts the real part of the original
+     * matrix into that.
+     */
+    std::unique_ptr<absolute_type> get_real() const;
+
+    /**
+     * Extracts the real part of the original matrix into a given real matrix.
+     */
+    void get_real(Dense<remove_complex<ValueType>> *result) const;
+
+    /**
+     * Creates a new real matrix and extracts the imaginary part of the
+     * original matrix into that.
+     */
+    std::unique_ptr<absolute_type> get_imag() const;
+
+    /**
+     * Extracts the imaginary part of the original matrix into a given real
+     * matrix.
+     */
+    void get_imag(Dense<remove_complex<ValueType>> *result) const;
+
     /**
      * Returns a pointer to the array of values of the matrix.
      *
diff --git a/omp/matrix/dense_kernels.cpp b/omp/matrix/dense_kernels.cpp
index afe74e3c5e4..15a9d6e9a99 100644
--- a/omp/matrix/dense_kernels.cpp
+++ b/omp/matrix/dense_kernels.cpp
@@ -805,6 +805,60 @@ void outplace_absolute_dense(std::shared_ptr<const OmpExecutor> exec,
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
 
+template <typename ValueType>
+void make_complex(std::shared_ptr<const OmpExecutor> exec,
+                  const matrix::Dense<ValueType> *source,
+                  matrix::Dense<to_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+
+#pragma omp parallel for
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = to_complex<ValueType>{source->at(row, col)};
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+
+template <typename ValueType>
+void get_real(std::shared_ptr<const OmpExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+
+#pragma omp parallel for
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = real(source->at(row, col));
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+
+template <typename ValueType>
+void get_imag(std::shared_ptr<const OmpExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+
+#pragma omp parallel for
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = imag(source->at(row, col));
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
+
 }  // namespace dense
 }  // namespace omp
 }  // namespace kernels
diff --git a/omp/test/matrix/dense_kernels.cpp b/omp/test/matrix/dense_kernels.cpp
index 06999748b9f..cf2d975c2ef 100644
--- a/omp/test/matrix/dense_kernels.cpp
+++ b/omp/test/matrix/dense_kernels.cpp
@@ -758,4 +758,76 @@ TEST_F(Dense, OutplaceAbsoluteMatrixIsEquivalentToRef)
 }
 
 
+TEST_F(Dense, MakeComplexIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = x->make_complex();
+    auto dcomplex_x = dx->make_complex();
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, MakeComplexWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto complex_x = ComplexMtx::create(ref, x->get_size());
+    x->make_complex(complex_x.get());
+    auto dcomplex_x = ComplexMtx::create(omp, x->get_size());
+    dx->make_complex(dcomplex_x.get());
+
+    GKO_ASSERT_MTX_NEAR(complex_x, dcomplex_x, 0);
+}
+
+
+TEST_F(Dense, GetRealIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = x->get_real();
+    auto dreal_x = dx->get_real();
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetRealWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto real_x = Mtx::create(ref, x->get_size());
+    x->get_real(real_x.get());
+    auto dreal_x = Mtx::create(omp, dx->get_size());
+    dx->get_real(dreal_x.get());
+
+    GKO_ASSERT_MTX_NEAR(real_x, dreal_x, 0);
+}
+
+
+TEST_F(Dense, GetImagIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = x->get_imag();
+    auto dimag_x = dx->get_imag();
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
+TEST_F(Dense, GetImagWithGivenResultIsEquivalentToRef)
+{
+    set_up_apply_data();
+
+    auto imag_x = Mtx::create(ref, x->get_size());
+    x->get_imag(imag_x.get());
+    auto dimag_x = Mtx::create(omp, dx->get_size());
+    dx->get_imag(dimag_x.get());
+
+    GKO_ASSERT_MTX_NEAR(imag_x, dimag_x, 0);
+}
+
+
 }  // namespace
diff --git a/reference/matrix/dense_kernels.cpp b/reference/matrix/dense_kernels.cpp
index f0c8f938d05..147204157e7 100644
--- a/reference/matrix/dense_kernels.cpp
+++ b/reference/matrix/dense_kernels.cpp
@@ -716,6 +716,54 @@ void outplace_absolute_dense(std::shared_ptr<const ReferenceExecutor> exec,
 GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL);
 
 
+template <typename ValueType>
+void make_complex(std::shared_ptr<const ReferenceExecutor> exec,
+                  const matrix::Dense<ValueType> *source,
+                  matrix::Dense<to_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = to_complex<ValueType>{source->at(row, col)};
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_MAKE_COMPLEX_KERNEL);
+
+
+template <typename ValueType>
+void get_real(std::shared_ptr<const ReferenceExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = real(source->at(row, col));
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_REAL_KERNEL);
+
+
+template <typename ValueType>
+void get_imag(std::shared_ptr<const ReferenceExecutor> exec,
+              const matrix::Dense<ValueType> *source,
+              matrix::Dense<remove_complex<ValueType>> *result)
+{
+    auto dim = source->get_size();
+    for (size_type row = 0; row < dim[0]; row++) {
+        for (size_type col = 0; col < dim[1]; col++) {
+            result->at(row, col) = imag(source->at(row, col));
+        }
+    }
+}
+
+GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GET_IMAG_KERNEL);
+
+
 }  // namespace dense
 }  // namespace reference
 }  // namespace kernels
diff --git a/reference/test/matrix/dense_kernels.cpp b/reference/test/matrix/dense_kernels.cpp
index b6195ffa8b0..d6618866de0 100644
--- a/reference/test/matrix/dense_kernels.cpp
+++ b/reference/test/matrix/dense_kernels.cpp
@@ -65,6 +65,8 @@ class Dense : public ::testing::Test {
 protected:
     using value_type = T;
     using Mtx = gko::matrix::Dense<value_type>;
+    using ComplexMtx = gko::to_complex<Mtx>;
+    using RealMtx = gko::remove_complex<Mtx>;
     Dense()
         : exec(gko::ReferenceExecutor::create()),
           mtx1(gko::initialize<Mtx>(4, {{1.0, 2.0, 3.0}, {1.5, 2.5, 3.5}},
@@ -2151,11 +2153,163 @@ TYPED_TEST(Dense, AdvancedAppliesToComplex)
 }
 
 
+TYPED_TEST(Dense, MakeComplex)
+{
+    using T = typename TestFixture::value_type;
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto complex_mtx = this->mtx5->make_complex();
+
+    GKO_ASSERT_MTX_NEAR(complex_mtx, this->mtx5, 0);
+}
+
+
+TYPED_TEST(Dense, MakeComplexWithGivenResult)
+{
+    using T = typename TestFixture::value_type;
+    using ComplexMtx = typename TestFixture::ComplexMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto complex_mtx = ComplexMtx::create(exec, this->mtx5->get_size());
+    this->mtx5->make_complex(complex_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(complex_mtx, this->mtx5, 0);
+}
+
+
+TYPED_TEST(Dense, MakeComplexWithGivenResultFailsForWrongDimensions)
+{
+    using T = typename TestFixture::value_type;
+    using ComplexMtx = typename TestFixture::ComplexMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto complex_mtx = ComplexMtx::create(exec);
+
+    ASSERT_THROW(this->mtx5->make_complex(complex_mtx.get()),
+                 gko::DimensionMismatch);
+}
+
+
+TYPED_TEST(Dense, GetReal)
+{
+    using T = typename TestFixture::value_type;
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto real_mtx = this->mtx5->get_real();
+
+    GKO_ASSERT_MTX_NEAR(real_mtx, this->mtx5, 0);
+}
+
+
+TYPED_TEST(Dense, GetRealWithGivenResult)
+{
+    using T = typename TestFixture::value_type;
+    using RealMtx = typename TestFixture::RealMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto real_mtx = RealMtx::create(exec, this->mtx5->get_size());
+    this->mtx5->get_real(real_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(real_mtx, this->mtx5, 0);
+}
+
+
+TYPED_TEST(Dense, GetRealWithGivenResultFailsForWrongDimensions)
+{
+    using T = typename TestFixture::value_type;
+    using RealMtx = typename TestFixture::RealMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto real_mtx = RealMtx::create(exec);
+    ASSERT_THROW(this->mtx5->get_real(real_mtx.get()), gko::DimensionMismatch);
+}
+
+
+TYPED_TEST(Dense, GetImag)
+{
+    using T = typename TestFixture::value_type;
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto imag_mtx = this->mtx5->get_imag();
+
+    GKO_ASSERT_MTX_NEAR(
+        imag_mtx, l({{0.0, 0.0, 0.0}, {0.0, 0.0, 0.0}, {0.0, 0.0, 0.0}}), 0);
+}
+
+
+TYPED_TEST(Dense, GetImagWithGivenResult)
+{
+    using T = typename TestFixture::value_type;
+    using RealMtx = typename TestFixture::RealMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto imag_mtx = RealMtx::create(exec, this->mtx5->get_size());
+    this->mtx5->get_imag(imag_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(
+        imag_mtx, l({{0.0, 0.0, 0.0}, {0.0, 0.0, 0.0}, {0.0, 0.0, 0.0}}), 0);
+}
+
+
+TYPED_TEST(Dense, GetImagWithGivenResultFailsForWrongDimensions)
+{
+    using T = typename TestFixture::value_type;
+    using RealMtx = typename TestFixture::RealMtx;
+    auto exec = this->mtx5->get_executor();
+    // clang-format off
+    // {1.0, -1.0, -0.5},
+    // {-2.0, 2.0, 4.5},
+    // {2.1, 3.4, 1.2}
+    // clang-format on
+
+    auto imag_mtx = RealMtx::create(exec);
+    ASSERT_THROW(this->mtx5->get_imag(imag_mtx.get()), gko::DimensionMismatch);
+}
+
+
 template <typename T>
 class DenseComplex : public ::testing::Test {
 protected:
     using value_type = T;
     using Mtx = gko::matrix::Dense<value_type>;
+    using RealMtx = gko::matrix::Dense<gko::remove_complex<value_type>>;
 };
 
 
@@ -2220,4 +2374,125 @@ TYPED_TEST(DenseComplex, OutplaceAbsolute)
 }
 
 
+TYPED_TEST(DenseComplex, MakeComplex)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto complex_mtx = mtx->make_complex();
+
+    GKO_ASSERT_MTX_NEAR(complex_mtx, mtx, 0.0);
+}
+
+
+TYPED_TEST(DenseComplex, MakeComplexWithGivenResult)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto complex_mtx = Mtx::create(exec, mtx->get_size());
+    mtx->make_complex(complex_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(complex_mtx, mtx, 0.0);
+}
+
+
+TYPED_TEST(DenseComplex, GetReal)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto real_mtx = mtx->get_real();
+
+    GKO_ASSERT_MTX_NEAR(
+        real_mtx, l({{1.0, 3.0, 0.0}, {-4.0, -1.0, 0.0}, {0.0, 0.0, 2.0}}),
+        0.0);
+}
+
+
+TYPED_TEST(DenseComplex, GetRealWithGivenResult)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using RealMtx = typename TestFixture::RealMtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto real_mtx = RealMtx::create(exec, mtx->get_size());
+    mtx->get_real(real_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(
+        real_mtx, l({{1.0, 3.0, 0.0}, {-4.0, -1.0, 0.0}, {0.0, 0.0, 2.0}}),
+        0.0);
+}
+
+
+TYPED_TEST(DenseComplex, GetImag)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto imag_mtx = mtx->get_imag();
+
+    GKO_ASSERT_MTX_NEAR(
+        imag_mtx, l({{0.0, 4.0, 2.0}, {-3.0, 0.0, 0.0}, {0.0, -1.5, 0.0}}),
+        0.0);
+}
+
+
+TYPED_TEST(DenseComplex, GetImagWithGivenResult)
+{
+    using Mtx = typename TestFixture::Mtx;
+    using RealMtx = typename TestFixture::RealMtx;
+    using T = typename TestFixture::value_type;
+    auto exec = gko::ReferenceExecutor::create();
+    // clang-format off
+    auto mtx = gko::initialize<Mtx>(
+        {{T{1.0, 0.0}, T{3.0, 4.0}, T{0.0, 2.0}},
+         {T{-4.0, -3.0}, T{-1.0, 0}, T{0.0, 0.0}},
+         {T{0.0, 0.0}, T{0.0, -1.5}, T{2.0, 0.0}}}, exec);
+    // clang-format on
+
+    auto imag_mtx = RealMtx::create(exec, mtx->get_size());
+    mtx->get_imag(imag_mtx.get());
+
+    GKO_ASSERT_MTX_NEAR(
+        imag_mtx, l({{0.0, 4.0, 2.0}, {-3.0, 0.0, 0.0}, {0.0, -1.5, 0.0}}),
+        0.0);
+}
+
+
 }  // namespace