GemmSYCL.cpp File Reference

(d32b4fc (Mon Sep 14 15:46:25 2026 -0700))

#include "open3d/ml/impl/GemmSYCL.h"
#include <cutlass/gemm/device/gemm_universal.h>
#include <cutlass/gemm/device/gemm_universal_adapter.h>
#include <cutlass/kernel_hardware_info.h>
#include <cutlass/layout/matrix.h>
#include <array>
#include <cute/tensor.hpp>
#include <cutlass/epilogue/collective/collective_builder.hpp>
#include <cutlass/gemm/collective/collective_builder.hpp>
#include <cutlass/util/packed_stride.hpp>
#include <mutex>
#include <sstream>
#include <stdexcept>
#include <sycl/sycl.hpp>
#include <vector>
#include "open3d/core/SYCLContext.h"
#include "open3d/utility/Logging.h"

Data Structures
struct	open3d::ml::impl::sycl_gemm_detail::TransposeLayout< cutlass::layout::ColumnMajor >
struct	open3d::ml::impl::sycl_gemm_detail::TransposeLayout< cutlass::layout::RowMajor >
struct	open3d::ml::impl::sycl_gemm_detail::GemmArgs

Namespaces
namespace	open3d
namespace	open3d::ml
namespace	open3d::ml::impl
namespace	open3d::ml::impl::sycl_gemm_detail

Typedefs
using	open3d::ml::impl::sycl_gemm_detail::Tile = cute::Shape< cute::_16, cute::_16, cute::_8 >

Functions
template<class Layout >
auto	open3d::ml::impl::sycl_gemm_detail::MakeStrideA (int64_t ld)
template<class Layout >
auto	open3d::ml::impl::sycl_gemm_detail::MakeStrideB (int64_t ld)
template<class TileShape , class LayoutA , class LayoutB >
std::pair< cutlass::Status, sycl::event >	open3d::ml::impl::sycl_gemm_detail::RunGemmXmxTf32RowMajorOutput (sycl::queue &queue, int m, int n, int k, float alpha, const float *A, int64_t lda, const float *B, int64_t ldb, float beta, const float *C, int64_t ldc, float *D, int64_t ldd, const std::vector< sycl::event > &deps={})
template<class TileShape , class LayoutA , class LayoutB >
std::pair< cutlass::Status, sycl::event >	open3d::ml::impl::sycl_gemm_detail::RunGemmIeeeFp32RowMajorOutput (sycl::queue &queue, int m, int n, int k, float alpha, const float *A, int64_t lda, const float *B, int64_t ldb, float beta, const float *C, int64_t ldc, float *D, int64_t ldd, const std::vector< sycl::event > &deps={})
template<class LayoutA , class LayoutB >
std::pair< cutlass::Status, sycl::event >	open3d::ml::impl::sycl_gemm_detail::RunTf32 (const GemmArgs &a)
	Runs the TF32/XMX path.
template<class LayoutA , class LayoutB >
std::pair< cutlass::Status, sycl::event >	open3d::ml::impl::sycl_gemm_detail::RunIeee (const GemmArgs &a)
	Runs the IEEE fp32 path.
bool	open3d::ml::impl::sycl_gemm_detail::IsTf32Aligned (int64_t value)
template<class LayoutA , class LayoutB >
sycl::event	open3d::ml::impl::GemmColumnMajorSYCL (sycl::queue &queue, int m, int n, int k, float alpha, const float *A, int64_t lda, const float *B, int64_t ldb, float beta, float *C, int64_t ldc, bool allow_tf32, const std::vector< sycl::event > &deps)
template sycl::event	open3d::ml::impl::GemmColumnMajorSYCL< cutlass::layout::ColumnMajor, cutlass::layout::ColumnMajor > (sycl::queue &, int, int, int, float, const float *, int64_t, const float *, int64_t, float, float *, int64_t, bool, const std::vector< sycl::event > &)
template sycl::event	open3d::ml::impl::GemmColumnMajorSYCL< cutlass::layout::ColumnMajor, cutlass::layout::RowMajor > (sycl::queue &, int, int, int, float, const float *, int64_t, const float *, int64_t, float, float *, int64_t, bool, const std::vector< sycl::event > &)