Yu Yang
8 years ago
committed by
GitHub
commit
0e77b31a02
@ -1,100 +1,101 @@
|
||||
#include <thrust/device_vector.h>
|
||||
#include <sstream>
|
||||
|
||||
#include "paddle/framework/dim.h"
|
||||
#include "gtest/gtest.h"
|
||||
#include "paddle/framework/dim.h"
|
||||
|
||||
__global__ void test(paddle::framework::Dim<2>* o) {
|
||||
o[0] = paddle::framework::make_dim(5, 6);
|
||||
o[0] = paddle::framework::make_dim(5, 6);
|
||||
}
|
||||
|
||||
__global__ void dyn_idx_gpu(int* o) {
|
||||
auto d = paddle::framework::make_dim(5, 6);
|
||||
o[0] = d[1];
|
||||
auto d = paddle::framework::make_dim(5, 6);
|
||||
o[0] = d[1];
|
||||
}
|
||||
|
||||
TEST(Dim, Equality) {
|
||||
// construct a Dim on the CPU
|
||||
auto a = paddle::framework::make_dim(3, 4);
|
||||
EXPECT_EQ(paddle::framework::get<0>(a), 3);
|
||||
EXPECT_EQ(paddle::framework::get<1>(a), 4);
|
||||
|
||||
// construct a Dim on the GPU
|
||||
thrust::device_vector<paddle::framework::Dim<2>> t(2);
|
||||
test<<<1,1>>>(thrust::raw_pointer_cast(t.data()));
|
||||
a = t[0];
|
||||
EXPECT_EQ(paddle::framework::get<0>(a), 5);
|
||||
EXPECT_EQ(paddle::framework::get<1>(a), 6);
|
||||
|
||||
// linearization
|
||||
auto b = paddle::framework::make_dim(7, 8);
|
||||
EXPECT_EQ(paddle::framework::linearize(a, b), 83);
|
||||
|
||||
// product
|
||||
EXPECT_EQ(paddle::framework::product(a), 30);
|
||||
|
||||
// mutate a Dim
|
||||
paddle::framework::get<1>(b) = 10;
|
||||
EXPECT_EQ(paddle::framework::get<0>(b), 7);
|
||||
EXPECT_EQ(paddle::framework::get<1>(b), 10);
|
||||
|
||||
// dynamic access
|
||||
paddle::framework::get(b, 0) = 8;
|
||||
b[1] = 11;
|
||||
EXPECT_EQ(paddle::framework::get<0>(b), 8);
|
||||
EXPECT_EQ(paddle::framework::get<1>(b), 11);
|
||||
EXPECT_EQ(paddle::framework::get(b, 0), 8);
|
||||
EXPECT_EQ(b[1], 11);
|
||||
|
||||
// dynamic access on GPU
|
||||
thrust::device_vector<int> r(1);
|
||||
dyn_idx_gpu<<<1,1>>>(thrust::raw_pointer_cast(r.data()));
|
||||
int res = r[0];
|
||||
EXPECT_EQ(res, 6);
|
||||
|
||||
// ex_prefix_mul
|
||||
paddle::framework::Dim<3> c = paddle::framework::ex_prefix_mul(paddle::framework::Dim<3>(3, 4, 5));
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 3);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 12);
|
||||
|
||||
// generate from an index
|
||||
auto size = paddle::framework::make_dim(4, 5, 2);
|
||||
c = paddle::framework::Dim<3>(14, size);
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 2);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 3);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 0);
|
||||
c = paddle::framework::Dim<3>(25, size);
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 1);
|
||||
// construct a Dim on the CPU
|
||||
auto a = paddle::framework::make_dim(3, 4);
|
||||
EXPECT_EQ(paddle::framework::get<0>(a), 3);
|
||||
EXPECT_EQ(paddle::framework::get<1>(a), 4);
|
||||
|
||||
// construct a Dim on the GPU
|
||||
thrust::device_vector<paddle::framework::Dim<2>> t(2);
|
||||
test<<<1, 1>>>(thrust::raw_pointer_cast(t.data()));
|
||||
a = t[0];
|
||||
EXPECT_EQ(paddle::framework::get<0>(a), 5);
|
||||
EXPECT_EQ(paddle::framework::get<1>(a), 6);
|
||||
|
||||
// linearization
|
||||
auto b = paddle::framework::make_dim(7, 8);
|
||||
EXPECT_EQ(paddle::framework::linearize(a, b), 83);
|
||||
|
||||
// product
|
||||
EXPECT_EQ(paddle::framework::product(a), 30);
|
||||
|
||||
// mutate a Dim
|
||||
paddle::framework::get<1>(b) = 10;
|
||||
EXPECT_EQ(paddle::framework::get<0>(b), 7);
|
||||
EXPECT_EQ(paddle::framework::get<1>(b), 10);
|
||||
|
||||
// dynamic access
|
||||
paddle::framework::get(b, 0) = 8;
|
||||
b[1] = 11;
|
||||
EXPECT_EQ(paddle::framework::get<0>(b), 8);
|
||||
EXPECT_EQ(paddle::framework::get<1>(b), 11);
|
||||
EXPECT_EQ(paddle::framework::get(b, 0), 8);
|
||||
EXPECT_EQ(b[1], 11);
|
||||
|
||||
// dynamic access on GPU
|
||||
thrust::device_vector<int> r(1);
|
||||
dyn_idx_gpu<<<1, 1>>>(thrust::raw_pointer_cast(r.data()));
|
||||
int res = r[0];
|
||||
EXPECT_EQ(res, 6);
|
||||
|
||||
// ex_prefix_mul
|
||||
paddle::framework::Dim<3> c =
|
||||
paddle::framework::ex_prefix_mul(paddle::framework::Dim<3>(3, 4, 5));
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 3);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 12);
|
||||
|
||||
// generate from an index
|
||||
auto size = paddle::framework::make_dim(4, 5, 2);
|
||||
c = paddle::framework::Dim<3>(14, size);
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 2);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 3);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 0);
|
||||
c = paddle::framework::Dim<3>(25, size);
|
||||
EXPECT_EQ(paddle::framework::get<0>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<1>(c), 1);
|
||||
EXPECT_EQ(paddle::framework::get<2>(c), 1);
|
||||
}
|
||||
|
||||
TEST(Dim, Bool) {
|
||||
auto a = paddle::framework::make_dim(3, 4);
|
||||
auto b = paddle::framework::make_dim(5, 6);
|
||||
auto c = paddle::framework::make_dim(3, 4);
|
||||
|
||||
// in_bounds check
|
||||
EXPECT_TRUE(paddle::framework::contained(a, b));
|
||||
EXPECT_FALSE(paddle::framework::contained(b, a));
|
||||
|
||||
// comparison
|
||||
EXPECT_TRUE(a == a);
|
||||
EXPECT_FALSE(a == b);
|
||||
EXPECT_TRUE(a == c);
|
||||
auto a = paddle::framework::make_dim(3, 4);
|
||||
auto b = paddle::framework::make_dim(5, 6);
|
||||
auto c = paddle::framework::make_dim(3, 4);
|
||||
|
||||
// in_bounds check
|
||||
EXPECT_TRUE(paddle::framework::contained(a, b));
|
||||
EXPECT_FALSE(paddle::framework::contained(b, a));
|
||||
|
||||
// comparison
|
||||
EXPECT_TRUE(a == a);
|
||||
EXPECT_FALSE(a == b);
|
||||
EXPECT_TRUE(a == c);
|
||||
}
|
||||
|
||||
TEST(Dim, Print) {
|
||||
{
|
||||
std::stringstream ss;
|
||||
auto a = paddle::framework::make_dim(2, 3);
|
||||
ss << a;
|
||||
EXPECT_EQ(ss.str(), "2, 3");
|
||||
}
|
||||
{
|
||||
std::stringstream ss;
|
||||
ss << paddle::framework::make_dim(8);
|
||||
EXPECT_EQ(ss.str(), "8");
|
||||
}
|
||||
{
|
||||
std::stringstream ss;
|
||||
auto a = paddle::framework::make_dim(2, 3);
|
||||
ss << a;
|
||||
EXPECT_EQ(ss.str(), "2, 3");
|
||||
}
|
||||
{
|
||||
std::stringstream ss;
|
||||
ss << paddle::framework::make_dim(8);
|
||||
EXPECT_EQ(ss.str(), "8");
|
||||
}
|
||||
}
|
||||
|
||||
@ -0,0 +1,15 @@
|
||||
/* 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/framework/enforce.h"
|
||||
@ -0,0 +1,19 @@
|
||||
/* 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/framework/tensor.h>
|
||||
|
||||
namespace paddle {
|
||||
namespace framework {}
|
||||
} // namespace paddle
|
||||
@ -1,6 +1,44 @@
|
||||
if(WITH_GPU)
|
||||
nv_library(add_op SRCS add_op.cc add_op.cu DEPS operator op_registry glog ddim)
|
||||
else()
|
||||
cc_library(add_op SRCS add_op.cc DEPS operator op_registry glog ddim)
|
||||
endif()
|
||||
function(op_library TARGET)
|
||||
# op_library is a function to create op library. The interface is same as
|
||||
# cc_library. But it handle split GPU/CPU code and link some common library
|
||||
# for ops.
|
||||
set(cc_srcs)
|
||||
set(cu_srcs)
|
||||
set(op_common_deps operator op_registry)
|
||||
set(options "")
|
||||
set(oneValueArgs "")
|
||||
set(multiValueArgs SRCS DEPS)
|
||||
cmake_parse_arguments(op_library "${options}" "${oneValueArgs}"
|
||||
"${multiValueArgs}" ${ARGN})
|
||||
|
||||
foreach(src ${op_library_SRCS})
|
||||
if (${src} MATCHES ".*\\.cu$")
|
||||
list(APPEND cu_srcs ${src})
|
||||
elseif(${src} MATCHES ".*\\.cc$")
|
||||
list(APPEND cc_srcs ${src})
|
||||
else()
|
||||
message(FATAL_ERROR "${TARGET} Source file ${src} should only be .cc or .cu")
|
||||
endif()
|
||||
endforeach()
|
||||
|
||||
list(LENGTH cc_srcs cc_srcs_len)
|
||||
if (${cc_srcs_len} EQUAL 0)
|
||||
message(FATAL_ERROR "The op library ${TARGET} should contains at least one .cc file")
|
||||
endif()
|
||||
|
||||
list(LENGTH cu_srcs cu_srcs_len)
|
||||
if (${cu_srcs_len} EQUAL 0)
|
||||
message(WARNING "The op library ${TARGET} not support GPU!")
|
||||
endif()
|
||||
|
||||
if (WITH_GPU)
|
||||
nv_library(${TARGET} SRCS ${cc_srcs} ${cu_srcs} DEPS ${op_library_DEPS}
|
||||
${op_common_deps})
|
||||
else()
|
||||
cc_library(${TARGET} SRCS ${cc_srcs} DEPS ${op_library_DEPS}
|
||||
${op_common_deps})
|
||||
endif()
|
||||
endfunction()
|
||||
|
||||
op_library(add_op SRCS add_op.cc add_op.cu)
|
||||
cc_test(add_op_test SRCS add_op_test.cc DEPS add_op)
|
||||
|
||||
Loading…
Reference in new issue