Skip to content
Snippets Groups Projects
Commit 3629ffc6 authored by Houssem ROUIS's avatar Houssem ROUIS
Browse files

Remove input_nbr attribute for Add operator

parent 1ae97a27
No related branches found
No related tags found
4 merge requests!61v0.4.0,!54Fix the BatchNorm operator,!47v0.4.0,!42[Fix] update backend cuda
Pipeline #57858 passed
...@@ -22,48 +22,27 @@ ...@@ -22,48 +22,27 @@
using namespace Aidge; using namespace Aidge;
TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") { TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") {
std::shared_ptr<Tensor> input1 = std::make_shared<Tensor>(Array4D<float,3,3,3,2> { SECTION("Same input") {
{ // std::shared_ptr<Tensor> input1 = std::make_shared<Tensor>(Array4D<float,3,3,3,2> {
{ // { //
{{20, 47},{21, 48},{22, 49}}, // { //
{{23, 50},{24, 51},{25, 52}}, // {{20, 47},{21, 48},{22, 49}}, //
{{26, 53},{27, 54},{28, 55}} // {{23, 50},{24, 51},{25, 52}}, //
}, // {{26, 53},{27, 54},{28, 55}} //
{ // }, //
{{29, 56},{30, 57},{31, 58}}, // { //
{{32, 59},{33, 60},{34, 61}}, // {{29, 56},{30, 57},{31, 58}}, //
{{35, 62},{36, 63},{37, 64}} // {{32, 59},{33, 60},{34, 61}}, //
}, // {{35, 62},{36, 63},{37, 64}} //
{ // }, //
{{38, 65},{39, 66},{40, 67}}, // { //
{{41, 68},{42, 69},{43, 70}}, // {{38, 65},{39, 66},{40, 67}}, //
{{44, 71},{45, 72},{46, 73}} // {{41, 68},{42, 69},{43, 70}}, //
} // {{44, 71},{45, 72},{46, 73}} //
} // } //
}); // } //
input1->setBackend("cuda"); }); //
SECTION("One input") { input1->setBackend("cuda");
std::shared_ptr<Node> myAdd = Add(1);
auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator());
op->associateInput(0, input1);
op->setBackend("cuda");
op->setDataType(DataType::Float32);
myAdd->forward();
float* computedOutput = new float[input1->size()]();
cudaMemcpy(computedOutput, op->getOutput(0)->getImpl()->rawPtr(), sizeof(float) * input1->size(), cudaMemcpyDeviceToHost);
float* targetOutput = new float[input1->size()]();
cudaMemcpy(targetOutput, input1->getImpl()->rawPtr(), sizeof(float) * input1->size(), cudaMemcpyDeviceToHost);
for(int i = 0; i < input1->size(); i++){
REQUIRE(fabs(computedOutput[i] - targetOutput[i]) < 1e-6);
}
delete[] computedOutput;
delete[] targetOutput;
}
SECTION("Two inputs") {
std::shared_ptr<Tensor> expectedOutput = std::make_shared<Tensor>(Array4D<float,3,3,3,2> { std::shared_ptr<Tensor> expectedOutput = std::make_shared<Tensor>(Array4D<float,3,3,3,2> {
{ {
{ {
...@@ -84,7 +63,7 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") { ...@@ -84,7 +63,7 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") {
} }
}); });
std::shared_ptr<Node> myAdd = Add(2); std::shared_ptr<Node> myAdd = Add();
auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator()); auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator());
op->associateInput(0, input1); op->associateInput(0, input1);
op->associateInput(1, input1); op->associateInput(1, input1);
...@@ -103,47 +82,6 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") { ...@@ -103,47 +82,6 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") {
delete[] computedOutput; delete[] computedOutput;
} }
SECTION("Three inputs") {
std::shared_ptr<Tensor> expectedOutput = std::make_shared<Tensor>(Array4D<float,3,3,3,2> {
{
{
{{ 60, 141},{ 63, 144},{ 66, 147}},
{{ 69, 150},{ 72, 153},{ 75, 156}},
{{ 78, 159},{ 81, 162},{ 84, 165}}
},
{
{{ 87, 168},{ 90, 171},{ 93, 174}},
{{ 96, 177},{ 99, 180},{102, 183}},
{{105, 186},{108, 189},{111, 192}}
},
{
{{114, 195},{117, 198},{120, 201}},
{{123, 204},{126, 207},{129, 210}},
{{132, 213},{135, 216},{138, 219}}
}
}
});
std::shared_ptr<Node> myAdd = Add(3);
auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator());
op->associateInput(0, input1);
op->associateInput(1, input1);
op->associateInput(2, input1);
op->setDataType(DataType::Float32);
op->setBackend("cuda");
myAdd->forward();
float* computedOutput = new float[input1->size()]();
cudaMemcpy(computedOutput, op->getOutput(0)->getImpl()->rawPtr(), sizeof(float) * expectedOutput->size(), cudaMemcpyDeviceToHost);
for(int i = 0; i < expectedOutput->size(); i++){
const float targetOutput = *(static_cast<float*>(expectedOutput->getImpl()->rawPtr()) + i);
REQUIRE(fabs(computedOutput[i] - targetOutput) < 1e-6);
}
delete[] computedOutput;
}
SECTION("Broadcasting") { SECTION("Broadcasting") {
std::shared_ptr<Tensor> input_0 = std::make_shared<Tensor>(Array4D<float,3,1,3,2> { std::shared_ptr<Tensor> input_0 = std::make_shared<Tensor>(Array4D<float,3,1,3,2> {
{ // { //
...@@ -168,47 +106,80 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") { ...@@ -168,47 +106,80 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") {
} // } //
}); // }); //
std::shared_ptr<Tensor> input_2 = std::make_shared<Tensor>(Array1D<float,2> {{100,200}}); std::shared_ptr<Tensor> input_2 = std::make_shared<Tensor>(Array1D<float,2> {{100,200}});
std::shared_ptr<Tensor> expectedOutput = std::make_shared<Tensor>(Array4D<float,3,3,3,2> {
{ //
{ //
{{ 120, 222},{ 124, 226},{ 128, 230}}, //
{{ 126, 228},{ 130, 232},{ 134, 236}}, //
{{ 132, 234},{ 136, 238},{ 140, 242}} //
}, //
{ //
{{ 126, 228},{ 130, 232},{ 134, 236}}, //
{{ 132, 234},{ 136, 238},{ 140, 242}}, //
{{ 138, 240},{ 142, 244},{ 146, 248}} //
}, //
{ //
{{ 132, 234},{ 136, 238},{140, 242}}, //
{{ 138, 240},{ 142, 244},{146, 248}}, //
{{ 144, 246},{ 148, 250},{152, 254}} //
} //
} //
}); //
input_0->setBackend("cuda"); input_0->setBackend("cuda");
input_1->setBackend("cuda"); input_1->setBackend("cuda");
input_2->setBackend("cuda"); input_2->setBackend("cuda");
std::shared_ptr<Node> myAdd = Add(3);
auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator());
op->associateInput(0, input_0);
op->associateInput(1, input_1);
op->associateInput(2, input_2);
op->setDataType(DataType::Float32);
op->setBackend("cuda");
myAdd->forward();
float* computedOutput = new float[input1->size()](); /// Input0(d0, 1, d2, d3) + Input1(1, d1, d2, d3) = Output(d0, d1, d2, d3)
cudaMemcpy(computedOutput, op->getOutput(0)->getImpl()->rawPtr(), sizeof(float) * expectedOutput->size(), cudaMemcpyDeviceToHost); std::shared_ptr<Tensor> expectedOutput0 = std::make_shared<Tensor>(Array4D<float,3,3,3,2> {
{ //
{ //
{{ 20, 22},{ 24, 26},{ 28, 30}}, //
{{ 26, 28},{ 30, 32},{ 34, 36}}, //
{{ 32, 34},{ 36, 38},{ 40, 42}} //
}, //
{ //
{{ 26, 28},{ 30, 32},{ 34, 36}}, //
{{ 32, 34},{ 36, 38},{ 40, 42}}, //
{{ 38, 40},{ 42, 44},{ 46, 48}} //
}, //
{ //
{{ 32, 34},{ 36, 38},{40, 42}}, //
{{ 38, 40},{ 42, 44},{46, 48}}, //
{{ 44, 46},{ 48, 50},{52, 54}} //
} //
} //
}); //
for(int i = 0; i < expectedOutput->size(); i++){ std::shared_ptr<Node> myAdd0 = Add();
const float targetOutput = *(static_cast<float*>(expectedOutput->getImpl()->rawPtr()) + i); auto op0 = std::static_pointer_cast<OperatorTensor>(myAdd0 -> getOperator());
REQUIRE(fabs(computedOutput[i] - targetOutput) < 1e-6); op0->associateInput(0, input_0);
op0->associateInput(1, input_1);
op0->setDataType(DataType::Float32);
op0->setBackend("cuda");
myAdd0->forward();
float* computedOutput0 = new float[expectedOutput0->size()]();
cudaMemcpy(computedOutput0, op0->getOutput(0)->getImpl()->rawPtr(), sizeof(float) * expectedOutput0->size(), cudaMemcpyDeviceToHost);
for(int i = 0; i < expectedOutput0->size(); i++){
const float targetOutput = *(static_cast<float*>(expectedOutput0->getImpl()->rawPtr()) + i);
REQUIRE(fabs(computedOutput0[i] - targetOutput) < 1e-6);
} }
delete[] computedOutput; delete[] computedOutput0;
/// Input0(d0, d1, d2, d3) + Input1(d3) = Output(d0, d1, d2, d3)
std::shared_ptr<Tensor> expectedOutput1 = std::make_shared<Tensor>(Array4D<float,3,1,3,2> {
{ //
{ //
{{100, 201},{102, 203},{104, 205}} //
}, //
{ //
{{106, 207},{108, 209},{110, 211}} //
}, //
{ //
{{112, 213},{114, 215},{116, 217}} //
} //
} //
}); //
std::shared_ptr<Node> myAdd1 = Add();
auto op1 = std::static_pointer_cast<OperatorTensor>(myAdd1 -> getOperator());
op1->associateInput(0, input_0);
op1->associateInput(1, input_2);
op1->setDataType(DataType::Float32);
op1->setBackend("cuda");
myAdd1->forward();
float* computedOutput1 = new float[expectedOutput1->size()]();
cudaMemcpy(computedOutput1, op1->getOutput(0)->getImpl()->rawPtr(), sizeof(float) * expectedOutput1->size(), cudaMemcpyDeviceToHost);
for(int i = 0; i < expectedOutput1->size(); i++){
const float targetOutput = *(static_cast<float*>(expectedOutput1->getImpl()->rawPtr()) + i);
REQUIRE(fabs(computedOutput1[i] - targetOutput) < 1e-6);
}
delete[] computedOutput1;
} }
SECTION("Random Input") { SECTION("Random Input") {
...@@ -231,11 +202,11 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") { ...@@ -231,11 +202,11 @@ TEST_CASE("[gpu/operator] Add(forward)", "[Add][GPU]") {
for (std::uint16_t trial = 0; trial < NBTRIALS; ++trial) for (std::uint16_t trial = 0; trial < NBTRIALS; ++trial)
{ {
// Create Add Operator CUDA // Create Add Operator CUDA
std::shared_ptr<Node> myAddCUDA = Add(2, "myaddcuda"); std::shared_ptr<Node> myAddCUDA = Add("myaddcuda");
auto op_cuda = std::static_pointer_cast<OperatorTensor>(myAddCUDA -> getOperator()); auto op_cuda = std::static_pointer_cast<OperatorTensor>(myAddCUDA -> getOperator());
// Create Add Operator CPU // Create Add Operator CPU
std::shared_ptr<Node> myAddCPU = Add(2, "myaddcpu"); std::shared_ptr<Node> myAddCPU = Add("myaddcpu");
auto op_cpu = std::static_pointer_cast<OperatorTensor>(myAddCPU -> getOperator()); auto op_cpu = std::static_pointer_cast<OperatorTensor>(myAddCPU -> getOperator());
op_cpu->setDataType(DataType::Float32); op_cpu->setDataType(DataType::Float32);
op_cpu->setBackend("cpu"); op_cpu->setBackend("cpu");
...@@ -360,16 +331,12 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") { ...@@ -360,16 +331,12 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") {
} // } //
}); // }); //
std::shared_ptr<Tensor> input_2 = std::make_shared<Tensor>(Array1D<float,2> {{100,200}});
input_0->setBackend("cuda"); input_0->setBackend("cuda");
input_1->setBackend("cuda"); input_1->setBackend("cuda");
input_2->setBackend("cuda"); std::shared_ptr<Node> myAdd = Add();
std::shared_ptr<Node> myAdd = Add(3);
auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator()); auto op = std::static_pointer_cast<OperatorTensor>(myAdd -> getOperator());
op->associateInput(0, input_0); op->associateInput(0, input_0);
op->associateInput(1, input_1); op->associateInput(1, input_1);
op->associateInput(2, input_2);
op->setDataType(DataType::Float32); op->setDataType(DataType::Float32);
op->setBackend("cuda"); op->setBackend("cuda");
myAdd->forward(); myAdd->forward();
...@@ -420,14 +387,11 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") { ...@@ -420,14 +387,11 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") {
} // } //
} // } //
}); // }); //
std::shared_ptr<Tensor> expectedInput3Grad = std::make_shared<Tensor>(Array1D<float,2> {{729, 756}});
float *computedGrad1Cuda = new float[expectedInput1Grad->size()](); float *computedGrad1Cuda = new float[expectedInput1Grad->size()]();
cudaMemcpy(computedGrad1Cuda, op->getInput(0)->grad()->getImpl()->rawPtr(), sizeof(float) * expectedInput1Grad->size(), cudaMemcpyDeviceToHost); cudaMemcpy(computedGrad1Cuda, op->getInput(0)->grad()->getImpl()->rawPtr(), sizeof(float) * expectedInput1Grad->size(), cudaMemcpyDeviceToHost);
float *computedGrad2Cuda = new float[expectedInput2Grad->size()](); float *computedGrad2Cuda = new float[expectedInput2Grad->size()]();
cudaMemcpy(computedGrad2Cuda, op->getInput(1)->grad()->getImpl()->rawPtr(), sizeof(float) * expectedInput2Grad->size(), cudaMemcpyDeviceToHost); cudaMemcpy(computedGrad2Cuda, op->getInput(1)->grad()->getImpl()->rawPtr(), sizeof(float) * expectedInput2Grad->size(), cudaMemcpyDeviceToHost);
float *computedGrad3Cuda = new float[expectedInput3Grad->size()]();
cudaMemcpy(computedGrad3Cuda, op->getInput(2)->grad()->getImpl()->rawPtr(), sizeof(float) * expectedInput3Grad->size(), cudaMemcpyDeviceToHost);
for(int i = 0; i < expectedInput1Grad->size(); i++){ for(int i = 0; i < expectedInput1Grad->size(); i++){
const float targetOutput = *(static_cast<float*>(expectedInput1Grad->getImpl()->rawPtr()) + i); const float targetOutput = *(static_cast<float*>(expectedInput1Grad->getImpl()->rawPtr()) + i);
...@@ -437,12 +401,7 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") { ...@@ -437,12 +401,7 @@ TEST_CASE("[gpu/operator] Add(backward)", "[Add][GPU]") {
const float targetOutput = *(static_cast<float*>(expectedInput2Grad->getImpl()->rawPtr()) + i); const float targetOutput = *(static_cast<float*>(expectedInput2Grad->getImpl()->rawPtr()) + i);
REQUIRE(fabs(computedGrad2Cuda[i] - targetOutput) < 1e-6); REQUIRE(fabs(computedGrad2Cuda[i] - targetOutput) < 1e-6);
} }
for(int i = 0; i < expectedInput3Grad->size(); i++){
const float targetOutput = *(static_cast<float*>(expectedInput3Grad->getImpl()->rawPtr()) + i);
REQUIRE(fabs(computedGrad3Cuda[i] - targetOutput) < 1e-6);
}
delete[] computedGrad1Cuda; delete[] computedGrad1Cuda;
delete[] computedGrad2Cuda; delete[] computedGrad2Cuda;
delete[] computedGrad3Cuda;
} }
\ No newline at end of file
0% Loading or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment