csr_accessor.cpp

/*******************************************************************************
* Copyright 2023 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.
*******************************************************************************/

#include <sycl/sycl.hpp>
#include <iostream>

#ifndef ONEDAL_DATA_PARALLEL
#define ONEDAL_DATA_PARALLEL
#endif

#include "oneapi/dal/table/csr_accessor.hpp"
#include "oneapi/dal/table/csr.hpp"

#include "example_util/dpc_helpers.hpp"

namespace dal = oneapi::dal;

void run(sycl::queue &q) {
    constexpr std::int64_t row_count = 4;
    constexpr std::int64_t column_count = 4;
    constexpr std::int64_t element_count = 7;

    // create arrays of data, column indices, and row offsets of the table
    // in sparse CSR storage format on host
    const float data_host[] = { 1.0f, 2.0f, 3.0f, 4.0f, 1.0f, 11.0f, 8.0f };
    const std::int64_t column_indices_host[] = { 1, 2, 4, 3, 2, 4, 2 };
    const std::int64_t row_offsets_host[] = { 1, 4, 5, 7, 8 };

    // allocate SYCL shared memory for storing data, column indices, and row offset arrays
    auto data = sycl::malloc_shared<float>(element_count, q);
    auto column_indices = sycl::malloc_shared<std::int64_t>(element_count, q);
    auto row_offsets = sycl::malloc_shared<std::int64_t>(row_count + 1, q);

    // copy data, column indices, and row offset arrays from the host to the SYCL shared memory
    auto data_event = q.memcpy(data, data_host, sizeof(float) * element_count);
    auto column_indices_event =
        q.memcpy(column_indices, column_indices_host, sizeof(std::int64_t) * element_count);
    auto row_offsets_event =
        q.memcpy(row_offsets, row_offsets_host, sizeof(std::int64_t) * (row_count + 1));

    // create a sparse table in CSR format from arrays of data, column indices, and row offsets
    // that are allocated in SYCL shared memory
    auto table = dal::csr_table{ q,
                                 data,
                                 column_indices,
                                 row_offsets,
                                 row_count,
                                 column_count,
                                 dal::detail::make_default_delete<const float>(q),
                                 dal::detail::make_default_delete<const std::int64_t>(q),
                                 dal::detail::make_default_delete<const std::int64_t>(q),
                                 dal::sparse_indexing::one_based,
                                 { data_event, column_indices_event, row_offsets_event } };
    dal::csr_accessor<const float> acc{ table };

    // pull the second and third rows of the sparse table
    // the pulled rows have one-based indices by default
    const auto [block_data, block_column_indices, block_row_offsets] = acc.pull(q, { 1, 3 });

    std::cout << "Print the original sparse data table as 3 arrays in CSR storage format:"
              << std::endl;
    std::cout << "Values of the table:" << std::endl;
    for (std::int64_t i = 0; i < element_count; i++) {
        std::cout << data_host[i] << ", ";
    }
    std::cout << std::endl << "Column indices of the table:" << std::endl;
    for (std::int64_t i = 0; i < element_count; i++) {
        std::cout << column_indices_host[i] << ", ";
    }
    std::cout << std::endl << "Row offsets of the table:" << std::endl;
    for (std::int64_t i = 0; i < row_count + 1; i++) {
        std::cout << row_offsets_host[i] << ", ";
    }
    std::cout << std::endl;

    std::cout << std::endl << "Print 2 rows from CSR table as dense float arrays" << std::endl;
    std::cout << "Values in the second and third rows of the table as dense float array:"
              << std::endl;
    for (std::int64_t i = 0; i < block_data.get_count(); i++) {
        std::cout << block_data[i] << ", ";
    }
    std::cout << std::endl
              << "Column indices of the data in the second and third rows from CSR table:"
              << std::endl;
    for (std::int64_t i = 0; i < block_column_indices.get_count(); i++) {
        std::cout << block_column_indices[i] << ", ";
    }
    std::cout << std::endl
              << "Row offsets of the second and third rows from CSR table:" << std::endl;
    for (std::int64_t i = 0; i < block_row_offsets.get_count(); i++) {
        std::cout << block_row_offsets[i] << ", ";
    }
    std::cout << std::endl;
}

int main(int argc, char const *argv[]) {
    for (auto d : list_devices()) {
        std::cout << "Running on " << d.get_platform().get_info<sycl::info::platform::name>()
                  << ", " << d.get_info<sycl::info::device::name>() << "\n"
                  << std::endl;
        auto q = sycl::queue{ d };
        run(q);
    }
    return 0;
}