| 64 | */ |
| 65 | template<typename Scalar> |
| 66 | void directTranspose(Scalar* dst, const Scalar* src, Index rows, Index cols, Index batches, cublasOperation_t transOp = CUBLAS_OP_T) |
| 67 | { |
| 68 | //thrust::complex<double> has no alignment requirements, |
| 69 | //while cublas cuComplexDouble requires 16B-alignment. |
| 70 | //If this is not fullfilled, a segfault is thrown. |
| 71 | //This hack enforces that. |
| 72 | #ifdef _MSC_VER |
| 73 | __declspec(align(16)) Scalar alpha(1); |
| 74 | __declspec(align(16)) Scalar beta(0); |
| 75 | #else |
| 76 | Scalar alpha __attribute__((aligned(16))) = 1; |
| 77 | Scalar beta __attribute__((aligned(16))) = 0; |
| 78 | #endif |
| 79 | |
| 80 | int m = static_cast<int>(rows); |
| 81 | int n = static_cast<int>(cols); |
| 82 | |
| 83 | cublasOperation_t transB = CUBLAS_OP_N; |
| 84 | |
| 85 | const Scalar* A = src; |
| 86 | int lda = n; |
| 87 | const Scalar* B = nullptr; |
| 88 | int ldb = m; |
| 89 | Scalar* C = dst; |
| 90 | int ldc = m; |
| 91 | size_t batch_offset = size_t(m) * n; |
| 92 | //TODO: parallelize over multiple streams |
| 93 | for (Index batch = 0; batch < batches; ++batch) { |
| 94 | internal::CublasApi::current().cublasGeam( |
| 95 | transOp, transB, m, n, |
| 96 | internal::CublasApi::cast(&alpha), internal::CublasApi::cast(A + batch*batch_offset), lda, |
| 97 | internal::CublasApi::cast(&beta), internal::CublasApi::cast(B), ldb, |
| 98 | internal::CublasApi::cast(C + batch*batch_offset), ldc); |
| 99 | } |
| 100 | |
| 101 | CUMAT_PROFILING_INC(EvalTranspose); |
| 102 | CUMAT_PROFILING_INC(EvalAny); |
| 103 | } |
| 104 | } |
| 105 | |
| 106 | /** |
no test coverage detected