diff --git a/paddle/fluid/operators/optimizers/adam_op.cu b/paddle/fluid/operators/optimizers/adam_op.cu index c7ffb53a0588267ef205971d5899d5aa36168072..3233d8a74365cbb5c142639cf1080d4a8abb8e4f 100644 --- a/paddle/fluid/operators/optimizers/adam_op.cu +++ b/paddle/fluid/operators/optimizers/adam_op.cu @@ -104,7 +104,7 @@ __global__ void SparseAdamCUDAKernelREG( for (; id < ndim; id += blockDim.x * gridDim.x) { auto row_idx = - math::BinarySearch(rows_, row_count, id / row_numel); + pten::funcs::BinarySearch(rows_, row_count, id / row_numel); if (lazy_mode && row_idx < 0) { return; } else { diff --git a/paddle/fluid/operators/optimizers/adam_op.h b/paddle/fluid/operators/optimizers/adam_op.h index bdeaa106282d2550e4fd928627da098168ea469a..20c7424e8617b23f2228fb90045835efeec630d9 100644 --- a/paddle/fluid/operators/optimizers/adam_op.h +++ b/paddle/fluid/operators/optimizers/adam_op.h @@ -21,10 +21,10 @@ limitations under the License. */ #include "paddle/fluid/framework/op_registry.h" #include "paddle/fluid/framework/threadpool.h" #include "paddle/fluid/operators/jit/kernels.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/operators/math/selected_rows_functor.h" #include "paddle/fluid/platform/for_range.h" #include "paddle/fluid/platform/profiler.h" +#include "paddle/pten/kernels/funcs/algorithm.h" namespace paddle { namespace operators { @@ -279,7 +279,7 @@ class SparseAdamFunctor { inline HOSTDEVICE void operator()(size_t i) const { auto row_idx = - math::BinarySearch(rows_, row_count_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_count_, i / row_numel_); if (lazy_mode_ && row_idx < 0) { return; } else { diff --git a/paddle/fluid/operators/optimizers/adamw_op.cu b/paddle/fluid/operators/optimizers/adamw_op.cu index 8bce415cb1ab9835d6c87c9617e9147187b2a2c8..72ec1023b283b8fe35bf9c1916253f2b6f987680 100644 --- a/paddle/fluid/operators/optimizers/adamw_op.cu +++ b/paddle/fluid/operators/optimizers/adamw_op.cu @@ -112,7 +112,7 @@ __global__ void SparseAdamWCUDAKernelREG( for (; id < ndim; id += blockDim.x * gridDim.x) { auto row_idx = - math::BinarySearch(rows_, row_count, id / row_numel); + pten::funcs::BinarySearch(rows_, row_count, id / row_numel); if (lazy_mode && row_idx < 0) { return; } else { diff --git a/paddle/fluid/operators/optimizers/adamw_op.h b/paddle/fluid/operators/optimizers/adamw_op.h index efd3a2b691f72d277965e5436295c2cbc01acc21..7f9d127766c722543e6cea981873481a2aa2367c 100644 --- a/paddle/fluid/operators/optimizers/adamw_op.h +++ b/paddle/fluid/operators/optimizers/adamw_op.h @@ -142,7 +142,7 @@ class SparseAdamWFunctor { inline HOSTDEVICE void operator()(size_t i) const { auto row_idx = - math::BinarySearch(rows_, row_count_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_count_, i / row_numel_); if (lazy_mode_ && row_idx < 0) { return; } else { diff --git a/paddle/fluid/operators/optimizers/lamb_op.h b/paddle/fluid/operators/optimizers/lamb_op.h index f1158703f028b6c2ebbb9e1596b240e81b5b0b2b..4bcacd45ff67132849c86407e62fd044d95d2f0f 100644 --- a/paddle/fluid/operators/optimizers/lamb_op.h +++ b/paddle/fluid/operators/optimizers/lamb_op.h @@ -19,10 +19,10 @@ limitations under the License. */ #include "paddle/fluid/framework/op_registry.h" #include "paddle/fluid/memory/buffer.h" #include "paddle/fluid/operators/amp/fp16_type_traits.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/operators/math/selected_rows_functor.h" #include "paddle/fluid/operators/math/squared_l2_norm.h" #include "paddle/fluid/platform/for_range.h" +#include "paddle/pten/kernels/funcs/algorithm.h" #include "paddle/pten/kernels/funcs/eigen/extensions.h" namespace paddle { @@ -233,7 +233,7 @@ struct SparseLambMomentREGUpdateFunctor { inline HOSTDEVICE void operator()(size_t i) const { if (skip_update_ && *skip_update_) return; auto row_idx = - math::BinarySearch(rows_, row_count_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_count_, i / row_numel_); T g = row_idx >= 0 ? grad_[row_idx * row_numel_ + i % row_numel_] : static_cast(0); update(i, g); @@ -312,7 +312,7 @@ struct SparseLambMomentMENUpdateFunctor { inline HOSTDEVICE void operator()(size_t i) const { if (skip_update_ && *skip_update_) return; auto row_idx = - math::BinarySearch(rows_, row_count_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_count_, i / row_numel_); T g = row_idx >= 0 ? grad_[row_idx * row_numel_ + i % row_numel_] : static_cast(0); update(i, g); diff --git a/paddle/fluid/operators/optimizers/momentum_op.h b/paddle/fluid/operators/optimizers/momentum_op.h index 79d76d52f48c8c2d5f1c62f4cd08977ce268c573..91d8d55c3ff0009a242d2dc32f03d3e168e221d9 100644 --- a/paddle/fluid/operators/optimizers/momentum_op.h +++ b/paddle/fluid/operators/optimizers/momentum_op.h @@ -18,10 +18,10 @@ limitations under the License. */ #include "paddle/fluid/framework/eigen.h" #include "paddle/fluid/framework/op_registry.h" #include "paddle/fluid/operators/amp/fp16_type_traits.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/operators/math/selected_rows_functor.h" #include "paddle/fluid/platform/float16.h" #include "paddle/fluid/platform/for_range.h" +#include "paddle/pten/kernels/funcs/algorithm.h" namespace paddle { namespace operators { @@ -345,7 +345,7 @@ class SparseMomentumFunctor { inline HOSTDEVICE void operator()(size_t i) { auto row_idx = - math::BinarySearch(rows_, row_height_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_height_, i / row_numel_); MT grad = row_idx >= 0 ? static_cast(grad_[row_idx * row_numel_ + i % row_numel_]) * @@ -418,7 +418,7 @@ class SparseMomentumFunctor { inline HOSTDEVICE void operator()(size_t i) { auto row_idx = - math::BinarySearch(rows_, row_height_, i / row_numel_); + pten::funcs::BinarySearch(rows_, row_height_, i / row_numel_); MT grad = row_idx >= 0 ? static_cast(grad_[row_idx * row_numel_ + i % row_numel_]) * diff --git a/paddle/fluid/operators/optimizers/rmsprop_op.h b/paddle/fluid/operators/optimizers/rmsprop_op.h index 27cdbf3df05ab71ee856c214d9c57dc6294b3300..1ecc5b49481600e249230175e756a74eeb4d68a3 100644 --- a/paddle/fluid/operators/optimizers/rmsprop_op.h +++ b/paddle/fluid/operators/optimizers/rmsprop_op.h @@ -16,9 +16,9 @@ limitations under the License. */ #include #include "paddle/fluid/framework/eigen.h" #include "paddle/fluid/framework/op_registry.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/operators/math/selected_rows_functor.h" #include "paddle/fluid/platform/for_range.h" +#include "paddle/pten/kernels/funcs/algorithm.h" namespace paddle { namespace operators { @@ -42,7 +42,8 @@ struct SparseRmspropGradFunctor { row_count_(row_count) {} HOSTDEVICE inline T operator()(int64_t idx) const { - auto row_idx = math::BinarySearch(rows_, row_count_, idx / row_numel_); + auto row_idx = + pten::funcs::BinarySearch(rows_, row_count_, idx / row_numel_); return row_idx >= 0 ? grad_[row_idx * row_numel_ + idx % row_numel_] : 0; } diff --git a/paddle/fluid/operators/optimizers/sparse_momentum_op.h b/paddle/fluid/operators/optimizers/sparse_momentum_op.h index a35fd36a295dfcca4135ac3c3dc9b7216f270a83..e2e5642e83909504da17c4b25ea7418c81adfbbb 100644 --- a/paddle/fluid/operators/optimizers/sparse_momentum_op.h +++ b/paddle/fluid/operators/optimizers/sparse_momentum_op.h @@ -21,7 +21,6 @@ #include "paddle/fluid/framework/eigen.h" #include "paddle/fluid/framework/op_registry.h" #include "paddle/fluid/operators/amp/fp16_type_traits.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/platform/float16.h" #include "paddle/fluid/platform/for_range.h" diff --git a/paddle/fluid/operators/searchsorted_op.h b/paddle/fluid/operators/searchsorted_op.h index f421bdd3cb7120366713aaec60d9ae4482948137..169f98fd7db772c3d6938b1ecc05c498c8343e8c 100644 --- a/paddle/fluid/operators/searchsorted_op.h +++ b/paddle/fluid/operators/searchsorted_op.h @@ -19,9 +19,9 @@ #include "paddle/fluid/framework/ddim.h" #include "paddle/fluid/framework/op_registry.h" #include "paddle/fluid/framework/operator.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/platform/enforce.h" #include "paddle/fluid/platform/for_range.h" +#include "paddle/pten/kernels/funcs/algorithm.h" namespace paddle { namespace operators { @@ -85,11 +85,11 @@ class GpuAndCpuSearchSortedCompute { out_data_[idx] = seq_size_; } else { if (right_) { - out_data_[idx] = static_cast( - math::UpperBound(sequence_ptr, seq_size_, *value_ptr)); + out_data_[idx] = static_cast(pten::funcs::UpperBound( + sequence_ptr, seq_size_, *value_ptr)); } else { - out_data_[idx] = static_cast( - math::LowerBound(sequence_ptr, seq_size_, *value_ptr)); + out_data_[idx] = static_cast(pten::funcs::LowerBound( + sequence_ptr, seq_size_, *value_ptr)); } } } diff --git a/paddle/fluid/operators/sequence_ops/sequence_reverse_op.h b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.h index 2094572a78a52775d371583764a3ad23f7fed190..3b8a802b4001bbef17bc9e2a4c51ead787c8b5e9 100644 --- a/paddle/fluid/operators/sequence_ops/sequence_reverse_op.h +++ b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.h @@ -16,8 +16,8 @@ #include #include "paddle/fluid/framework/op_registry.h" -#include "paddle/fluid/operators/math/algorithm.h" #include "paddle/fluid/platform/for_range.h" +#include "paddle/pten/kernels/funcs/algorithm.h" namespace paddle { namespace operators { @@ -94,7 +94,7 @@ struct SequenceReverseFunctor { HOSTDEVICE void operator()(size_t idx_x) const { auto row_idx_x = idx_x / row_numel_; - auto lod_idx = math::UpperBound(lod_, lod_count_, row_idx_x); + auto lod_idx = pten::funcs::UpperBound(lod_, lod_count_, row_idx_x); auto row_idx_y = lod_[lod_idx - 1] + (lod_[lod_idx] - 1 - row_idx_x); auto idx_y = row_idx_y * row_numel_ + idx_x % row_numel_; y_[idx_y] = x_[idx_x]; diff --git a/paddle/fluid/operators/math/algorithm.h b/paddle/pten/kernels/funcs/algorithm.h similarity index 95% rename from paddle/fluid/operators/math/algorithm.h rename to paddle/pten/kernels/funcs/algorithm.h index cbe1a03d90d850ad5de379041edf844262c7ac79..e88becc0b681b036eba810012392cebf82b151b0 100644 --- a/paddle/fluid/operators/math/algorithm.h +++ b/paddle/pten/kernels/funcs/algorithm.h @@ -20,9 +20,8 @@ #include "paddle/pten/core/hostdevice.h" -namespace paddle { -namespace operators { -namespace math { +namespace pten { +namespace funcs { template HOSTDEVICE inline int64_t BinarySearch(const T *x, int64_t num, const T &val) { @@ -85,6 +84,5 @@ HOSTDEVICE inline size_t UpperBound(const T1 *x, size_t num, const T2 &val) { #endif // @} End Group UpperBound } -} // namespace math -} // namespace operators -} // namespace paddle +} // namespace funcs +} // namespace pten