Skip to content

P
Paddle

Crayon鑫 / Paddle
与 Fork 源项目一致

Fork自 PaddlePaddle / Paddle

通知 1
Star 1
Fork 0
P
Paddle
未验证 提交 a1e40dc6 编写于 作者: F Feiyu Chan 提交者： GitHub
浏览文件
操作

[Pten] Migration of eigen numeric extensions and functors in paddle/fluid/operatos/eigen (#39124) 

* migration of functors in paddle/fluid/operators/eigen and paddle/fluid/platform/eigen_ext.h
* update path of data types like float16.h in includes in extensions.h
上级 f783b846
隐藏空白更改
内联 并排
Showing with 855 addition and 606 deletion
paddle/fluid/framework/data_type.h
浏览文件 @ a1e40dc6
...@@ -20,9 +20,9 @@ limitations under the License. */ ...@@ -20,9 +20,9 @@ limitations under the License. */
#include "paddle/fluid/framework/framework.pb.h" #include "paddle/fluid/framework/framework.pb.h"
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/fluid/platform/bfloat16.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/fluid/platform/complex.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/enforce.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/fluid/platform/float16.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace paddle {
namespace framework { namespace framework {
......
paddle/fluid/operators/CMakeLists.txt
浏览文件 @ a1e40dc6
...@@ -8,7 +8,6 @@ set(pybind_file_final ${PADDLE_BINARY_DIR}/paddle/fluid/pybind/pybind.h) ...@@ -8,7 +8,6 @@ set(pybind_file_final ${PADDLE_BINARY_DIR}/paddle/fluid/pybind/pybind.h)
file(WRITE ${pybind_file} "// Generated by the paddle/fluid/operators/CMakeLists.txt. DO NOT EDIT!\n\n") file(WRITE ${pybind_file} "// Generated by the paddle/fluid/operators/CMakeLists.txt. DO NOT EDIT!\n\n")
add_subdirectory(math) add_subdirectory(math)
add_subdirectory(eigen)
add_subdirectory(controlflow) add_subdirectory(controlflow)
add_subdirectory(detection) add_subdirectory(detection)
add_subdirectory(elementwise) add_subdirectory(elementwise)
......
paddle/fluid/operators/eigen/CMakeLists.txt 已删除 100644 → 0
浏览文件 @ f783b846
file(GLOB EIGEN_CC_SOURCES RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*.cc")
file(GLOB EIGEN_CU_SOURCES RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*.cu")
if(WITH_GPU)
nv_library(eigen_function SRCS ${EIGEN_CC_SOURCES} ${EIGEN_CU_SOURCES} DEPS eigen3)
elseif(WITH_ROCM)
hip_library(eigen_function SRCS ${EIGEN_CC_SOURCES} ${EIGEN_CU_SOURCES} DEPS eigen3)
else()
cc_library(eigen_function SRCS ${EIGEN_CC_SOURCES} DEPS eigen3)
endif()
paddle/fluid/operators/eigen/eigen_function.h
浏览文件 @ a1e40dc6
...@@ -18,243 +18,71 @@ limitations under the License. */ ...@@ -18,243 +18,71 @@ limitations under the License. */
#ifndef NOMINMAX #ifndef NOMINMAX
#define NOMINMAX #define NOMINMAX
#endif #endif
#include "unsupported/Eigen/CXX11/Tensor" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace paddle {
namespace operators { namespace operators {
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenBroadcast { using EigenBroadcast = pten::funcs::EigenBroadcast<EigenDevice, T, Rank>;
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev, OutType out, InType in,
const Array& bcast);
static void Eval(const EigenDevice& dev, OutType32BitIndex out,
InType32BitIndex in, const Array& bcast);
};
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenBroadcastGrad { using EigenBroadcastGrad =
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>; pten::funcs::EigenBroadcastGrad<EigenDevice, T, Rank>;
using Array2 = Eigen::DSizes<Eigen::DenseIndex, Rank * 2>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, InType in,
const Array& reduce_dims, const Array2& reshape_dims);
};
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenConstant { using EigenConstant = pten::funcs::EigenConstant<EigenDevice, T, Rank>;
using Type = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, Type out, const T value);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenSign { using EigenSign = pten::funcs::EigenSign<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenReverse { using EigenReverse = pten::funcs::EigenReverse<EigenDevice, T, Rank>;
using Array = Eigen::DSizes<bool, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in,
const Array& reverse);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenAdd { using EigenAdd = pten::funcs::EigenAdd<EigenDevice, T>;
using InType = Eigen::TensorMap<Eigen::TensorFixedSize<
const T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in,
const T value);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenSub { using EigenSub = pten::funcs::EigenSub<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& left,
const InType& right);
};
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenSlice { using EigenSlice = pten::funcs::EigenSlice<EigenDevice, T, Rank>;
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>;
using Array32Bit = Eigen::DSizes<int, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in,
const Array& offsets, const Array& extents);
static void Eval(const EigenDevice& dev, OutType32BitIndex out,
const InType32BitIndex& in, const Array32Bit& offsets,
const Array32Bit& extents);
};
template <typename EigenDevice, typename T, int Rank> template <typename EigenDevice, typename T, int Rank>
struct EigenPad { using EigenPad = pten::funcs::EigenPad<EigenDevice, T, Rank>;
using Array = std::array<std::pair<int64_t, int64_t>, Rank>;
using Array32Bit = std::array<std::pair<int, int>, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in,
const Array& padding, const T value);
static void Eval(const EigenDevice& dev, OutType32BitIndex out,
const InType32BitIndex& in, const Array32Bit& padding,
const T value);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenScale { using EigenScale = pten::funcs::EigenScale<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in,
const T scale, const T bias, const bool bias_after_scale);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenErf { using EigenErf = pten::funcs::EigenErf<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenErfGrad { using EigenErfGrad = pten::funcs::EigenErfGrad<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType din, const InType& in,
const InType& dout);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenRankLoss { using EigenRankLoss = pten::funcs::EigenRankLoss<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& label,
const InType& left, const InType& right);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenRankLossGrad { using EigenRankLossGrad = pten::funcs::EigenRankLossGrad<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void EvalLeft(const EigenDevice& dev, OutType dleft,
const InType& dout, const InType& label,
const InType& left, const InType& right);
static void EvalRight(const EigenDevice& dev, OutType dright,
const InType& dout, const InType& label,
const InType& left, const InType& right);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenLogLoss { using EigenLogLoss = pten::funcs::EigenLogLoss<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& pred,
const InType& label, const T& epsilon);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenLogLossGrad { using EigenLogLossGrad = pten::funcs::EigenLogLossGrad<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType dpred, const InType& dloss,
const InType& pred, const InType& label, const T& epsilon);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenHingeLoss { using EigenHingeLoss = pten::funcs::EigenHingeLoss<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType loss, const InType& pred,
const InType& label);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenHingeLossGrad { using EigenHingeLossGrad = pten::funcs::EigenHingeLossGrad<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType dpred, const InType& dloss,
const InType& pred, const InType& label);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenL1Norm { using EigenL1Norm = pten::funcs::EigenL1Norm<EigenDevice, T>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T> template <typename EigenDevice, typename T>
struct EigenL1NormGrad { using EigenL1NormGrad = pten::funcs::EigenL1NormGrad<EigenDevice, T>;
using Array = Eigen::DSizes<Eigen::DenseIndex, 1>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType din, const InType& dout,
const InType& in, const Array& bcast);
};
} // namespace operators } // namespace operators
} // namespace paddle } // namespace paddle
paddle/fluid/operators/kernel_primitives/functor_primitives.h
浏览文件 @ a1e40dc6
...@@ -15,9 +15,9 @@ ...@@ -15,9 +15,9 @@
#pragma once #pragma once
#include "paddle/fluid/operators/amp/fp16_type_traits.h" #include "paddle/fluid/operators/amp/fp16_type_traits.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/enforce.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/fluid/platform/float16.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace paddle {
namespace operators { namespace operators {
......
paddle/fluid/operators/optimizers/lamb_op.h
浏览文件 @ a1e40dc6
...@@ -22,8 +22,8 @@ limitations under the License. */ ...@@ -22,8 +22,8 @@ limitations under the License. */
#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/operators/math/algorithm.h"
#include "paddle/fluid/operators/math/selected_rows_functor.h" #include "paddle/fluid/operators/math/selected_rows_functor.h"
#include "paddle/fluid/operators/math/squared_l2_norm.h" #include "paddle/fluid/operators/math/squared_l2_norm.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/for_range.h" #include "paddle/fluid/platform/for_range.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace paddle {
namespace operators { namespace operators {
......
paddle/fluid/platform/bfloat16_test.cc
浏览文件 @ a1e40dc6
...@@ -10,7 +10,7 @@ See the License for the specific language governing permissions and ...@@ -10,7 +10,7 @@ See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/fluid/platform/bfloat16.h"
#include "paddle/fluid/platform/eigen_ext.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h"
#define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h #define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h
#include "gtest/gtest.h" #include "gtest/gtest.h"
......
paddle/fluid/platform/complex_test.cc
浏览文件 @ a1e40dc6
...@@ -14,7 +14,7 @@ ...@@ -14,7 +14,7 @@
#include "paddle/fluid/platform/complex.h" #include "paddle/fluid/platform/complex.h"
#include <complex> #include <complex>
#include "paddle/fluid/platform/eigen_ext.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h"
#define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h #define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h
#include "gtest/gtest.h" #include "gtest/gtest.h"
...@@ -267,56 +267,56 @@ TEST(complex, print) { ...@@ -267,56 +267,56 @@ TEST(complex, print) {
TEST(complex, isinf) { TEST(complex, isinf) {
// *********** complex<float> ************* // *********** complex<float> *************
complex<float> a; complex<float> a;
a.real = float(INFINITY); a.real = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(a), true); EXPECT_EQ(std::isinf(a), true);
a.imag = float(INFINITY); a.imag = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(a), true); EXPECT_EQ(std::isinf(a), true);
complex<float> b = float(INFINITY); complex<float> b = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(b), true); EXPECT_EQ(std::isinf(b), true);
complex<float> c(float(INFINITY), 0); complex<float> c(static_cast<float>(INFINITY), 0);
EXPECT_EQ(std::isinf(c), true); EXPECT_EQ(std::isinf(c), true);
// *********** complex<double> ************* // *********** complex<double> *************
complex<double> a1; complex<double> a1;
a1.real = double(INFINITY); a1.real = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(a1), true); EXPECT_EQ(std::isinf(a1), true);
a1.imag = double(INFINITY); a1.imag = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(a1), true); EXPECT_EQ(std::isinf(a1), true);
complex<double> b1 = double(INFINITY); complex<double> b1 = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(b1), true); EXPECT_EQ(std::isinf(b1), true);
complex<double> c1(double(INFINITY), 0); complex<double> c1(static_cast<double>(INFINITY), 0);
EXPECT_EQ(std::isinf(c1), true); EXPECT_EQ(std::isinf(c1), true);
} }
TEST(complex, isnan) { TEST(complex, isnan) {
// *********** complex<float> ************* // *********** complex<float> *************
complex<float> a; complex<float> a;
a.real = float(NAN); a.real = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(a), true); EXPECT_EQ(std::isnan(a), true);
a.imag = float(NAN); a.imag = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(a), true); EXPECT_EQ(std::isnan(a), true);
complex<float> b = float(NAN); complex<float> b = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(b), true); EXPECT_EQ(std::isnan(b), true);
complex<float> c(float(NAN), 0); complex<float> c(static_cast<float>(NAN), 0);
EXPECT_EQ(std::isnan(c), true); EXPECT_EQ(std::isnan(c), true);
// *********** complex<double> ************* // *********** complex<double> *************
complex<double> a1; complex<double> a1;
a1.real = double(NAN); a1.real = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(a1), true); EXPECT_EQ(std::isnan(a1), true);
a1.imag = double(NAN); a1.imag = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(a1), true); EXPECT_EQ(std::isnan(a1), true);
complex<double> b1 = double(NAN); complex<double> b1 = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(b1), true); EXPECT_EQ(std::isnan(b1), true);
complex<double> c1(double(NAN), 0); complex<double> c1(static_cast<double>(NAN), 0);
EXPECT_EQ(std::isnan(c1), true); EXPECT_EQ(std::isnan(c1), true);
} }
......
paddle/fluid/platform/complex_test.cu
浏览文件 @ a1e40dc6
...@@ -23,8 +23,8 @@ ...@@ -23,8 +23,8 @@
#include "paddle/fluid/framework/lod_tensor.h" #include "paddle/fluid/framework/lod_tensor.h"
#include "paddle/fluid/framework/tensor_util.h" #include "paddle/fluid/framework/tensor_util.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/enforce.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
#if defined(PADDLE_WITH_CUDA) || defined(PADDLE_WITH_HIP) #if defined(PADDLE_WITH_CUDA) || defined(PADDLE_WITH_HIP)
namespace paddle { namespace paddle {
...@@ -303,59 +303,59 @@ TEST(complex, print) { ...@@ -303,59 +303,59 @@ TEST(complex, print) {
TEST(complex, isinf) { TEST(complex, isinf) {
// *********** complex<float> ************* // *********** complex<float> *************
complex<float> a; complex<float> a;
a.real = float(INFINITY); a.real = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(a), true); EXPECT_EQ(std::isinf(a), true);
a.imag = float(INFINITY); a.imag = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(a), true); EXPECT_EQ(std::isinf(a), true);
complex<float> b = float(INFINITY); complex<float> b = static_cast<float>(INFINITY);
EXPECT_EQ(std::isinf(b), true); EXPECT_EQ(std::isinf(b), true);
complex<float> c(float(INFINITY), 0); complex<float> c(static_cast<float>(INFINITY), 0);
EXPECT_EQ(std::isinf(c), true); EXPECT_EQ(std::isinf(c), true);
// *********** complex<double> ************* // *********** complex<double> *************
complex<double> a1; complex<double> a1;
a1.real = double(INFINITY); a1.real = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(a1), true); EXPECT_EQ(std::isinf(a1), true);
a1.imag = double(INFINITY); a1.imag = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(a1), true); EXPECT_EQ(std::isinf(a1), true);
complex<double> b1 = double(INFINITY); complex<double> b1 = static_cast<double>(INFINITY);
EXPECT_EQ(std::isinf(b1), true); EXPECT_EQ(std::isinf(b1), true);
complex<double> c1(double(INFINITY), 0); complex<double> c1(static_cast<double>(INFINITY), 0);
EXPECT_EQ(std::isinf(c1), true); EXPECT_EQ(std::isinf(c1), true);
} }
TEST(complex, isnan) { TEST(complex, isnan) {
// *********** complex<float> ************* // *********** complex<float> *************
complex<float> a; complex<float> a;
a.real = float(NAN); a.real = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(a), true); EXPECT_EQ(std::isnan(a), true);
a.imag = float(NAN); a.imag = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(a), true); EXPECT_EQ(std::isnan(a), true);
complex<float> b = float(NAN); complex<float> b = static_cast<float>(NAN);
EXPECT_EQ(std::isnan(b), true); EXPECT_EQ(std::isnan(b), true);
complex<float> c(float(NAN), 0); complex<float> c(static_cast<float>(NAN), 0);
EXPECT_EQ(std::isnan(c), true); EXPECT_EQ(std::isnan(c), true);
// *********** complex<double> ************* // *********** complex<double> *************
complex<double> a1; complex<double> a1;
a1.real = double(NAN); a1.real = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(a1), true); EXPECT_EQ(std::isnan(a1), true);
a1.imag = double(NAN); a1.imag = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(a1), true); EXPECT_EQ(std::isnan(a1), true);
complex<double> b1 = double(NAN); complex<double> b1 = static_cast<double>(NAN);
EXPECT_EQ(std::isnan(b1), true); EXPECT_EQ(std::isnan(b1), true);
complex<double> c1(double(NAN), 0); complex<double> c1(static_cast<double>(NAN), 0);
EXPECT_EQ(std::isnan(c1), true); EXPECT_EQ(std::isnan(c1), true);
} }
} // namespace platform } // namespace platform
} // namespace paddle } // namespace paddle
#endif #endif
\ No newline at end of file
paddle/fluid/platform/float16_test.cc
浏览文件 @ a1e40dc6
...@@ -14,8 +14,8 @@ limitations under the License. */ ...@@ -14,8 +14,8 @@ limitations under the License. */
#define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h #define GLOG_NO_ABBREVIATED_SEVERITIES // msvc conflict logging with windows.h
#include "gtest/gtest.h" #include "gtest/gtest.h"
#include "paddle/fluid/framework/lod_tensor.h" #include "paddle/fluid/framework/lod_tensor.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/enforce.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace paddle {
namespace platform { namespace platform {
......
paddle/fluid/platform/float16_test.cu
浏览文件 @ a1e40dc6
...@@ -19,8 +19,8 @@ limitations under the License. */ ...@@ -19,8 +19,8 @@ limitations under the License. */
#include "paddle/fluid/framework/lod_tensor.h" #include "paddle/fluid/framework/lod_tensor.h"
#include "paddle/fluid/framework/tensor_util.h" #include "paddle/fluid/framework/tensor_util.h"
#include "paddle/fluid/platform/eigen_ext.h"
#include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/enforce.h"
#include "paddle/pten/kernels/funcs/eigen/extensions.h"
#define ARITHMETIC_KERNEL(op_type, sign) \ #define ARITHMETIC_KERNEL(op_type, sign) \
__global__ void op_type(const half *in1, const half *in2, half *out) { \ __global__ void op_type(const half *in1, const half *in2, half *out) { \
......
paddle/pten/kernels/funcs/eigen/CMakeLists.txt
浏览文件 @ a1e40dc6
file(GLOB EIGEN_CC_SOURCES RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*.cc")
file(GLOB EIGEN_CU_SOURCES RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*.cu")
if(WITH_GPU)
nv_library(eigen_function SRCS ${EIGEN_CC_SOURCES} ${EIGEN_CU_SOURCES} DEPS eigen3)
elseif(WITH_ROCM)
hip_library(eigen_function SRCS ${EIGEN_CC_SOURCES} ${EIGEN_CU_SOURCES} DEPS eigen3)
else()
cc_library(eigen_function SRCS ${EIGEN_CC_SOURCES} DEPS eigen3)
endif()
paddle/fluid/operators/eigen/broadcast.cc → paddle/pten/kernels/funcs/eigen/broadcast.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenBroadcast<Eigen::DefaultDevice, T, Rank> { struct EigenBroadcast<Eigen::DefaultDevice, T, Rank> {
...@@ -31,13 +31,17 @@ struct EigenBroadcast<Eigen::DefaultDevice, T, Rank> { ...@@ -31,13 +31,17 @@ struct EigenBroadcast<Eigen::DefaultDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, InType in, static void Eval(const Eigen::DefaultDevice& dev,
OutType out,
InType in,
const Array& bcast) { const Array& bcast) {
out.device(dev) = in.broadcast(bcast); out.device(dev) = in.broadcast(bcast);
} }
static void Eval(const Eigen::DefaultDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::DefaultDevice& dev,
InType32BitIndex in, const Array& bcast) { OutType32BitIndex out,
InType32BitIndex in,
const Array& bcast) {
out.device(dev) = in.broadcast(bcast); out.device(dev) = in.broadcast(bcast);
} }
}; };
...@@ -50,8 +54,11 @@ struct EigenBroadcastGrad<Eigen::DefaultDevice, T, Rank> { ...@@ -50,8 +54,11 @@ struct EigenBroadcastGrad<Eigen::DefaultDevice, T, Rank> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, InType in, static void Eval(const Eigen::DefaultDevice& dev,
const Array& reduce_dims, const Array2& reshape_dims) { OutType out,
InType in,
const Array& reduce_dims,
const Array2& reshape_dims) {
out.device(dev) = out.device(dev) =
in.reshape(reshape_dims).sum(reduce_dims).reshape(out.dimensions()); in.reshape(reshape_dims).sum(reduce_dims).reshape(out.dimensions());
} }
...@@ -65,14 +72,14 @@ struct EigenBroadcastGrad<Eigen::DefaultDevice, T, Rank> { ...@@ -65,14 +72,14 @@ struct EigenBroadcastGrad<Eigen::DefaultDevice, T, Rank> {
template struct FUNCTOR<Eigen::DefaultDevice, T, 5>; \ template struct FUNCTOR<Eigen::DefaultDevice, T, 5>; \
template struct FUNCTOR<Eigen::DefaultDevice, T, 6> template struct FUNCTOR<Eigen::DefaultDevice, T, 6>
INSTANTIATION(EigenBroadcast, bool); INSTANTIATION(EigenBroadcast, bool);
INSTANTIATION(EigenBroadcast, platform::float16); INSTANTIATION(EigenBroadcast, dtype::float16);
INSTANTIATION(EigenBroadcast, float); INSTANTIATION(EigenBroadcast, float);
INSTANTIATION(EigenBroadcast, double); INSTANTIATION(EigenBroadcast, double);
INSTANTIATION(EigenBroadcast, int); INSTANTIATION(EigenBroadcast, int);
INSTANTIATION(EigenBroadcast, int64_t); INSTANTIATION(EigenBroadcast, int64_t);
INSTANTIATION(EigenBroadcastGrad, bool); INSTANTIATION(EigenBroadcastGrad, bool);
INSTANTIATION(EigenBroadcastGrad, float); INSTANTIATION(EigenBroadcastGrad, float);
INSTANTIATION(EigenBroadcastGrad, platform::float16); INSTANTIATION(EigenBroadcastGrad, dtype::float16);
INSTANTIATION(EigenBroadcastGrad, double); INSTANTIATION(EigenBroadcastGrad, double);
INSTANTIATION(EigenBroadcastGrad, int); INSTANTIATION(EigenBroadcastGrad, int);
INSTANTIATION(EigenBroadcastGrad, int64_t); INSTANTIATION(EigenBroadcastGrad, int64_t);
...@@ -82,5 +89,5 @@ template struct EigenBroadcastGrad<Eigen::DefaultDevice, int, 0>; ...@@ -82,5 +89,5 @@ template struct EigenBroadcastGrad<Eigen::DefaultDevice, int, 0>;
template struct EigenBroadcastGrad<Eigen::DefaultDevice, int64_t, 0>; template struct EigenBroadcastGrad<Eigen::DefaultDevice, int64_t, 0>;
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/broadcast.cu → paddle/pten/kernels/funcs/eigen/broadcast.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenBroadcast<Eigen::GpuDevice, T, Rank> { struct EigenBroadcast<Eigen::GpuDevice, T, Rank> {
...@@ -31,13 +31,17 @@ struct EigenBroadcast<Eigen::GpuDevice, T, Rank> { ...@@ -31,13 +31,17 @@ struct EigenBroadcast<Eigen::GpuDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, InType in, static void Eval(const Eigen::GpuDevice& dev,
OutType out,
InType in,
const Array& bcast) { const Array& bcast) {
out.device(dev) = in.broadcast(bcast); out.device(dev) = in.broadcast(bcast);
} }
static void Eval(const Eigen::GpuDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::GpuDevice& dev,
InType32BitIndex in, const Array& bcast) { OutType32BitIndex out,
InType32BitIndex in,
const Array& bcast) {
out.device(dev) = in.broadcast(bcast); out.device(dev) = in.broadcast(bcast);
} }
}; };
...@@ -50,8 +54,11 @@ struct EigenBroadcastGrad<Eigen::GpuDevice, T, Rank> { ...@@ -50,8 +54,11 @@ struct EigenBroadcastGrad<Eigen::GpuDevice, T, Rank> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, InType in, static void Eval(const Eigen::GpuDevice& dev,
const Array& reduce_dims, const Array2& reshape_dims) { OutType out,
InType in,
const Array& reduce_dims,
const Array2& reshape_dims) {
out.device(dev) = out.device(dev) =
in.reshape(reshape_dims).sum(reduce_dims).reshape(out.dimensions()); in.reshape(reshape_dims).sum(reduce_dims).reshape(out.dimensions());
} }
...@@ -65,23 +72,23 @@ struct EigenBroadcastGrad<Eigen::GpuDevice, T, Rank> { ...@@ -65,23 +72,23 @@ struct EigenBroadcastGrad<Eigen::GpuDevice, T, Rank> {
template struct FUNCTOR<Eigen::GpuDevice, T, 5>; \ template struct FUNCTOR<Eigen::GpuDevice, T, 5>; \
template struct FUNCTOR<Eigen::GpuDevice, T, 6> template struct FUNCTOR<Eigen::GpuDevice, T, 6>
INSTANTIATION(EigenBroadcast, bool); INSTANTIATION(EigenBroadcast, bool);
INSTANTIATION(EigenBroadcast, platform::float16); INSTANTIATION(EigenBroadcast, dtype::float16);
INSTANTIATION(EigenBroadcast, float); INSTANTIATION(EigenBroadcast, float);
INSTANTIATION(EigenBroadcast, double); INSTANTIATION(EigenBroadcast, double);
INSTANTIATION(EigenBroadcast, int); INSTANTIATION(EigenBroadcast, int);
INSTANTIATION(EigenBroadcast, int64_t); INSTANTIATION(EigenBroadcast, int64_t);
INSTANTIATION(EigenBroadcastGrad, bool); INSTANTIATION(EigenBroadcastGrad, bool);
INSTANTIATION(EigenBroadcastGrad, float); INSTANTIATION(EigenBroadcastGrad, float);
INSTANTIATION(EigenBroadcastGrad, platform::float16); INSTANTIATION(EigenBroadcastGrad, dtype::float16);
INSTANTIATION(EigenBroadcastGrad, double); INSTANTIATION(EigenBroadcastGrad, double);
INSTANTIATION(EigenBroadcastGrad, int); INSTANTIATION(EigenBroadcastGrad, int);
INSTANTIATION(EigenBroadcastGrad, int64_t); INSTANTIATION(EigenBroadcastGrad, int64_t);
template struct EigenBroadcastGrad<Eigen::GpuDevice, float, 0>; template struct EigenBroadcastGrad<Eigen::GpuDevice, float, 0>;
template struct EigenBroadcastGrad<Eigen::GpuDevice, platform::float16, 0>; template struct EigenBroadcastGrad<Eigen::GpuDevice, dtype::float16, 0>;
template struct EigenBroadcastGrad<Eigen::GpuDevice, double, 0>; template struct EigenBroadcastGrad<Eigen::GpuDevice, double, 0>;
template struct EigenBroadcastGrad<Eigen::GpuDevice, int, 0>; template struct EigenBroadcastGrad<Eigen::GpuDevice, int, 0>;
template struct EigenBroadcastGrad<Eigen::GpuDevice, int64_t, 0>; template struct EigenBroadcastGrad<Eigen::GpuDevice, int64_t, 0>;
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/constant.cc → paddle/pten/kernels/funcs/eigen/constant.cc
浏览文件 @ a1e40dc6
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenConstant<Eigen::DefaultDevice, T, Rank> { struct EigenConstant<Eigen::DefaultDevice, T, Rank> {
...@@ -27,5 +27,5 @@ struct EigenConstant<Eigen::DefaultDevice, T, Rank> { ...@@ -27,5 +27,5 @@ struct EigenConstant<Eigen::DefaultDevice, T, Rank> {
template struct EigenConstant<Eigen::DefaultDevice, float, 1>; template struct EigenConstant<Eigen::DefaultDevice, float, 1>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/constant.cu → paddle/pten/kernels/funcs/eigen/constant.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenConstant<Eigen::GpuDevice, T, Rank> { struct EigenConstant<Eigen::GpuDevice, T, Rank> {
...@@ -27,5 +27,5 @@ struct EigenConstant<Eigen::GpuDevice, T, Rank> { ...@@ -27,5 +27,5 @@ struct EigenConstant<Eigen::GpuDevice, T, Rank> {
template struct EigenConstant<Eigen::GpuDevice, float, 1>; template struct EigenConstant<Eigen::GpuDevice, float, 1>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/pten/kernels/funcs/eigen/eigen_function.h 0 → 100644
浏览文件 @ a1e40dc6
/* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
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. */
#pragma once
#ifndef _USE_MATH_DEFINES
#define _USE_MATH_DEFINES
#endif
#ifndef NOMINMAX
#define NOMINMAX
#endif
#include "unsupported/Eigen/CXX11/Tensor"
namespace pten {
namespace funcs {
template <typename EigenDevice, typename T, int Rank>
struct EigenBroadcast {
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev,
OutType out,
InType in,
const Array& bcast);
static void Eval(const EigenDevice& dev,
OutType32BitIndex out,
InType32BitIndex in,
const Array& bcast);
};
template <typename EigenDevice, typename T, int Rank>
struct EigenBroadcastGrad {
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>;
using Array2 = Eigen::DSizes<Eigen::DenseIndex, Rank * 2>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
InType in,
const Array& reduce_dims,
const Array2& reshape_dims);
};
template <typename EigenDevice, typename T, int Rank>
struct EigenConstant {
using Type = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, Type out, const T value);
};
template <typename EigenDevice, typename T>
struct EigenSign {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T, int Rank>
struct EigenReverse {
using Array = Eigen::DSizes<bool, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& in,
const Array& reverse);
};
template <typename EigenDevice, typename T>
struct EigenAdd {
using InType = Eigen::TensorMap<Eigen::TensorFixedSize<const T,
Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& in,
const T value);
};
template <typename EigenDevice, typename T>
struct EigenSub {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& left,
const InType& right);
};
template <typename EigenDevice, typename T, int Rank>
struct EigenSlice {
using Array = Eigen::DSizes<Eigen::DenseIndex, Rank>;
using Array32Bit = Eigen::DSizes<int, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& in,
const Array& offsets,
const Array& extents);
static void Eval(const EigenDevice& dev,
OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& offsets,
const Array32Bit& extents);
};
template <typename EigenDevice, typename T, int Rank>
struct EigenPad {
using Array = std::array<std::pair<int64_t, int64_t>, Rank>;
using Array32Bit = std::array<std::pair<int, int>, Rank>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using InType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<const T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType32BitIndex =
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& in,
const Array& padding,
const T value);
static void Eval(const EigenDevice& dev,
OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& padding,
const T value);
};
template <typename EigenDevice, typename T>
struct EigenScale {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& in,
const T scale,
const T bias,
const bool bias_after_scale);
};
template <typename EigenDevice, typename T>
struct EigenErf {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T>
struct EigenErfGrad {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType din,
const InType& in,
const InType& dout);
};
template <typename EigenDevice, typename T>
struct EigenRankLoss {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& label,
const InType& left,
const InType& right);
};
template <typename EigenDevice, typename T>
struct EigenRankLossGrad {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void EvalLeft(const EigenDevice& dev,
OutType dleft,
const InType& dout,
const InType& label,
const InType& left,
const InType& right);
static void EvalRight(const EigenDevice& dev,
OutType dright,
const InType& dout,
const InType& label,
const InType& left,
const InType& right);
};
template <typename EigenDevice, typename T>
struct EigenLogLoss {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType out,
const InType& pred,
const InType& label,
const T& epsilon);
};
template <typename EigenDevice, typename T>
struct EigenLogLossGrad {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label,
const T& epsilon);
};
template <typename EigenDevice, typename T>
struct EigenHingeLoss {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType loss,
const InType& pred,
const InType& label);
};
template <typename EigenDevice, typename T>
struct EigenHingeLossGrad {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label);
};
template <typename EigenDevice, typename T>
struct EigenL1Norm {
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev, OutType out, const InType& in);
};
template <typename EigenDevice, typename T>
struct EigenL1NormGrad {
using Array = Eigen::DSizes<Eigen::DenseIndex, 1>;
using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const EigenDevice& dev,
OutType din,
const InType& dout,
const InType& in,
const Array& bcast);
};
} // namespace funcs
} // namespace pten
paddle/fluid/operators/eigen/elementwise.cc → paddle/pten/kernels/funcs/eigen/elementwise.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,19 +11,25 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,19 +11,25 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenAdd<Eigen::DefaultDevice, T> { struct EigenAdd<Eigen::DefaultDevice, T> {
using InType = Eigen::TensorMap<Eigen::TensorFixedSize< using InType = Eigen::TensorMap<Eigen::TensorFixedSize<const T,
const T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Sizes<>,
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize< Eigen::RowMajor,
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
const InType& in, const T value) { Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev,
OutType out,
const InType& in,
const T value) {
out.device(dev) = in + value; out.device(dev) = in + value;
} }
}; };
...@@ -39,13 +45,15 @@ struct EigenSub<Eigen::DefaultDevice, T> { ...@@ -39,13 +45,15 @@ struct EigenSub<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& left, const InType& right) { OutType out,
const InType& left,
const InType& right) {
out.device(dev) = left - right; out.device(dev) = left - right;
} }
}; };
template struct EigenSub<Eigen::DefaultDevice, float>; template struct EigenSub<Eigen::DefaultDevice, float>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/elementwise.cu → paddle/pten/kernels/funcs/eigen/elementwise.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,18 +11,24 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,18 +11,24 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenAdd<Eigen::GpuDevice, T> { struct EigenAdd<Eigen::GpuDevice, T> {
using InType = Eigen::TensorMap<Eigen::TensorFixedSize< using InType = Eigen::TensorMap<Eigen::TensorFixedSize<const T,
const T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Sizes<>,
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize< Eigen::RowMajor,
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in, using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev,
OutType out,
const InType& in,
const T value) { const T value) {
out.device(dev) = in + value; out.device(dev) = in + value;
} }
...@@ -39,7 +45,9 @@ struct EigenSub<Eigen::GpuDevice, T> { ...@@ -39,7 +45,9 @@ struct EigenSub<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& left, static void Eval(const Eigen::GpuDevice& dev,
OutType out,
const InType& left,
const InType& right) { const InType& right) {
out.device(dev) = left - right; out.device(dev) = left - right;
} }
...@@ -47,5 +55,5 @@ struct EigenSub<Eigen::GpuDevice, T> { ...@@ -47,5 +55,5 @@ struct EigenSub<Eigen::GpuDevice, T> {
template struct EigenSub<Eigen::GpuDevice, float>; template struct EigenSub<Eigen::GpuDevice, float>;
} // namespace operators } // namespace fucns
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/erf.cc → paddle/pten/kernels/funcs/eigen/erf.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/eigen_ext.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenErf<Eigen::DefaultDevice, T> { struct EigenErf<Eigen::DefaultDevice, T> {
...@@ -24,7 +24,8 @@ struct EigenErf<Eigen::DefaultDevice, T> { ...@@ -24,7 +24,8 @@ struct EigenErf<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
OutType out,
const InType& in) { const InType& in) {
out.device(dev) = in.erf(); out.device(dev) = in.erf();
} }
...@@ -36,8 +37,10 @@ struct EigenErfGrad<Eigen::DefaultDevice, T> { ...@@ -36,8 +37,10 @@ struct EigenErfGrad<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType din, static void Eval(const Eigen::DefaultDevice& dev,
const InType& in, const InType& dout) { OutType din,
const InType& in,
const InType& dout) {
din.device(dev) = din.device(dev) =
dout * static_cast<T>(M_2_SQRTPI) * (-(in.square())).exp(); dout * static_cast<T>(M_2_SQRTPI) * (-(in.square())).exp();
} }
...@@ -46,10 +49,10 @@ struct EigenErfGrad<Eigen::DefaultDevice, T> { ...@@ -46,10 +49,10 @@ struct EigenErfGrad<Eigen::DefaultDevice, T> {
#define INSTANTIATION(FUNCTOR) \ #define INSTANTIATION(FUNCTOR) \
template struct FUNCTOR<Eigen::DefaultDevice, float>; \ template struct FUNCTOR<Eigen::DefaultDevice, float>; \
template struct FUNCTOR<Eigen::DefaultDevice, double>; \ template struct FUNCTOR<Eigen::DefaultDevice, double>; \
template struct FUNCTOR<Eigen::DefaultDevice, platform::float16> template struct FUNCTOR<Eigen::DefaultDevice, dtype::float16>
INSTANTIATION(EigenErf); INSTANTIATION(EigenErf);
INSTANTIATION(EigenErfGrad); INSTANTIATION(EigenErfGrad);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/erf.cu → paddle/pten/kernels/funcs/eigen/erf.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -14,12 +14,12 @@ limitations under the License. */ ...@@ -14,12 +14,12 @@ limitations under the License. */
#ifndef _USE_MATH_DEFINES #ifndef _USE_MATH_DEFINES
#define _USE_MATH_DEFINES #define _USE_MATH_DEFINES
#endif #endif
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/eigen_ext.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenErf<Eigen::GpuDevice, T> { struct EigenErf<Eigen::GpuDevice, T> {
...@@ -38,7 +38,9 @@ struct EigenErfGrad<Eigen::GpuDevice, T> { ...@@ -38,7 +38,9 @@ struct EigenErfGrad<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType din, const InType& in, static void Eval(const Eigen::GpuDevice& dev,
OutType din,
const InType& in,
const InType& dout) { const InType& dout) {
din.device(dev) = din.device(dev) =
dout * static_cast<T>(M_2_SQRTPI) * (-(in.square())).exp(); dout * static_cast<T>(M_2_SQRTPI) * (-(in.square())).exp();
...@@ -48,10 +50,10 @@ struct EigenErfGrad<Eigen::GpuDevice, T> { ...@@ -48,10 +50,10 @@ struct EigenErfGrad<Eigen::GpuDevice, T> {
#define INSTANTIATION(FUNCTOR) \ #define INSTANTIATION(FUNCTOR) \
template struct FUNCTOR<Eigen::GpuDevice, float>; \ template struct FUNCTOR<Eigen::GpuDevice, float>; \
template struct FUNCTOR<Eigen::GpuDevice, double>; \ template struct FUNCTOR<Eigen::GpuDevice, double>; \
template struct FUNCTOR<Eigen::GpuDevice, platform::float16> template struct FUNCTOR<Eigen::GpuDevice, dtype::float16>
INSTANTIATION(EigenErf); INSTANTIATION(EigenErf);
INSTANTIATION(EigenErfGrad); INSTANTIATION(EigenErfGrad);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/platform/eigen_ext.h → paddle/pten/kernels/funcs/eigen/extensions.h
浏览文件 @ a1e40dc6
// Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. // Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
...@@ -14,25 +14,25 @@ ...@@ -14,25 +14,25 @@
#pragma once #pragma once
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/pten/common/bfloat16.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/common/float16.h"
#include "paddle/pten/core/hostdevice.h" #include "paddle/pten/core/hostdevice.h"
#include "unsupported/Eigen/CXX11/Tensor" #include "unsupported/Eigen/CXX11/Tensor"
namespace Eigen { namespace Eigen {
using float16 = paddle::platform::float16; using float16 = pten::dtype::float16;
template <typename T> template <typename T>
using complex = paddle::platform::complex<T>; using complex = pten::dtype::complex<T>;
template <typename T> template <typename T>
struct NumTraits; struct NumTraits;
template <> template <>
struct NumTraits<paddle::platform::bfloat16> struct NumTraits<pten::dtype::bfloat16>
: GenericNumTraits<paddle::platform::bfloat16> { : GenericNumTraits<pten::dtype::bfloat16> {
enum { enum {
IsSigned = true, IsSigned = true,
IsInteger = false, IsInteger = false,
...@@ -40,23 +40,23 @@ struct NumTraits<paddle::platform::bfloat16> ...@@ -40,23 +40,23 @@ struct NumTraits<paddle::platform::bfloat16>
RequireInitialization = false RequireInitialization = false
}; };
HOSTDEVICE static inline paddle::platform::bfloat16 epsilon() { HOSTDEVICE static inline pten::dtype::bfloat16 epsilon() {
return paddle::platform::raw_uint16_to_bfloat16(0x3400); return pten::dtype::raw_uint16_to_bfloat16(0x3400);
} }
HOSTDEVICE static inline paddle::platform::bfloat16 dummy_precision() { HOSTDEVICE static inline pten::dtype::bfloat16 dummy_precision() {
return paddle::platform::bfloat16(1e-5f); return pten::dtype::bfloat16(1e-5f);
} }
HOSTDEVICE static inline paddle::platform::bfloat16 highest() { HOSTDEVICE static inline pten::dtype::bfloat16 highest() {
return paddle::platform::raw_uint16_to_bfloat16(0x7f7f); return pten::dtype::raw_uint16_to_bfloat16(0x7f7f);
} }
HOSTDEVICE static inline paddle::platform::bfloat16 lowest() { HOSTDEVICE static inline pten::dtype::bfloat16 lowest() {
return paddle::platform::raw_uint16_to_bfloat16(0xff7f); return pten::dtype::raw_uint16_to_bfloat16(0xff7f);
} }
HOSTDEVICE static inline paddle::platform::bfloat16 infinity() { HOSTDEVICE static inline pten::dtype::bfloat16 infinity() {
return paddle::platform::raw_uint16_to_bfloat16(0x7f80); return pten::dtype::raw_uint16_to_bfloat16(0x7f80);
} }
HOSTDEVICE static inline paddle::platform::bfloat16 quiet_NaN() { HOSTDEVICE static inline pten::dtype::bfloat16 quiet_NaN() {
return paddle::platform::raw_uint16_to_bfloat16(0xffc1); return pten::dtype::raw_uint16_to_bfloat16(0xffc1);
} }
}; };
...@@ -114,20 +114,20 @@ struct NumTraits<float16> : GenericNumTraits<float16> { ...@@ -114,20 +114,20 @@ struct NumTraits<float16> : GenericNumTraits<float16> {
}; };
HOSTDEVICE static inline float16 epsilon() { HOSTDEVICE static inline float16 epsilon() {
return paddle::platform::raw_uint16_to_float16(0x0800); return pten::dtype::raw_uint16_to_float16(0x0800);
} }
HOSTDEVICE static inline float16 dummy_precision() { return float16(1e-2f); } HOSTDEVICE static inline float16 dummy_precision() { return float16(1e-2f); }
HOSTDEVICE static inline float16 highest() { HOSTDEVICE static inline float16 highest() {
return paddle::platform::raw_uint16_to_float16(0x7bff); return pten::dtype::raw_uint16_to_float16(0x7bff);
} }
HOSTDEVICE static inline float16 lowest() { HOSTDEVICE static inline float16 lowest() {
return paddle::platform::raw_uint16_to_float16(0xfbff); return pten::dtype::raw_uint16_to_float16(0xfbff);
} }
HOSTDEVICE static inline float16 infinity() { HOSTDEVICE static inline float16 infinity() {
return paddle::platform::raw_uint16_to_float16(0x7c00); return pten::dtype::raw_uint16_to_float16(0x7c00);
} }
HOSTDEVICE static inline float16 quiet_NaN() { HOSTDEVICE static inline float16 quiet_NaN() {
return paddle::platform::raw_uint16_to_float16(0x7c01); return pten::dtype::raw_uint16_to_float16(0x7c01);
} }
}; };
...@@ -136,96 +136,86 @@ namespace numext { ...@@ -136,96 +136,86 @@ namespace numext {
//////////// bfloat methods ///////////// //////////// bfloat methods /////////////
template <> template <>
HOSTDEVICE inline bool(isnan)(const paddle::platform::bfloat16& a) { HOSTDEVICE inline bool(isnan)(const pten::dtype::bfloat16& a) {
return (paddle::platform::isnan)(a); return (pten::dtype::isnan)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isinf)(const paddle::platform::bfloat16& a) { HOSTDEVICE inline bool(isinf)(const pten::dtype::bfloat16& a) {
return (paddle::platform::isinf)(a); return (pten::dtype::isinf)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isfinite)(const paddle::platform::bfloat16& a) { HOSTDEVICE inline bool(isfinite)(const pten::dtype::bfloat16& a) {
return (paddle::platform::isfinite)(a); return (pten::dtype::isfinite)(a);
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 exp( HOSTDEVICE inline pten::dtype::bfloat16 exp(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::expf(static_cast<float>(a)));
return paddle::platform::bfloat16(::expf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 expm1( HOSTDEVICE inline pten::dtype::bfloat16 expm1(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::expm1f(static_cast<float>(a)));
return paddle::platform::bfloat16(::expm1f(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 erf( HOSTDEVICE inline pten::dtype::bfloat16 erf(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::erff(static_cast<float>(a)));
return paddle::platform::bfloat16(::erff(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 log( HOSTDEVICE inline pten::dtype::bfloat16 log(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::logf(static_cast<float>(a)));
return paddle::platform::bfloat16(::logf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 tanh( HOSTDEVICE inline pten::dtype::bfloat16 tanh(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::tanhf(static_cast<float>(a)));
return paddle::platform::bfloat16(::tanhf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 sqrt( HOSTDEVICE inline pten::dtype::bfloat16 sqrt(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::sqrtf(static_cast<float>(a)));
return paddle::platform::bfloat16(::sqrtf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 ceil( HOSTDEVICE inline pten::dtype::bfloat16 ceil(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::ceilf(static_cast<float>(a)));
return paddle::platform::bfloat16(::ceilf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 floor( HOSTDEVICE inline pten::dtype::bfloat16 floor(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::floorf(static_cast<float>(a)));
return paddle::platform::bfloat16(::floorf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 round( HOSTDEVICE inline pten::dtype::bfloat16 round(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::roundf(static_cast<float>(a)));
return paddle::platform::bfloat16(::roundf(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 pow( HOSTDEVICE inline pten::dtype::bfloat16 pow(const pten::dtype::bfloat16& a,
const paddle::platform::bfloat16& a, const paddle::platform::bfloat16& b) { const pten::dtype::bfloat16& b) {
return paddle::platform::bfloat16( return pten::dtype::bfloat16(
::powf(static_cast<float>(a), static_cast<float>(b))); ::powf(static_cast<float>(a), static_cast<float>(b)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 abs( HOSTDEVICE inline pten::dtype::bfloat16 abs(const pten::dtype::bfloat16& a) {
const paddle::platform::bfloat16& a) { return pten::dtype::bfloat16(::fabs(static_cast<float>(a)));
return paddle::platform::bfloat16(::fabs(static_cast<float>(a)));
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 mini( HOSTDEVICE inline pten::dtype::bfloat16 mini(const pten::dtype::bfloat16& a,
const paddle::platform::bfloat16& a, const paddle::platform::bfloat16& b) { const pten::dtype::bfloat16& b) {
return b < a ? b : a; return b < a ? b : a;
} }
template <> template <>
HOSTDEVICE inline paddle::platform::bfloat16 maxi( HOSTDEVICE inline pten::dtype::bfloat16 maxi(const pten::dtype::bfloat16& a,
const paddle::platform::bfloat16& a, const paddle::platform::bfloat16& b) { const pten::dtype::bfloat16& b) {
return a < b ? b : a; return a < b ? b : a;
} }
...@@ -233,17 +223,17 @@ HOSTDEVICE inline paddle::platform::bfloat16 maxi( ...@@ -233,17 +223,17 @@ HOSTDEVICE inline paddle::platform::bfloat16 maxi(
template <> template <>
HOSTDEVICE inline bool(isnan)(const complex<float>& a) { HOSTDEVICE inline bool(isnan)(const complex<float>& a) {
return (paddle::platform::isnan)(a); return (pten::dtype::isnan)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isinf)(const complex<float>& a) { HOSTDEVICE inline bool(isinf)(const complex<float>& a) {
return (paddle::platform::isinf)(a); return (pten::dtype::isinf)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isfinite)(const complex<float>& a) { HOSTDEVICE inline bool(isfinite)(const complex<float>& a) {
return (paddle::platform::isfinite)(a); return (pten::dtype::isfinite)(a);
} }
template <> template <>
...@@ -256,17 +246,17 @@ HOSTDEVICE inline complex<float> exp(const complex<float>& a) { ...@@ -256,17 +246,17 @@ HOSTDEVICE inline complex<float> exp(const complex<float>& a) {
template <> template <>
HOSTDEVICE inline complex<float> log(const complex<float>& a) { HOSTDEVICE inline complex<float> log(const complex<float>& a) {
return paddle::platform::log(a); return pten::dtype::log(a);
} }
template <> template <>
HOSTDEVICE inline complex<float> tanh(const complex<float>& a) { HOSTDEVICE inline complex<float> tanh(const complex<float>& a) {
return paddle::platform::tanh(a); return pten::dtype::tanh(a);
} }
template <> template <>
HOSTDEVICE inline complex<float> sqrt(const complex<float>& a) { HOSTDEVICE inline complex<float> sqrt(const complex<float>& a) {
return paddle::platform::sqrt(a); return pten::dtype::sqrt(a);
} }
template <> template <>
...@@ -287,29 +277,29 @@ HOSTDEVICE inline complex<float> round(const complex<float>& a) { ...@@ -287,29 +277,29 @@ HOSTDEVICE inline complex<float> round(const complex<float>& a) {
template <> template <>
HOSTDEVICE inline complex<float> pow(const complex<float>& a, HOSTDEVICE inline complex<float> pow(const complex<float>& a,
const complex<float>& b) { const complex<float>& b) {
return paddle::platform::pow(a, b); return pten::dtype::pow(a, b);
} }
template <> template <>
HOSTDEVICE inline float abs(const complex<float>& a) { HOSTDEVICE inline float abs(const complex<float>& a) {
return paddle::platform::abs(a); return pten::dtype::abs(a);
} }
//////////// complex<double> methods ///////////// //////////// complex<double> methods /////////////
template <> template <>
HOSTDEVICE inline bool(isnan)(const complex<double>& a) { HOSTDEVICE inline bool(isnan)(const complex<double>& a) {
return (paddle::platform::isnan)(a); return (pten::dtype::isnan)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isinf)(const complex<double>& a) { HOSTDEVICE inline bool(isinf)(const complex<double>& a) {
return (paddle::platform::isinf)(a); return (pten::dtype::isinf)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isfinite)(const complex<double>& a) { HOSTDEVICE inline bool(isfinite)(const complex<double>& a) {
return (paddle::platform::isfinite)(a); return (pten::dtype::isfinite)(a);
} }
template <> template <>
...@@ -322,17 +312,17 @@ HOSTDEVICE inline complex<double> exp(const complex<double>& a) { ...@@ -322,17 +312,17 @@ HOSTDEVICE inline complex<double> exp(const complex<double>& a) {
template <> template <>
HOSTDEVICE inline complex<double> log(const complex<double>& a) { HOSTDEVICE inline complex<double> log(const complex<double>& a) {
return paddle::platform::log(a); return pten::dtype::log(a);
} }
template <> template <>
HOSTDEVICE inline complex<double> tanh(const complex<double>& a) { HOSTDEVICE inline complex<double> tanh(const complex<double>& a) {
return paddle::platform::tanh(a); return pten::dtype::tanh(a);
} }
template <> template <>
HOSTDEVICE inline complex<double> sqrt(const complex<double>& a) { HOSTDEVICE inline complex<double> sqrt(const complex<double>& a) {
return paddle::platform::sqrt(a); return pten::dtype::sqrt(a);
} }
template <> template <>
...@@ -353,29 +343,29 @@ HOSTDEVICE inline complex<double> round(const complex<double>& a) { ...@@ -353,29 +343,29 @@ HOSTDEVICE inline complex<double> round(const complex<double>& a) {
template <> template <>
HOSTDEVICE inline complex<double> pow(const complex<double>& a, HOSTDEVICE inline complex<double> pow(const complex<double>& a,
const complex<double>& b) { const complex<double>& b) {
return paddle::platform::pow(a, b); return pten::dtype::pow(a, b);
} }
template <> template <>
HOSTDEVICE inline double abs(const complex<double>& a) { HOSTDEVICE inline double abs(const complex<double>& a) {
return paddle::platform::abs(a); return pten::dtype::abs(a);
} }
//////////// float16 methods ///////////// //////////// float16 methods /////////////
template <> template <>
HOSTDEVICE inline bool(isnan)(const float16& a) { HOSTDEVICE inline bool(isnan)(const float16& a) {
return (paddle::platform::isnan)(a); return (pten::dtype::isnan)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isinf)(const float16& a) { HOSTDEVICE inline bool(isinf)(const float16& a) {
return (paddle::platform::isinf)(a); return (pten::dtype::isinf)(a);
} }
template <> template <>
HOSTDEVICE inline bool(isfinite)(const float16& a) { HOSTDEVICE inline bool(isfinite)(const float16& a) {
return (paddle::platform::isfinite)(a); return (pten::dtype::isfinite)(a);
} }
template <> template <>
......
paddle/fluid/operators/eigen/l1_norm.cc → paddle/pten/kernels/funcs/eigen/l1_norm.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,18 +11,21 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,18 +11,21 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenL1Norm<Eigen::DefaultDevice, T> { struct EigenL1Norm<Eigen::DefaultDevice, T> {
using InType = Eigen::TensorMap< using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize< using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Sizes<>,
static void Eval(const Eigen::DefaultDevice& dev, OutType out, Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev,
OutType out,
const InType& in) { const InType& in) {
out.device(dev) = in.abs().sum(); out.device(dev) = in.abs().sum();
} }
...@@ -35,8 +38,11 @@ struct EigenL1NormGrad<Eigen::DefaultDevice, T> { ...@@ -35,8 +38,11 @@ struct EigenL1NormGrad<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType din, static void Eval(const Eigen::DefaultDevice& dev,
const InType& dout, const InType& in, const Array& bcast) { OutType din,
const InType& dout,
const InType& in,
const Array& bcast) {
din.device(dev) = dout.broadcast(bcast) * in.sign(); din.device(dev) = dout.broadcast(bcast) * in.sign();
} }
}; };
...@@ -44,5 +50,5 @@ struct EigenL1NormGrad<Eigen::DefaultDevice, T> { ...@@ -44,5 +50,5 @@ struct EigenL1NormGrad<Eigen::DefaultDevice, T> {
template struct EigenL1Norm<Eigen::DefaultDevice, float>; template struct EigenL1Norm<Eigen::DefaultDevice, float>;
template struct EigenL1NormGrad<Eigen::DefaultDevice, float>; template struct EigenL1NormGrad<Eigen::DefaultDevice, float>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/l1_norm.cu → paddle/pten/kernels/funcs/eigen/l1_norm.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,17 +11,19 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,17 +11,19 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenL1Norm<Eigen::GpuDevice, T> { struct EigenL1Norm<Eigen::GpuDevice, T> {
using InType = Eigen::TensorMap< using InType = Eigen::TensorMap<
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap<Eigen::TensorFixedSize< using OutType = Eigen::TensorMap<Eigen::TensorFixedSize<T,
T, Eigen::Sizes<>, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Sizes<>,
Eigen::RowMajor,
Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in) { static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in) {
out.device(dev) = in.abs().sum(); out.device(dev) = in.abs().sum();
} }
...@@ -34,8 +36,11 @@ struct EigenL1NormGrad<Eigen::GpuDevice, T> { ...@@ -34,8 +36,11 @@ struct EigenL1NormGrad<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType din, const InType& dout, static void Eval(const Eigen::GpuDevice& dev,
const InType& in, const Array& bcast) { OutType din,
const InType& dout,
const InType& in,
const Array& bcast) {
din.device(dev) = dout.broadcast(bcast) * in.sign(); din.device(dev) = dout.broadcast(bcast) * in.sign();
} }
}; };
...@@ -43,5 +48,5 @@ struct EigenL1NormGrad<Eigen::GpuDevice, T> { ...@@ -43,5 +48,5 @@ struct EigenL1NormGrad<Eigen::GpuDevice, T> {
template struct EigenL1Norm<Eigen::GpuDevice, float>; template struct EigenL1Norm<Eigen::GpuDevice, float>;
template struct EigenL1NormGrad<Eigen::GpuDevice, float>; template struct EigenL1NormGrad<Eigen::GpuDevice, float>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/loss.cc → paddle/pten/kernels/funcs/eigen/loss.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenRankLoss<Eigen::DefaultDevice, T> { struct EigenRankLoss<Eigen::DefaultDevice, T> {
...@@ -22,8 +22,10 @@ struct EigenRankLoss<Eigen::DefaultDevice, T> { ...@@ -22,8 +22,10 @@ struct EigenRankLoss<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& label, const InType& left, OutType out,
const InType& label,
const InType& left,
const InType& right) { const InType& right) {
out.device(dev) = out.device(dev) =
(1.0f + (left - right).exp()).log() - label * (left - right); (1.0f + (left - right).exp()).log() - label * (left - right);
...@@ -37,15 +39,21 @@ struct EigenRankLossGrad<Eigen::DefaultDevice, T> { ...@@ -37,15 +39,21 @@ struct EigenRankLossGrad<Eigen::DefaultDevice, T> {
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void EvalLeft(const Eigen::DefaultDevice& dev, OutType dleft, static void EvalLeft(const Eigen::DefaultDevice& dev,
const InType& dout, const InType& label, OutType dleft,
const InType& left, const InType& right) { const InType& dout,
const InType& label,
const InType& left,
const InType& right) {
dleft.device(dev) = dout * (1.0f / (1.0f + (right - left).exp()) - label); dleft.device(dev) = dout * (1.0f / (1.0f + (right - left).exp()) - label);
} }
static void EvalRight(const Eigen::DefaultDevice& dev, OutType dright, static void EvalRight(const Eigen::DefaultDevice& dev,
const InType& dout, const InType& label, OutType dright,
const InType& left, const InType& right) { const InType& dout,
const InType& label,
const InType& left,
const InType& right) {
dright.device(dev) = -dout * (1.0f / (1.0f + (right - left).exp()) - label); dright.device(dev) = -dout * (1.0f / (1.0f + (right - left).exp()) - label);
} }
}; };
...@@ -59,8 +67,11 @@ struct EigenLogLoss<Eigen::DefaultDevice, T> { ...@@ -59,8 +67,11 @@ struct EigenLogLoss<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& pred, const InType& label, const T& epsilon) { OutType out,
const InType& pred,
const InType& label,
const T& epsilon) {
out.device(dev) = (-(label * (pred + epsilon).log()) - out.device(dev) = (-(label * (pred + epsilon).log()) -
((static_cast<T>(1) - label) * ((static_cast<T>(1) - label) *
(static_cast<T>(1) - pred + epsilon).log())); (static_cast<T>(1) - pred + epsilon).log()));
...@@ -73,8 +84,11 @@ struct EigenLogLossGrad<Eigen::DefaultDevice, T> { ...@@ -73,8 +84,11 @@ struct EigenLogLossGrad<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType dpred, static void Eval(const Eigen::DefaultDevice& dev,
const InType& dloss, const InType& pred, const InType& label, OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label,
const T& epsilon) { const T& epsilon) {
dpred.device(dev) = dpred.device(dev) =
dloss * dloss *
...@@ -92,8 +106,10 @@ struct EigenHingeLoss<Eigen::DefaultDevice, T> { ...@@ -92,8 +106,10 @@ struct EigenHingeLoss<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType loss, static void Eval(const Eigen::DefaultDevice& dev,
const InType& pred, const InType& label) { OutType loss,
const InType& pred,
const InType& label) {
loss.device(dev) = (static_cast<T>(1) - loss.device(dev) = (static_cast<T>(1) -
pred * (static_cast<T>(2) * label - static_cast<T>(1))) pred * (static_cast<T>(2) * label - static_cast<T>(1)))
.cwiseMax(static_cast<T>(0)); .cwiseMax(static_cast<T>(0));
...@@ -106,8 +122,10 @@ struct EigenHingeLossGrad<Eigen::DefaultDevice, T> { ...@@ -106,8 +122,10 @@ struct EigenHingeLossGrad<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType dpred, static void Eval(const Eigen::DefaultDevice& dev,
const InType& dloss, const InType& pred, OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label) { const InType& label) {
auto alt_labels = static_cast<T>(2) * label - static_cast<T>(1); auto alt_labels = static_cast<T>(2) * label - static_cast<T>(1);
dpred.device(dev) = dpred.device(dev) =
...@@ -119,5 +137,5 @@ struct EigenHingeLossGrad<Eigen::DefaultDevice, T> { ...@@ -119,5 +137,5 @@ struct EigenHingeLossGrad<Eigen::DefaultDevice, T> {
template struct EigenHingeLoss<Eigen::DefaultDevice, float>; template struct EigenHingeLoss<Eigen::DefaultDevice, float>;
template struct EigenHingeLossGrad<Eigen::DefaultDevice, float>; template struct EigenHingeLossGrad<Eigen::DefaultDevice, float>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/loss.cu → paddle/pten/kernels/funcs/eigen/loss.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenRankLoss<Eigen::GpuDevice, T> { struct EigenRankLoss<Eigen::GpuDevice, T> {
...@@ -22,8 +22,10 @@ struct EigenRankLoss<Eigen::GpuDevice, T> { ...@@ -22,8 +22,10 @@ struct EigenRankLoss<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, static void Eval(const Eigen::GpuDevice& dev,
const InType& label, const InType& left, OutType out,
const InType& label,
const InType& left,
const InType& right) { const InType& right) {
out.device(dev) = out.device(dev) =
(1.0f + (left - right).exp()).log() - label * (left - right); (1.0f + (left - right).exp()).log() - label * (left - right);
...@@ -37,15 +39,21 @@ struct EigenRankLossGrad<Eigen::GpuDevice, T> { ...@@ -37,15 +39,21 @@ struct EigenRankLossGrad<Eigen::GpuDevice, T> {
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void EvalLeft(const Eigen::GpuDevice& dev, OutType dleft, static void EvalLeft(const Eigen::GpuDevice& dev,
const InType& dout, const InType& label, OutType dleft,
const InType& left, const InType& right) { const InType& dout,
const InType& label,
const InType& left,
const InType& right) {
dleft.device(dev) = dout * (1.0f / (1.0f + (right - left).exp()) - label); dleft.device(dev) = dout * (1.0f / (1.0f + (right - left).exp()) - label);
} }
static void EvalRight(const Eigen::GpuDevice& dev, OutType dright, static void EvalRight(const Eigen::GpuDevice& dev,
const InType& dout, const InType& label, OutType dright,
const InType& left, const InType& right) { const InType& dout,
const InType& label,
const InType& left,
const InType& right) {
dright.device(dev) = -dout * (1.0f / (1.0f + (right - left).exp()) - label); dright.device(dev) = -dout * (1.0f / (1.0f + (right - left).exp()) - label);
} }
}; };
...@@ -59,8 +67,11 @@ struct EigenLogLoss<Eigen::GpuDevice, T> { ...@@ -59,8 +67,11 @@ struct EigenLogLoss<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& pred, static void Eval(const Eigen::GpuDevice& dev,
const InType& label, const T& epsilon) { OutType out,
const InType& pred,
const InType& label,
const T& epsilon) {
out.device(dev) = (-(label * (pred + epsilon).log()) - out.device(dev) = (-(label * (pred + epsilon).log()) -
((static_cast<T>(1) - label) * ((static_cast<T>(1) - label) *
(static_cast<T>(1) - pred + epsilon).log())); (static_cast<T>(1) - pred + epsilon).log()));
...@@ -73,8 +84,11 @@ struct EigenLogLossGrad<Eigen::GpuDevice, T> { ...@@ -73,8 +84,11 @@ struct EigenLogLossGrad<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType dpred, static void Eval(const Eigen::GpuDevice& dev,
const InType& dloss, const InType& pred, const InType& label, OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label,
const T& epsilon) { const T& epsilon) {
dpred.device(dev) = dpred.device(dev) =
dloss * dloss *
...@@ -92,8 +106,10 @@ struct EigenHingeLoss<Eigen::GpuDevice, T> { ...@@ -92,8 +106,10 @@ struct EigenHingeLoss<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType loss, static void Eval(const Eigen::GpuDevice& dev,
const InType& pred, const InType& label) { OutType loss,
const InType& pred,
const InType& label) {
loss.device(dev) = (static_cast<T>(1) - loss.device(dev) = (static_cast<T>(1) -
pred * (static_cast<T>(2) * label - static_cast<T>(1))) pred * (static_cast<T>(2) * label - static_cast<T>(1)))
.cwiseMax(static_cast<T>(0)); .cwiseMax(static_cast<T>(0));
...@@ -106,8 +122,10 @@ struct EigenHingeLossGrad<Eigen::GpuDevice, T> { ...@@ -106,8 +122,10 @@ struct EigenHingeLossGrad<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType dpred, static void Eval(const Eigen::GpuDevice& dev,
const InType& dloss, const InType& pred, OutType dpred,
const InType& dloss,
const InType& pred,
const InType& label) { const InType& label) {
auto alt_labels = static_cast<T>(2) * label - static_cast<T>(1); auto alt_labels = static_cast<T>(2) * label - static_cast<T>(1);
dpred.device(dev) = dpred.device(dev) =
...@@ -119,5 +137,5 @@ struct EigenHingeLossGrad<Eigen::GpuDevice, T> { ...@@ -119,5 +137,5 @@ struct EigenHingeLossGrad<Eigen::GpuDevice, T> {
template struct EigenHingeLoss<Eigen::GpuDevice, float>; template struct EigenHingeLoss<Eigen::GpuDevice, float>;
template struct EigenHingeLossGrad<Eigen::GpuDevice, float>; template struct EigenHingeLossGrad<Eigen::GpuDevice, float>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/pad.cc → paddle/pten/kernels/funcs/eigen/pad.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,11 +11,11 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenPad<Eigen::DefaultDevice, T, Rank> { struct EigenPad<Eigen::DefaultDevice, T, Rank> {
...@@ -32,13 +32,18 @@ struct EigenPad<Eigen::DefaultDevice, T, Rank> { ...@@ -32,13 +32,18 @@ struct EigenPad<Eigen::DefaultDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& in, const Array& padding, const T value) { OutType out,
const InType& in,
const Array& padding,
const T value) {
out.device(dev) = in.pad(padding, value); out.device(dev) = in.pad(padding, value);
} }
static void Eval(const Eigen::DefaultDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::DefaultDevice& dev,
const InType32BitIndex& in, const Array32Bit& padding, OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& padding,
const T value) { const T value) {
out.device(dev) = in.pad(padding, value); out.device(dev) = in.pad(padding, value);
} }
...@@ -56,9 +61,9 @@ INSTANTIATION(EigenPad, int); ...@@ -56,9 +61,9 @@ INSTANTIATION(EigenPad, int);
INSTANTIATION(EigenPad, int64_t); INSTANTIATION(EigenPad, int64_t);
INSTANTIATION(EigenPad, float); INSTANTIATION(EigenPad, float);
INSTANTIATION(EigenPad, double); INSTANTIATION(EigenPad, double);
INSTANTIATION(EigenPad, platform::complex<float>); INSTANTIATION(EigenPad, dtype::complex<float>);
INSTANTIATION(EigenPad, platform::complex<double>); INSTANTIATION(EigenPad, dtype::complex<double>);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/pad.cu → paddle/pten/kernels/funcs/eigen/pad.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/bfloat16.h"
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenPad<Eigen::GpuDevice, T, Rank> { struct EigenPad<Eigen::GpuDevice, T, Rank> {
...@@ -34,13 +34,18 @@ struct EigenPad<Eigen::GpuDevice, T, Rank> { ...@@ -34,13 +34,18 @@ struct EigenPad<Eigen::GpuDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in, static void Eval(const Eigen::GpuDevice& dev,
const Array& padding, const T value) { OutType out,
const InType& in,
const Array& padding,
const T value) {
out.device(dev) = in.pad(padding, value); out.device(dev) = in.pad(padding, value);
} }
static void Eval(const Eigen::GpuDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::GpuDevice& dev,
const InType32BitIndex& in, const Array32Bit& padding, OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& padding,
const T value) { const T value) {
out.device(dev) = in.pad(padding, value); out.device(dev) = in.pad(padding, value);
} }
...@@ -58,11 +63,11 @@ INSTANTIATION(EigenPad, int); ...@@ -58,11 +63,11 @@ INSTANTIATION(EigenPad, int);
INSTANTIATION(EigenPad, int64_t); INSTANTIATION(EigenPad, int64_t);
INSTANTIATION(EigenPad, float); INSTANTIATION(EigenPad, float);
INSTANTIATION(EigenPad, double); INSTANTIATION(EigenPad, double);
INSTANTIATION(EigenPad, platform::float16); INSTANTIATION(EigenPad, dtype::float16);
INSTANTIATION(EigenPad, platform::bfloat16); INSTANTIATION(EigenPad, dtype::bfloat16);
INSTANTIATION(EigenPad, platform::complex<float>); INSTANTIATION(EigenPad, dtype::complex<float>);
INSTANTIATION(EigenPad, platform::complex<double>); INSTANTIATION(EigenPad, dtype::complex<double>);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/reverse.cc → paddle/pten/kernels/funcs/eigen/reverse.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenReverse<Eigen::DefaultDevice, T, Rank> { struct EigenReverse<Eigen::DefaultDevice, T, Rank> {
...@@ -23,8 +23,10 @@ struct EigenReverse<Eigen::DefaultDevice, T, Rank> { ...@@ -23,8 +23,10 @@ struct EigenReverse<Eigen::DefaultDevice, T, Rank> {
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap< using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& in, const Array& reverse) { OutType out,
const InType& in,
const Array& reverse) {
out.device(dev) = in.reverse(reverse); out.device(dev) = in.reverse(reverse);
} }
}; };
...@@ -44,5 +46,5 @@ INSTANTIATION(EigenReverse, float); ...@@ -44,5 +46,5 @@ INSTANTIATION(EigenReverse, float);
INSTANTIATION(EigenReverse, double); INSTANTIATION(EigenReverse, double);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/reverse.cu → paddle/pten/kernels/funcs/eigen/reverse.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenReverse<Eigen::GpuDevice, T, Rank> { struct EigenReverse<Eigen::GpuDevice, T, Rank> {
...@@ -23,7 +23,9 @@ struct EigenReverse<Eigen::GpuDevice, T, Rank> { ...@@ -23,7 +23,9 @@ struct EigenReverse<Eigen::GpuDevice, T, Rank> {
Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = Eigen::TensorMap< using OutType = Eigen::TensorMap<
Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<T, Rank, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in, static void Eval(const Eigen::GpuDevice& dev,
OutType out,
const InType& in,
const Array& reverse) { const Array& reverse) {
out.device(dev) = in.reverse(reverse); out.device(dev) = in.reverse(reverse);
} }
...@@ -44,5 +46,5 @@ INSTANTIATION(EigenReverse, float); ...@@ -44,5 +46,5 @@ INSTANTIATION(EigenReverse, float);
INSTANTIATION(EigenReverse, double); INSTANTIATION(EigenReverse, double);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/scale.cc → paddle/pten/kernels/funcs/eigen/scale.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/bfloat16.h"
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenScale<Eigen::DefaultDevice, T> { struct EigenScale<Eigen::DefaultDevice, T> {
...@@ -24,8 +24,11 @@ struct EigenScale<Eigen::DefaultDevice, T> { ...@@ -24,8 +24,11 @@ struct EigenScale<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& in, const T scale, const T bias, OutType out,
const InType& in,
const T scale,
const T bias,
const bool bias_after_scale) { const bool bias_after_scale) {
if (bias_after_scale) { if (bias_after_scale) {
out.device(dev) = scale * in + bias; out.device(dev) = scale * in + bias;
...@@ -37,14 +40,14 @@ struct EigenScale<Eigen::DefaultDevice, T> { ...@@ -37,14 +40,14 @@ struct EigenScale<Eigen::DefaultDevice, T> {
template struct EigenScale<Eigen::DefaultDevice, float>; template struct EigenScale<Eigen::DefaultDevice, float>;
template struct EigenScale<Eigen::DefaultDevice, double>; template struct EigenScale<Eigen::DefaultDevice, double>;
template struct EigenScale<Eigen::DefaultDevice, platform::bfloat16>; template struct EigenScale<Eigen::DefaultDevice, dtype::bfloat16>;
template struct EigenScale<Eigen::DefaultDevice, uint8_t>; template struct EigenScale<Eigen::DefaultDevice, uint8_t>;
template struct EigenScale<Eigen::DefaultDevice, int8_t>; template struct EigenScale<Eigen::DefaultDevice, int8_t>;
template struct EigenScale<Eigen::DefaultDevice, int16_t>; template struct EigenScale<Eigen::DefaultDevice, int16_t>;
template struct EigenScale<Eigen::DefaultDevice, int>; template struct EigenScale<Eigen::DefaultDevice, int>;
template struct EigenScale<Eigen::DefaultDevice, int64_t>; template struct EigenScale<Eigen::DefaultDevice, int64_t>;
template struct EigenScale<Eigen::DefaultDevice, platform::complex<float>>; template struct EigenScale<Eigen::DefaultDevice, dtype::complex<float>>;
template struct EigenScale<Eigen::DefaultDevice, platform::complex<double>>; template struct EigenScale<Eigen::DefaultDevice, dtype::complex<double>>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/scale.cu → paddle/pten/kernels/funcs/eigen/scale.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenScale<Eigen::GpuDevice, T> { struct EigenScale<Eigen::GpuDevice, T> {
...@@ -24,8 +24,12 @@ struct EigenScale<Eigen::GpuDevice, T> { ...@@ -24,8 +24,12 @@ struct EigenScale<Eigen::GpuDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in, static void Eval(const Eigen::GpuDevice& dev,
const T scale, const T bias, const bool bias_after_scale) { OutType out,
const InType& in,
const T scale,
const T bias,
const bool bias_after_scale) {
if (bias_after_scale) { if (bias_after_scale) {
out.device(dev) = scale * in + bias; out.device(dev) = scale * in + bias;
} else { } else {
...@@ -41,9 +45,9 @@ template struct EigenScale<Eigen::GpuDevice, int8_t>; ...@@ -41,9 +45,9 @@ template struct EigenScale<Eigen::GpuDevice, int8_t>;
template struct EigenScale<Eigen::GpuDevice, int16_t>; template struct EigenScale<Eigen::GpuDevice, int16_t>;
template struct EigenScale<Eigen::GpuDevice, int>; template struct EigenScale<Eigen::GpuDevice, int>;
template struct EigenScale<Eigen::GpuDevice, int64_t>; template struct EigenScale<Eigen::GpuDevice, int64_t>;
template struct EigenScale<Eigen::GpuDevice, platform::float16>; template struct EigenScale<Eigen::GpuDevice, dtype::float16>;
template struct EigenScale<Eigen::GpuDevice, platform::complex<float>>; template struct EigenScale<Eigen::GpuDevice, dtype::complex<float>>;
template struct EigenScale<Eigen::GpuDevice, platform::complex<double>>; template struct EigenScale<Eigen::GpuDevice, dtype::complex<double>>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/sign.cc → paddle/pten/kernels/funcs/eigen/sign.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,10 +11,10 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenSign<Eigen::DefaultDevice, T> { struct EigenSign<Eigen::DefaultDevice, T> {
...@@ -22,7 +22,8 @@ struct EigenSign<Eigen::DefaultDevice, T> { ...@@ -22,7 +22,8 @@ struct EigenSign<Eigen::DefaultDevice, T> {
Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::Tensor<const T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
using OutType = using OutType =
Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>; Eigen::TensorMap<Eigen::Tensor<T, 1, Eigen::RowMajor, Eigen::DenseIndex>>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
OutType out,
const InType& in) { const InType& in) {
out.device(dev) = in.sign(); out.device(dev) = in.sign();
} }
...@@ -31,5 +32,5 @@ struct EigenSign<Eigen::DefaultDevice, T> { ...@@ -31,5 +32,5 @@ struct EigenSign<Eigen::DefaultDevice, T> {
template struct EigenSign<Eigen::DefaultDevice, float>; template struct EigenSign<Eigen::DefaultDevice, float>;
template struct EigenSign<Eigen::DefaultDevice, double>; template struct EigenSign<Eigen::DefaultDevice, double>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/sign.cu → paddle/pten/kernels/funcs/eigen/sign.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,12 +11,12 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/eigen_ext.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T> template <typename T>
struct EigenSign<Eigen::GpuDevice, T> { struct EigenSign<Eigen::GpuDevice, T> {
...@@ -31,7 +31,7 @@ struct EigenSign<Eigen::GpuDevice, T> { ...@@ -31,7 +31,7 @@ struct EigenSign<Eigen::GpuDevice, T> {
template struct EigenSign<Eigen::GpuDevice, float>; template struct EigenSign<Eigen::GpuDevice, float>;
template struct EigenSign<Eigen::GpuDevice, double>; template struct EigenSign<Eigen::GpuDevice, double>;
template struct EigenSign<Eigen::GpuDevice, platform::float16>; template struct EigenSign<Eigen::GpuDevice, dtype::float16>;
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/slice.cc → paddle/pten/kernels/funcs/eigen/slice.cc
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/bfloat16.h"
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenSlice<Eigen::DefaultDevice, T, Rank> { struct EigenSlice<Eigen::DefaultDevice, T, Rank> {
...@@ -34,14 +34,18 @@ struct EigenSlice<Eigen::DefaultDevice, T, Rank> { ...@@ -34,14 +34,18 @@ struct EigenSlice<Eigen::DefaultDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::DefaultDevice& dev, OutType out, static void Eval(const Eigen::DefaultDevice& dev,
const InType& in, const Array& offsets, OutType out,
const InType& in,
const Array& offsets,
const Array& extents) { const Array& extents) {
out.device(dev) = in.slice(offsets, extents); out.device(dev) = in.slice(offsets, extents);
} }
static void Eval(const Eigen::DefaultDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::DefaultDevice& dev,
const InType32BitIndex& in, const Array32Bit& offsets, OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& offsets,
const Array32Bit& extents) { const Array32Bit& extents) {
out.device(dev) = in.slice(offsets, extents); out.device(dev) = in.slice(offsets, extents);
} }
...@@ -65,11 +69,11 @@ INSTANTIATION(EigenSlice, int16_t); ...@@ -65,11 +69,11 @@ INSTANTIATION(EigenSlice, int16_t);
INSTANTIATION(EigenSlice, int64_t); INSTANTIATION(EigenSlice, int64_t);
INSTANTIATION(EigenSlice, float); INSTANTIATION(EigenSlice, float);
INSTANTIATION(EigenSlice, double); INSTANTIATION(EigenSlice, double);
INSTANTIATION(EigenSlice, platform::float16); INSTANTIATION(EigenSlice, dtype::float16);
INSTANTIATION(EigenSlice, platform::bfloat16); INSTANTIATION(EigenSlice, dtype::bfloat16);
INSTANTIATION(EigenSlice, platform::complex<float>); INSTANTIATION(EigenSlice, dtype::complex<float>);
INSTANTIATION(EigenSlice, platform::complex<double>); INSTANTIATION(EigenSlice, dtype::complex<double>);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
paddle/fluid/operators/eigen/slice.cu → paddle/pten/kernels/funcs/eigen/slice.cu
浏览文件 @ a1e40dc6
/* Copyright (c) 2021 PaddlePaddle Authors. All Rights Reserved. /* Copyright (c) 2022 PaddlePaddle Authors. All Rights Reserved.
Licensed under the Apache License, Version 2.0 (the "License"); Licensed under the Apache License, Version 2.0 (the "License");
you may not use this file except in compliance with the License. you may not use this file except in compliance with the License.
...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS, ...@@ -11,13 +11,13 @@ distributed under the License is distributed on an "AS IS" BASIS,
WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
See the License for the specific language governing permissions and See the License for the specific language governing permissions and
limitations under the License. */ limitations under the License. */
#include "paddle/fluid/operators/eigen/eigen_function.h" #include "paddle/pten/common/bfloat16.h"
#include "paddle/fluid/platform/bfloat16.h" #include "paddle/pten/common/complex.h"
#include "paddle/fluid/platform/complex.h" #include "paddle/pten/common/float16.h"
#include "paddle/fluid/platform/float16.h" #include "paddle/pten/kernels/funcs/eigen/eigen_function.h"
namespace paddle { namespace pten {
namespace operators { namespace funcs {
template <typename T, int Rank> template <typename T, int Rank>
struct EigenSlice<Eigen::GpuDevice, T, Rank> { struct EigenSlice<Eigen::GpuDevice, T, Rank> {
...@@ -34,13 +34,18 @@ struct EigenSlice<Eigen::GpuDevice, T, Rank> { ...@@ -34,13 +34,18 @@ struct EigenSlice<Eigen::GpuDevice, T, Rank> {
Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>, Eigen::TensorMap<Eigen::Tensor<T, Rank, Eigen::RowMajor, int>,
Eigen::Aligned>; Eigen::Aligned>;
static void Eval(const Eigen::GpuDevice& dev, OutType out, const InType& in, static void Eval(const Eigen::GpuDevice& dev,
const Array& offsets, const Array& extents) { OutType out,
const InType& in,
const Array& offsets,
const Array& extents) {
out.device(dev) = in.slice(offsets, extents); out.device(dev) = in.slice(offsets, extents);
} }
static void Eval(const Eigen::GpuDevice& dev, OutType32BitIndex out, static void Eval(const Eigen::GpuDevice& dev,
const InType32BitIndex& in, const Array32Bit& offsets, OutType32BitIndex out,
const InType32BitIndex& in,
const Array32Bit& offsets,
const Array32Bit& extents) { const Array32Bit& extents) {
out.device(dev) = in.slice(offsets, extents); out.device(dev) = in.slice(offsets, extents);
} }
...@@ -58,11 +63,11 @@ INSTANTIATION(EigenSlice, int); ...@@ -58,11 +63,11 @@ INSTANTIATION(EigenSlice, int);
INSTANTIATION(EigenSlice, int64_t); INSTANTIATION(EigenSlice, int64_t);
INSTANTIATION(EigenSlice, float); INSTANTIATION(EigenSlice, float);
INSTANTIATION(EigenSlice, double); INSTANTIATION(EigenSlice, double);
INSTANTIATION(EigenSlice, platform::float16); INSTANTIATION(EigenSlice, dtype::float16);
INSTANTIATION(EigenSlice, platform::bfloat16); INSTANTIATION(EigenSlice, dtype::bfloat16);
INSTANTIATION(EigenSlice, platform::complex<float>); INSTANTIATION(EigenSlice, dtype::complex<float>);
INSTANTIATION(EigenSlice, platform::complex<double>); INSTANTIATION(EigenSlice, dtype::complex<double>);
#undef INSTANTIATION #undef INSTANTIATION
} // namespace operators } // namespace funcs
} // namespace paddle } // namespace pten
Markdown is supported
0% 或 .
You are about to add 0 people to the discussion. Proceed with caution.
先完成此消息的编辑！
想要评论请 注册 或