mmr/tkDNN
Archived
1
0
Fork 0
Code Issues 46 Pull Requests 12 Releases 4 Activity

works but it need cleaning

Browse Source
This commit is contained in:
Francesco Gatti
2020-02-16 16:28:39 +01:00
parent 8736a8c3da
commit 0ac292ea48
9 changed files with 219 additions and 71 deletions
Download Patch File Download Diff File Expand all files Collapse all files
include/tkDNN/Layer.h
+2 -1
View File
@@ -237,7 +237,7 @@ public:
virtual dnnType* infer(dataDim_t &dim, dnnType* srcData);
const bool bidirectional = false; /**> is the net bidir */
const bool bidirectional = true; /**> is the net bidir */
bool returnSeq = false; /**> if false return only the result of last timestep */
int stateSize = 0; /**> number of hidden states */
int seqLen = 0; /**> number of timesteps */
@@ -260,6 +260,7 @@ protected:
cudnnFilterDescriptor_t w_desc_;
dnnType *w_ptr;
dnnType *w_h;
dnnType *wf_ptr, *wb_ptr; // params pointer forward and backward layer
};
include/tkDNN/utils.h
+1 -1
View File
@@ -91,7 +91,7 @@
void printCenteredTitle(const char *title, char fill, int dim);
bool fileExist(const char *fname);
void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek = 0);
int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true);
int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true, int limit = 10);
void printDeviceVector(int size, dnnType* vec_d, bool device = true);
void resize(int size, dnnType **data);
src/LSTM.cpp
+131 -32
View File
@@ -37,7 +37,7 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig
net->dataType, 3, dimA, strideA));
dimA[0] = batchSize;
dimA[1] = bidirectional ? stateSize*2 : stateSize;
dimA[1] = stateSize;
dimA[2] = 1;
strideA[0] = dimA[2] * dimA[1];
strideA[1] = dimA[2];
@@ -51,7 +51,7 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig
// set the state tensors
dimA[0] = numLayers * (bidirectional ? 2 : 1);
dimA[0] = numLayers;
dimA[1] = batchSize;
dimA[2] = stateSize;
strideA[0] = dimA[2] * dimA[1];
@@ -91,7 +91,8 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig
checkCUDNN(cudnnSetRNNDescriptor(net->cudnnHandle,
rnnDesc, stateSize, numLayers, dropoutDesc,
cudnnRNNInputMode_t::CUDNN_LINEAR_INPUT,
(bidirectional ? cudnnDirectionMode_t::CUDNN_BIDIRECTIONAL : cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL),
//(bidirectional ? cudnnDirectionMode_t::CUDNN_BIDIRECTIONAL : cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL),
cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL,
cudnnRNNMode_t::CUDNN_LSTM,
cudnnRNNAlgo_t::CUDNN_RNN_ALGO_STANDARD,
net->dataType));
@@ -119,23 +120,26 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig
net->dataType, net->tensorFormat, 3, dim_w));
// load params
readBinaryFile(fname_weights, cudnn_params, &w_h, &w_ptr);
//allocate data for infer result
int dstDim = input_dim.n * stateSize*(bidirectional ? 2 : 1) * input_dim.h * input_dim.w;
checkCuda( cudaMalloc(&dstData, dstDim*sizeof(dnnType)) );
readBinaryFile(fname_weights, cudnn_params*2, &w_h, &w_ptr);
// set forward and backward params
wf_ptr = w_ptr;
wb_ptr = w_ptr + cudnn_params;
std::cout<<"wf: "<<wf_ptr<<" wb "<<wb_ptr<<"\n";
// set output dim
output_dim = input_dim;
output_dim.c = stateSize*(bidirectional ? 2 : 1);
//allocate data for infer result
checkCuda( cudaMalloc(&dstData, output_dim.tot()*sizeof(dnnType)) );
if(!returnSeq) {
output_dim.h = 1;
output_dim.w = 1;
}
/*
// Query weight layout
cudnnFilterDescriptor_t m_desc;
checkCUDNN(cudnnCreateFilterDescriptor(&m_desc));
@@ -192,6 +196,7 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig
}
checkCUDNN(cudnnDestroyFilterDescriptor(m_desc));
*/
}
LSTM::~LSTM() {
@@ -210,30 +215,124 @@ LSTM::~LSTM() {
dnnType* LSTM::infer(dataDim_t &dim, dnnType* srcData) {
std::cout<<"LSTM infer\n";
// reset states
checkCuda( cudaMemset(hx_ptr, 0, stateDataDim*sizeof(float)) );
checkCuda( cudaMemset(cx_ptr, 0, stateDataDim*sizeof(float)) );
dnnType *trans;
checkCuda( cudaMalloc(&trans, dim.tot()*sizeof(dnnType)));
matrixTranspose(net->cublasHandle, srcData, trans, dim.c, dim.h*dim.w*dim.l);
srcData = trans;
// reposition in invered order
dnnType *srcBack;
checkCuda( cudaMalloc(&srcBack, dim.tot()*sizeof(dnnType)));
for(int i=0; i<input_dim.w; i++) {
int off_0 = i*(input_dim.c);
int off_1 = (i+1)*(input_dim.c);
std::cout<<off_0<<" "<<off_1<<"\n";
checkCuda( cudaMemcpy(srcBack + dim.tot() - off_1, srcData + off_0, input_dim.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
}
dataDim_t singleOutput = input_dim;
singleOutput.c = stateSize;
dnnType *dstF = dstData;
dnnType *dstB = dstData + singleOutput.tot();
std::cout<<"INPUT:\n";
printDeviceVector(input_dim.tot(), srcData);
// forward
{
// reset states
checkCuda( cudaMemset(hx_ptr, 0, stateDataDim*sizeof(float)) );
checkCuda( cudaMemset(cx_ptr, 0, stateDataDim*sizeof(float)) );
checkCUDNN(cudnnRNNForwardInference(net->cudnnHandle,
rnnDesc,
seqLen, // number of time steps (nT)
x_desc_vec_.data(), // input array of desc (nT*nC_in)
srcData, // input pointer
hx_desc_, // initial hidden state desc
hx_ptr, // initial hidden state pointer
cx_desc_, // initial cell state desc
cx_ptr, // initial cell state pointer
w_desc_, // weights desc
wf_ptr, // weights pointer
y_desc_vec_.data(), // output desc (nT*nC_out)
dstF, // output pointer
hy_desc_, // final hidden state desc
hy_ptr, // final hidden state pointer
cy_desc_, // final cell state desc
cy_ptr, // final cell state pointer
work_space_, // workspace pointer
workspace_byte_)); // workspace size
}
std::cout<<"OUTPUT F:\n";
printDeviceVector(singleOutput.tot(), dstF);
std::cout<<"INPUT:\n";
printDeviceVector(input_dim.tot(), srcBack);
// backward
{
// reset states
checkCuda( cudaMemset(hx_ptr, 0, stateDataDim*sizeof(float)) );
checkCuda( cudaMemset(cx_ptr, 0, stateDataDim*sizeof(float)) );
checkCUDNN(cudnnRNNForwardInference(net->cudnnHandle,
rnnDesc,
seqLen, // number of time steps (nT)
x_desc_vec_.data(), // input array of desc (nT*nC_in)
srcBack, // input pointer
hx_desc_, // initial hidden state desc
hx_ptr, // initial hidden state pointer
cx_desc_, // initial cell state desc
cx_ptr, // initial cell state pointer
w_desc_, // weights desc
wb_ptr, // weights pointer
y_desc_vec_.data(), // output desc (nT*nC_out)
dstB, // output pointer
hy_desc_, // final hidden state desc
hy_ptr, // final hidden state pointer
cy_desc_, // final cell state desc
cy_ptr, // final cell state pointer
work_space_, // workspace pointer
workspace_byte_)); // workspace size
}
// reposition in invered order
dnnType *dstBack;
checkCuda( cudaMalloc(&dstBack, singleOutput.tot()*sizeof(dnnType)));
for(int i=0; i<singleOutput.w; i++) {
int off_0 = i*(singleOutput.c);
int off_1 = (i+1)*(singleOutput.c);
std::cout<<off_0<<" "<<off_1<<"\n";
checkCuda( cudaMemcpy(dstBack + singleOutput.tot() - off_1, dstB + off_0, singleOutput.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
}
dstB = dstBack;
std::cout<<"OUTPUT B:\n";
printDeviceVector(singleOutput.tot(), dstB);
checkCuda( cudaMalloc(&trans, singleOutput.tot()*2*sizeof(dnnType)));
if(returnSeq) {
// forward transpose
matrixTranspose(net->cublasHandle, dstF, trans,
singleOutput.h*singleOutput.w*singleOutput.l, singleOutput.c);
// backward transpose
matrixTranspose(net->cublasHandle, dstB, trans + singleOutput.tot(),
singleOutput.h*singleOutput.w*singleOutput.l, singleOutput.c);
dstData = trans;
} else {
// copy last of forward
checkCuda( cudaMemcpy(trans, dstF + singleOutput.tot() - singleOutput.c, singleOutput.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
// copy first of backward
checkCuda( cudaMemcpy(trans + singleOutput.c, dstB, singleOutput.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
dstData = trans;
}
checkCUDNN(cudnnRNNForwardInference(net->cudnnHandle,
rnnDesc,
seqLen, // number of time steps (nT)
x_desc_vec_.data(), // input array of desc (nT*nC_in)
srcData, // input pointer
hx_desc_, // initial hidden state desc
hx_ptr, // initial hidden state pointer
cx_desc_, // initial cell state desc
cx_ptr, // initial cell state pointer
w_desc_, // weights desc
w_ptr, // weights pointer
y_desc_vec_.data(), // output desc (nT*nC_out)
dstData, // output pointer
hy_desc_, // final hidden state desc
hy_ptr, // final hidden state pointer
cy_desc_, // final cell state desc
cy_ptr, // final cell state pointer
work_space_, // workspace pointer
workspace_byte_)); // workspace size
dim = output_dim;
return dstData;
src/utils.cpp
+2 -2
View File
@@ -68,7 +68,7 @@ void printDeviceVector(int size, dnnType* vec_d, bool device)
delete [] vec;
}
int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device) {
int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device, int limit) {
dnnType *data_h, *correct_h;
const float eps = 0.02f;
@@ -92,7 +92,7 @@ int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device) {
diffs += 1;
if(diffs == 1)
std::cout<<"\n";
if(diffs < 10)
if(diffs < limit)
std::cout<<" | [ "<<i<<" ]: "<<data_h[i]<<" "<<correct_h[i]<<"\n";
}
}
tests/imuodom/imuodom.cpp
+15 -11
View File
@@ -6,7 +6,6 @@ const char *i1_bin = "../tests/imuodom/layers/input1.bin";
const char *i2_bin = "../tests/imuodom/layers/input2.bin";
const char *o0_bin = "../tests/imuodom/layers/output0.bin";
const char *o1_bin = "../tests/imuodom/layers/output1.bin";
const char *output_bin = "../tests/imuodom/layers/output.bin";
const char *c0_bin = "../tests/imuodom/layers/conv1d_7.bin";
const char *c1_bin = "../tests/imuodom/layers/conv1d_8.bin";
@@ -17,6 +16,7 @@ const char *c5_bin = "../tests/imuodom/layers/conv1d_12.bin";
const char *l0_bin = "../tests/imuodom/layers/bidirectional_3.bin";
const char *l1_bin = "../tests/imuodom/layers/bidirectional_4.bin";
const char *d0_bin = "../tests/imuodom/layers/dense_3.bin";
const char *d1_bin = "../tests/imuodom/layers/dense_4.bin";
int main() {
@@ -51,10 +51,14 @@ int main() {
tk::dnn::Layer *concat_l[3] = { &x0_2, &x1_2, &x2_2 };
tk::dnn::Route concat (&net, concat_l, 3);
//tk::dnn::LSTM lstm0(&net, 128, true, l0_bin);
//tk::dnn::LSTM lstm1(&net, 128, false, l1_bin);
//tk::dnn::Dense d0 (&net, 3, d0_bin);
tk::dnn::LSTM lstm0(&net, 128, true, l0_bin);
tk::dnn::LSTM lstm1(&net, 128, false, l1_bin);
tk::dnn::Dense d0 (&net, 3, d0_bin);
tk::dnn::Layer *lstm1_l[1] = { &lstm1 };
tk::dnn::Route lstm1_link (&net, lstm1_l, 1);
tk::dnn::Dense d1 (&net, 4, d1_bin);
net.print();
dnnType *data;
@@ -66,12 +70,12 @@ int main() {
TIMER_STOP
// Print real test
//std::cout<<"\n==== CHECK RESULT ====\n";
//dnnType *out;
//dnnType *out_h;
//readBinaryFile(output_bin, dim.tot(), &out_h, &out);
//checkResult(dim.tot(), data, out);
printDeviceVector(100, data);
std::cout<<"\n==== CHECK RESULT ====\n";
dnnType *out0, *out1;
dnnType *out0_h, *out1_h;
readBinaryFile(o0_bin, d0.output_dim.tot(), &out0_h, &out0);
readBinaryFile(o1_bin, d1.output_dim.tot(), &out1_h, &out1);
checkResult(d0.output_dim.tot(), d0.dstData, out0);
checkResult(d1.output_dim.tot(), d1.dstData, out1);
return 0;
}
tests/imuodom/infer.py
+14 -11
View File
@@ -34,30 +34,33 @@ if __name__ == '__main__':
[yhat_delta_p, yhat_delta_q] = model.predict([x_angle, x_gyro, x_acc], batch_size=1, verbose=1)
layer_name = 'bidirectional_3'
intermediate_layer_model = Model(inputs=model.input,
outputs=model.get_layer(layer_name).output)
intermediate_output = intermediate_layer_model.predict([x_angle, x_gyro, x_acc])
#layer_name = 'dense_4'
#intermediate_layer_model = Model(inputs=model.input,
# outputs=model.get_layer(layer_name).output)
#intermediate_output = intermediate_layer_model.predict([x_angle, x_gyro, x_acc])
x_angle = np.array([x_angle])
x_gyro = np.array([x_gyro])
x_acc = np.array([x_acc])
intermediate_output = np.array([intermediate_output])
#intermediate_output = np.array([intermediate_output])
x_angle = x_angle.transpose(0, 3, 1, 2)
x_gyro = x_gyro.transpose(0, 3, 1, 2)
x_acc = x_acc.transpose(0, 3, 1, 2)
intermediate_output = intermediate_output.transpose(0, 3, 1, 2)
#intermediate_output = intermediate_output.transpose(0, 3, 1, 2)
#print("Aggregate:")
#print(intermediate_output.tolist())
print("x0: ", np.shape(x_angle))
print("out: ",np.shape(intermediate_output))
#print("out: ",np.shape(intermediate_output))
x_angle = np.array(x_angle.flatten(), dtype=np.float32)
x_gyro = np.array(x_gyro.flatten(), dtype=np.float32)
x_acc = np.array(x_acc.flatten(), dtype=np.float32)
yhat_delta_p = np.array(yhat_delta_p.flatten(), dtype=np.float32)
yhat_delta_q = np.array(yhat_delta_q.flatten(), dtype=np.float32)
intermediate_output = np.array(intermediate_output.flatten(), dtype=np.float32)
#intermediate_output = np.array(intermediate_output.flatten(), dtype=np.float32)
f = open("layers/input0.bin", mode='wb')
@@ -70,6 +73,6 @@ if __name__ == '__main__':
bin_write(f, yhat_delta_p)
f = open("layers/output1.bin", mode='wb')
bin_write(f, yhat_delta_q)
f = open("layers/output.bin", mode='wb')
bin_write(f, intermediate_output)
#f = open("layers/output.bin", mode='wb')
#bin_write(f, intermediate_output)
tests/simple/test_model.py
+10 -5
View File
@@ -2,6 +2,7 @@ import keras
import numpy as np
from keras.models import Sequential
from keras.layers import Input, Dense, Activation, Flatten, Dropout, ELU, Reshape, Lambda, Conv1D
from keras.layers import Bidirectional, CuDNNLSTM
from keras.layers.convolutional import Convolution2D, Convolution3D
from keras.layers.pooling import MaxPooling2D, MaxPooling3D, AveragePooling3D
from keras.models import Sequential, Model
@@ -17,9 +18,11 @@ def bin_write(f, data):
f.write(bin)
def create_model():
x1 = Input((6, 16), name='x1')
x1 = Input((3, 8), name='x1')
conv = Conv1D(4, 2)(x1)
model = Model([x1], [conv])
lstm = Bidirectional(CuDNNLSTM(5, return_sequences=True))(conv)
lstm2 = Bidirectional(CuDNNLSTM(5, return_sequences=False))(lstm)
model = Model([x1], [lstm2])
model.summary()
return model
@@ -30,14 +33,16 @@ if __name__ == '__main__':
model = create_model()
model.save("net.hdf5")
x = np.random.rand(1,1,6,16)
np.random.seed(2)
x = np.random.rand(1,1,3,8)
r = model.predict( x[0], batch_size=1)
r = np.array([r])
r = np.array([r])
x = x.transpose(0, 3, 1, 2)
r = r.transpose(0, 3, 1, 2)
#r = r.transpose(0, 3, 1, 2)
print("in: ", np.shape(x))
print("out: ", np.shape(r))
print("output: ", r.tolist())
x = np.array(x.flatten(), dtype=np.float32)
f = open("input.bin", mode='wb')
tests/simple/test_simple.cpp
+7 -1
View File
@@ -3,14 +3,20 @@
const char *input_bin = "../tests/simple/input.bin";
const char *c0_bin = "../tests/simple/layers/conv1d_1.bin";
const char *l1_bin = "../tests/simple/layers/bidirectional_1.bin";
const char *l2_bin = "../tests/simple/layers/bidirectional_2.bin";
const char *output_bin = "../tests/simple/output.bin";
int main() {
// Network layout
tk::dnn::dataDim_t dim(1, 16, 1, 6);
tk::dnn::dataDim_t dim(1, 8, 1, 3);
tk::dnn::Network net(dim);
tk::dnn::Conv2d l0(&net, 4, 1, 2, 1, 1, 0, 0, c0_bin);
tk::dnn::LSTM l1(&net, 5, true, l1_bin);
tk::dnn::LSTM l2(&net, 5, false, l2_bin);
net.print();
// Load input
dnnType *data;
tests/weights_exporter.py
+37 -7
View File
@@ -42,17 +42,47 @@ def export_layer(name, weights, bias):
bin_write(f, bias)
print ("WEIGHTS saved\n")
def export_bidir(name, params):
def export_bidir(name, params, paramsb):
print ("######## EXPORT", name, "LAYER ########")
f = open(name + ".bin", mode='wb')
print("FORWARD")
ker = params[0]
rec_ker = params[1]
bias = params[2]
print ("export kernels: ", np.shape(ker))
units = np.shape(ker)[1] // 4
bin_write(f, ker[:,:units])
bin_write(f, ker[:,units:units*2])
bin_write(f, ker[:,units*2:units*3])
bin_write(f, ker[:,units*3:])
print ("export recurrent kernels: ", np.shape(rec_ker))
bin_write(f, rec_ker[:,:units])
bin_write(f, rec_ker[:,units:units*2])
bin_write(f, rec_ker[:,units*2:units*3])
bin_write(f, rec_ker[:,units*3:])
print ("export kernels: ", np.shape(ker))
bin_write(f, bias)
print("WEIGHTS saved\n")
for w in params:
#w = w.transpose()
print(np.shape(w))
bin_write(f, w)
print("BACKWARD")
ker = paramsb[0]
rec_ker = paramsb[1]
bias = paramsb[2]
print ("export kernels: ", np.shape(ker))
units = np.shape(ker)[1] // 4
bin_write(f, ker[:,:units])
bin_write(f, ker[:,units:units*2])
bin_write(f, ker[:,units*2:units*3])
bin_write(f, ker[:,units*3:])
print ("export recurrent kernels: ", np.shape(rec_ker))
bin_write(f, rec_ker[:,:units])
bin_write(f, rec_ker[:,units:units*2])
bin_write(f, rec_ker[:,units*2:units*3])
bin_write(f, rec_ker[:,units*3:])
print ("export kernels: ", np.shape(ker))
bin_write(f, bias)
print("WEIGHTS saved\n")
#https://github.com/fchollet/keras/wiki/Converting-convolution-kernels-from-Theano-to-TensorFlow-and-vice-versa
@@ -100,7 +130,7 @@ if __name__ == '__main__':
export_layer(args.output + "/" + name, wgs[0], wgs[1])
elif name.startswith("bidirectional"):
wgs = l.forward_layer.get_weights()
export_bidir(args.output + "/" + name, wgs)
export_bidir(args.output + "/" + name, l.forward_layer.get_weights(), l.backward_layer.get_weights())
else:
print ("skip:", name, "has no weights")
continue