Merge branch 'develop' of https://github.com/PaddlePaddle/Paddle into mnist
qiaolongfei
8 years ago
commit
d41a551693
@ -0,0 +1,72 @@
|
|||||||
|
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve.
|
||||||
|
|
||||||
|
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. */
|
||||||
|
|
||||||
|
#include "paddle/operators/gather_op.h"
|
||||||
|
#include "paddle/framework/ddim.h"
|
||||||
|
|
||||||
|
namespace paddle {
|
||||||
|
namespace operators {
|
||||||
|
|
||||||
|
class GatherOp : public framework::OperatorWithKernel {
|
||||||
|
public:
|
||||||
|
using framework::OperatorWithKernel::OperatorWithKernel;
|
||||||
|
|
||||||
|
protected:
|
||||||
|
void InferShape(const framework::InferShapeContext &ctx) const override {
|
||||||
|
int batch_size = ctx.Input<Tensor>("Index")->dims()[0];
|
||||||
|
PADDLE_ENFORCE_GE(batch_size, 0, "Batch size must be >0");
|
||||||
|
framework::DDim output_dims(ctx.Input<Tensor>("X")->dims());
|
||||||
|
output_dims[0] = batch_size;
|
||||||
|
ctx.Output<Tensor>("Out")->Resize(output_dims);
|
||||||
|
}
|
||||||
|
};
|
||||||
|
|
||||||
|
class GatherGradOp : public framework::OperatorWithKernel {
|
||||||
|
public:
|
||||||
|
using framework::OperatorWithKernel::OperatorWithKernel;
|
||||||
|
|
||||||
|
protected:
|
||||||
|
void InferShape(const framework::InferShapeContext &ctx) const override {
|
||||||
|
auto X_grad = ctx.Output<Tensor>(framework::GradVarName("X"));
|
||||||
|
auto X = ctx.Input<Tensor>("X");
|
||||||
|
|
||||||
|
X_grad->Resize(X->dims());
|
||||||
|
}
|
||||||
|
};
|
||||||
|
|
||||||
|
class GatherOpMaker : public framework::OpProtoAndCheckerMaker {
|
||||||
|
public:
|
||||||
|
GatherOpMaker(framework::OpProto *proto, framework::OpAttrChecker *op_checker)
|
||||||
|
: OpProtoAndCheckerMaker(proto, op_checker) {
|
||||||
|
AddInput("X", "The source input of gather op");
|
||||||
|
AddInput("Index", "The index input of gather op");
|
||||||
|
AddOutput("Out", "The output of add op");
|
||||||
|
AddComment(R"DOC(
|
||||||
|
Gather Operator by selecting from the first axis,
|
||||||
|
|
||||||
|
Out = X[Index]
|
||||||
|
)DOC");
|
||||||
|
}
|
||||||
|
};
|
||||||
|
} // namespace operators
|
||||||
|
} // namespace paddle
|
||||||
|
|
||||||
|
namespace ops = paddle::operators;
|
||||||
|
REGISTER_OP(gather, ops::GatherOp, ops::GatherOpMaker, gather_grad,
|
||||||
|
ops::GatherGradOp);
|
||||||
|
REGISTER_OP_CPU_KERNEL(gather,
|
||||||
|
ops::GatherOpKernel<paddle::platform::CPUPlace, float>);
|
||||||
|
REGISTER_OP_CPU_KERNEL(
|
||||||
|
gather_grad,
|
||||||
|
ops::GatherGradientOpKernel<paddle::platform::CPUPlace, float>);
|
||||||
@ -0,0 +1,20 @@
|
|||||||
|
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve.
|
||||||
|
|
||||||
|
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. */
|
||||||
|
|
||||||
|
#define EIGEN_USE_GPU
|
||||||
|
#include "paddle/operators/gather_op.h"
|
||||||
|
|
||||||
|
namespace ops = paddle::operators;
|
||||||
|
REGISTER_OP_GPU_KERNEL(gather,
|
||||||
|
ops::GatherOpKernel<paddle::platform::GPUPlace, float>);
|
||||||
@ -0,0 +1,53 @@
|
|||||||
|
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve.
|
||||||
|
|
||||||
|
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
|
||||||
|
#include "gather.h"
|
||||||
|
#include "paddle/framework/eigen.h"
|
||||||
|
#include "paddle/framework/op_registry.h"
|
||||||
|
#include "scatter.h"
|
||||||
|
|
||||||
|
namespace paddle {
|
||||||
|
namespace operators {
|
||||||
|
|
||||||
|
using Tensor = framework::Tensor;
|
||||||
|
|
||||||
|
template <typename Place, typename T>
|
||||||
|
class GatherOpKernel : public framework::OpKernel {
|
||||||
|
public:
|
||||||
|
void Compute(const framework::ExecutionContext &ctx) const override {
|
||||||
|
auto *X = ctx.Input<Tensor>("X");
|
||||||
|
auto *Index = ctx.Input<Tensor>("Index");
|
||||||
|
auto *Y = ctx.Output<Tensor>("Out");
|
||||||
|
|
||||||
|
Y->mutable_data<T>(ctx.GetPlace());
|
||||||
|
Gather<T>(ctx.GetPlace(), X, Index, Y);
|
||||||
|
}
|
||||||
|
};
|
||||||
|
|
||||||
|
template <typename Place, typename T>
|
||||||
|
class GatherGradientOpKernel : public framework::OpKernel {
|
||||||
|
public:
|
||||||
|
void Compute(const framework::ExecutionContext &ctx) const override {
|
||||||
|
auto *Index = ctx.Input<Tensor>("Index");
|
||||||
|
auto *dX = ctx.Output<Tensor>(framework::GradVarName("X"));
|
||||||
|
auto *dO = ctx.Input<Tensor>(framework::GradVarName("Out"));
|
||||||
|
|
||||||
|
dX->mutable_data<T>(ctx.GetPlace());
|
||||||
|
ScatterUpdate<T>(ctx.GetPlace(), dO, Index, dX);
|
||||||
|
}
|
||||||
|
};
|
||||||
|
|
||||||
|
} // namespace operators
|
||||||
|
} // namespace paddle
|
||||||
@ -1,53 +1,65 @@
|
|||||||
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve.
|
/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve.
|
||||||
|
|
||||||
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.
|
||||||
You may obtain a copy of the License at
|
You may obtain a copy of the License at
|
||||||
|
|
||||||
http://www.apache.org/licenses/LICENSE-2.0
|
http://www.apache.org/licenses/LICENSE-2.0
|
||||||
|
|
||||||
Unless required by applicable law or agreed to in writing, software
|
Unless required by applicable law or agreed to in writing, software
|
||||||
distributed under the License is distributed on an "AS IS" BASIS,
|
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 <memory>
|
#include <thrust/device_ptr.h>
|
||||||
#include <random>
|
#include <thrust/iterator/counting_iterator.h>
|
||||||
#include "paddle/platform/dynload/curand.h"
|
#include <thrust/random.h>
|
||||||
#include "paddle/platform/gpu_info.h"
|
#include <thrust/transform.h>
|
||||||
|
|
||||||
#include "paddle/framework/op_registry.h"
|
#include "paddle/framework/op_registry.h"
|
||||||
|
#include "paddle/framework/operator.h"
|
||||||
|
|
||||||
namespace paddle {
|
namespace paddle {
|
||||||
namespace operators {
|
namespace operators {
|
||||||
|
|
||||||
template <typename T>
|
template <typename T>
|
||||||
class GaussianRandomKernel : public framework::OpKernel {
|
struct GaussianGenerator {
|
||||||
|
T mean_, std_;
|
||||||
|
unsigned int seed_;
|
||||||
|
|
||||||
|
__host__ __device__ GaussianGenerator(T mean, T std, int seed)
|
||||||
|
: mean_(mean), std_(std), seed_(seed) {}
|
||||||
|
|
||||||
|
__host__ __device__ T operator()(const unsigned int n) const {
|
||||||
|
thrust::minstd_rand rng;
|
||||||
|
rng.seed(seed_);
|
||||||
|
thrust::normal_distribution<T> dist(mean_, std_);
|
||||||
|
rng.discard(n);
|
||||||
|
return dist(rng);
|
||||||
|
}
|
||||||
|
};
|
||||||
|
|
||||||
|
template <typename T>
|
||||||
|
class GPUGaussianRandomKernel : public framework::OpKernel {
|
||||||
public:
|
public:
|
||||||
void Compute(const framework::ExecutionContext& context) const override {
|
void Compute(const framework::ExecutionContext& context) const override {
|
||||||
float mean = context.op_.GetAttr<float>("mean");
|
auto* tensor = context.Output<framework::Tensor>("Out");
|
||||||
float std = context.op_.GetAttr<float>("std");
|
|
||||||
auto* tensor = context.Output<framework::Tensor>(0);
|
|
||||||
T* data = tensor->mutable_data<T>(context.GetPlace());
|
T* data = tensor->mutable_data<T>(context.GetPlace());
|
||||||
|
unsigned int seed =
|
||||||
int seed = context.op_.GetAttr<int>("seed");
|
static_cast<unsigned int>(context.op_.GetAttr<int>("seed"));
|
||||||
if (seed == 0) {
|
if (seed == 0) {
|
||||||
std::random_device rd;
|
std::random_device rd;
|
||||||
seed = rd();
|
seed = rd();
|
||||||
}
|
}
|
||||||
curandGenerator_t g;
|
T mean = static_cast<T>(context.op_.GetAttr<float>("mean"));
|
||||||
PADDLE_ENFORCE(platform::dynload::curandCreateGenerator(
|
T std = static_cast<T>(context.op_.GetAttr<float>("std"));
|
||||||
&g, CURAND_RNG_PSEUDO_DEFAULT));
|
thrust::counting_iterator<unsigned int> index_sequence_begin(0);
|
||||||
PADDLE_ENFORCE(
|
ssize_t N = framework::product(tensor->dims());
|
||||||
platform::dynload::curandSetPseudoRandomGeneratorSeed(g, seed));
|
thrust::transform(index_sequence_begin, index_sequence_begin + N,
|
||||||
platform::dynload::curandGenerateNormal(
|
thrust::device_ptr<T>(data),
|
||||||
g, data, framework::product(tensor->dims()), mean, std);
|
GaussianGenerator<T>(mean, std, seed));
|
||||||
}
|
}
|
||||||
};
|
};
|
||||||
|
|
||||||
} // namespace operators
|
} // namespace operators
|
||||||
} // namespace paddle
|
} // namespace paddle
|
||||||
|
|
||||||
namespace ops = paddle::operators;
|
REGISTER_OP_GPU_KERNEL(gaussian_random,
|
||||||
REGISTER_OP_GPU_KERNEL(gaussian_random, ops::GaussianRandomKernel<float>);
|
paddle::operators::GPUGaussianRandomKernel<float>);
|
||||||
|
|||||||
Some files were not shown because too many files have changed in this diff Show More
Loading…
Reference in new issue