-
Notifications
You must be signed in to change notification settings - Fork 8
Build with sycl_thrust for device examples if building oneMKL.
#53
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
Changes from all commits
8e272ad
e79ee7f
a841a09
1f4679e
53adaa6
0b9af32
61ad2ac
053369b
e00ccc2
56534e8
4e7c172
16dc7d1
2adca57
e0ed0be
File filter
Filter by extension
Conversations
Jump to
Diff view
Diff view
There are no files selected for viewing
| Original file line number | Diff line number | Diff line change |
|---|---|---|
|
|
@@ -56,8 +56,7 @@ int main(int argc, char** argv) { | |
| std::span<value_t> y_span(d_y.data().get(), m); | ||
|
|
||
| // y = A * x | ||
| spblas::spmv_state_t state; | ||
| spblas::multiply(state, a, x_span, y_span); | ||
| spblas::multiply(a, x_span, y_span); | ||
|
Contributor
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. where did the state go ?
Collaborator
Author
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. This should be resolved now---both rocSPARSE and cuSPARSE now take an optional
Contributor
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. so no state now in this simple example ?
Collaborator
Author
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. Correct, although you can create a state and pass it in if you want: spblas::operation_info_t state;
spblas::multiply(state, a, x_span, y_span); |
||
|
|
||
| thrust::copy(d_y.begin(), d_y.end(), y.begin()); | ||
|
|
||
|
|
||
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,34 @@ | ||
| #pragma once | ||
|
|
||
| #include <cusparse.h> | ||
| #include <memory> | ||
|
|
||
| namespace spblas { | ||
| namespace __cusparse { | ||
|
|
||
| class abstract_operation_state_t { | ||
| public: | ||
| // Common state that all operations need | ||
| cusparseHandle_t handle() const { | ||
| return handle_; | ||
| } | ||
|
|
||
| // Make std::default_delete a friend so unique_ptr can delete us | ||
| friend struct std::default_delete<abstract_operation_state_t>; | ||
|
|
||
| protected: | ||
| abstract_operation_state_t() { | ||
| cusparseCreate(&handle_); | ||
| } | ||
|
|
||
| virtual ~abstract_operation_state_t() { | ||
| if (handle_) { | ||
| cusparseDestroy(handle_); | ||
| } | ||
| } | ||
|
|
||
| cusparseHandle_t handle_; | ||
| }; | ||
|
|
||
| } // namespace __cusparse | ||
| } // namespace spblas |
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,41 @@ | ||
| #pragma once | ||
|
|
||
| #include <cusparse.h> | ||
|
|
||
| #include <spblas/detail/types.hpp> | ||
| #include <spblas/detail/view_inspectors.hpp> | ||
| #include <spblas/vendor/cusparse/exception.hpp> | ||
| #include <spblas/vendor/cusparse/types.hpp> | ||
|
|
||
| namespace spblas { | ||
|
|
||
| namespace __cusparse { | ||
|
|
||
| template <matrix M> | ||
| requires __detail::is_csr_view_v<M> | ||
| cusparseSpMatDescr_t create_cusparse_handle(M&& m) { | ||
| cusparseSpMatDescr_t mat_descr; | ||
| __cusparse::throw_if_error(cusparseCreateCsr( | ||
| &mat_descr, __backend::shape(m)[0], __backend::shape(m)[1], | ||
| m.values().size(), m.rowptr().data(), m.colind().data(), | ||
| m.values().data(), detail::cusparse_index_type_v<tensor_offset_t<M>>, | ||
| detail::cusparse_index_type_v<tensor_index_t<M>>, | ||
| CUSPARSE_INDEX_BASE_ZERO, detail::cuda_data_type_v<tensor_scalar_t<M>>)); | ||
|
|
||
| return mat_descr; | ||
| } | ||
|
|
||
| template <vector V> | ||
| requires __ranges::contiguous_range<V> | ||
| cusparseDnVecDescr_t create_cusparse_handle(V&& v) { | ||
| cusparseDnVecDescr_t vec_descr; | ||
| __cusparse::throw_if_error( | ||
| cusparseCreateDnVec(&vec_descr, __backend::shape(v), __ranges::data(v), | ||
| detail::cuda_data_type_v<tensor_scalar_t<V>>)); | ||
|
|
||
| return vec_descr; | ||
| } | ||
|
|
||
| } // namespace __cusparse | ||
|
|
||
| } // namespace spblas |
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,32 @@ | ||
| #pragma once | ||
|
|
||
| #include <cusparse.h> | ||
| #include <spblas/detail/view_inspectors.hpp> | ||
|
|
||
| namespace spblas { | ||
| namespace __cusparse { | ||
|
|
||
| // | ||
| // Takes in a CSR or CSR_transpose (aka CSC) or CSC or CSC_transpose | ||
| // and returns the cusparseOperation_t value associated with it being | ||
| // represented in the CSR format | ||
| // | ||
| // CSR = CSR + NON_TRANSPOSE | ||
| // CSR_transpose = CSR + TRANSPOSE | ||
| // CSC = CSR + TRANSPOSE | ||
| // CSC_transpose = CSR + NON_TRANSPOSE | ||
| // | ||
| template <matrix M> | ||
| cusparseOperation_t get_transpose(M&& m) { | ||
| static_assert(__detail::has_csr_base<M> || __detail::has_csc_base<M>); | ||
| if constexpr (__detail::has_base<M>) { | ||
| return get_transpose(m.base()); | ||
| } else if constexpr (__detail::is_csr_view_v<M>) { | ||
| return CUSPARSE_OPERATION_NON_TRANSPOSE; | ||
| } else if constexpr (__detail::is_csc_view_v<M>) { | ||
| return CUSPARSE_OPERATION_TRANSPOSE; | ||
| } | ||
| } | ||
|
|
||
| } // namespace __cusparse | ||
| } // namespace spblas |
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,55 @@ | ||
| #pragma once | ||
|
|
||
| #include <cusparse.h> | ||
| #include <memory> | ||
|
|
||
| #include "abstract_operation_state.hpp" | ||
|
|
||
| namespace spblas { | ||
| namespace __cusparse { | ||
|
|
||
| class spmv_state_t : public abstract_operation_state_t { | ||
| public: | ||
| spmv_state_t() = default; | ||
| ~spmv_state_t() { | ||
| if (a_descr_) { | ||
| cusparseDestroySpMat(a_descr_); | ||
| } | ||
| if (b_descr_) { | ||
| cusparseDestroyDnVec(b_descr_); | ||
| } | ||
| if (c_descr_) { | ||
| cusparseDestroyDnVec(c_descr_); | ||
| } | ||
| } | ||
|
|
||
| // Accessors for the descriptors | ||
| cusparseSpMatDescr_t a_descriptor() const { | ||
| return a_descr_; | ||
| } | ||
| cusparseDnVecDescr_t b_descriptor() const { | ||
| return b_descr_; | ||
| } | ||
| cusparseDnVecDescr_t c_descriptor() const { | ||
| return c_descr_; | ||
| } | ||
|
|
||
| // Setters for the descriptors | ||
| void set_a_descriptor(cusparseSpMatDescr_t descr) { | ||
| a_descr_ = descr; | ||
| } | ||
| void set_b_descriptor(cusparseDnVecDescr_t descr) { | ||
| b_descr_ = descr; | ||
| } | ||
| void set_c_descriptor(cusparseDnVecDescr_t descr) { | ||
| c_descr_ = descr; | ||
| } | ||
|
|
||
| private: | ||
| cusparseSpMatDescr_t a_descr_ = nullptr; | ||
| cusparseDnVecDescr_t b_descr_ = nullptr; | ||
| cusparseDnVecDescr_t c_descr_ = nullptr; | ||
| }; | ||
|
|
||
| } // namespace __cusparse | ||
| } // namespace spblas |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
can you explain this change? My understanding was that we will have the following elements in
what is operation_info_t ?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
operation_info_tis the state. I believe in the current proposal we refer to it asoperation_state_t. It just hasn't been renamed here (I can do that in a separate PR). (Perhaps confusingly, there is a non-user-visibleoperation_state_tclass, which is an implementation detail. That's how vendor backends store their data inside the state.)I think one semi-open question is whether to have different state objects for each type of operation. Personally I think it creates more complexity for the user, as they have to juggle different kinds of state objects.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
no, it should be _state_t as the paper described.
I still believe that the first step should be individual state for each component.
They can be much easier merged later rather than create a big object then later split up.