基于gpu与cuda c的神经网络实现

本神经网络为3层结构，一个输入层（N0N个神经元）+一个隐层（N1N个神经元）+一个输出层（N2N个神经元）

cuda加速为神经网络的3大主要计算部分   神经元输入计算，神经元错误计算，神经权重调整

#include<iostream>
#include<cmath>
#include<stdio.h>
#include<stdlib.h>
using namespace std;
#define N0N 3
#define N1N 2
#define N2N 1
#define HANDLE_ERROR(x) {if(x!=0) printf("error is caught on gpu: %d\n",x);}
//训练用的参数
struct neuroNode
{
        float I;
 float O;
 float Err;
 float bias;
};
float pic[N0N]={0};
neuroNode node1[N1N];
neuroNode node2[N2N];
float w01[N0N][N1N]={0};
float w12[N1N][N2N]={0};
float l=0.9;

//cuda第一加速部分
__global__ void cuda_fun_mul(float *a,float *b,float *c)
{
     int tid=threadIdx.x;
     c[tid]=a[tid]*b[tid];                 
}

//cuda第二加速部分
__global__ void cuda_fun_mul2(float *a,float *b,float *cc,int *NxN,float *l)
{
     int bid=blockIdx.x;
     int tid=threadIdx.x;
     cc[bid*(*NxN)+tid]=(*l)*a[bid]*b[tid];                 
}

//保存网络
void saveNet()
{
   int i,j,k;
   freopen("net.txt","w",stdout);
   //保存权重
   printf("biasnode1\n");
   for(i=0;i<N1N;i++)
    printf("%.2f ",node1[i].bias);
   printf("\n");
   printf("biasnode2\n");
   for(i=0;i<N2N;i++)
    printf("%.2f ",node2[i].bias);
   printf("\n");
   //保存偏量
   printf("w01[N0N][N1N]\n");
   for(i=0;i<N0N;i++)
   {
    for(j=0;j<N1N;j++)
    {
        printf("%.2f ",w01[i][j]);
    }
    printf("\n");
   }
   printf("w12[N1N][N2N]\n");
   for(i=0;i<N1N;i++)
   {
    for(j=0;j<N2N;j++)
    {
        printf("%.2f ",w12[i][j]);
    }
    printf("\n");
   }
   fclose(stdout); 
}
int main()
{
    int i,j,k;
   
     pic[0]=1;
  pic[1]=0;
  pic[2]=1;
  w01[0][0]=0.2;
  w01[0][1]=-0.3;
  w01[1][0]=0.4;
  w01[1][1]=0.1;
  w01[2][0]=-0.5;
  w01[2][1]=0.2;

  w12[0][0]=-0.3;
  w12[1][0]=-0.2;

  node1[0].bias=-0.4;
  node1[1].bias=0.2;
  node2[0].bias=0.1;

   
    int Nmax=N0N;
    if(Nmax<N1N)
        Nmax=N1N;
    if(Nmax<N2N)
        Nmax=N2N;
    float a[Nmax];
    float b[Nmax];
    float c[Nmax];
    float cc[Nmax*Nmax];
    int NxN;
   
    float *dev_a;
    float *dev_b;
    float *dev_c;
    float *dev_cc;
    HANDLE_ERROR(cudaMalloc((void**)&dev_a,sizeof(float)*Nmax));
    HANDLE_ERROR(cudaMalloc((void**)&dev_b,sizeof(float)*Nmax));
    HANDLE_ERROR(cudaMalloc((void**)&dev_c,sizeof(float)*Nmax));
    HANDLE_ERROR(cudaMalloc((void**)&dev_cc,sizeof(float)*Nmax*Nmax));
    float *dev_l;
    int *dev_NxN;
    HANDLE_ERROR(cudaMalloc((void**)&dev_l,sizeof(float)));
    HANDLE_ERROR(cudaMalloc((void**)&dev_NxN,sizeof(int)));

    HANDLE_ERROR(cudaMemcpy(dev_l,&l,sizeof(float),cudaMemcpyHostToDevice));
    

//训练神经网络

//正向过程

//1st层神经网络训练

  for(j=0;j<N1N;j++)//1st层神经网络训练，every neuro unit
  {
          node1[j].I=node1[j].bias;
            
           for(i=0;i<N0N;i++)//every colum of a pic，1层第j神经元 与 0层第i输入元的关系
           {
               a[i]=w01[i][j];
               b[i]=pic[i];
            }
            HANDLE_ERROR(cudaMemcpy(dev_a,a,sizeof(float)*N0N,cudaMemcpyHostToDevice));
            HANDLE_ERROR(cudaMemcpy(dev_b,b,sizeof(float)*N0N,cudaMemcpyHostToDevice));
            cuda_fun_mul<<<1,N0N>>>(dev_a,dev_b,dev_c);
            HANDLE_ERROR(cudaMemcpy(c,dev_c,sizeof(float)*N0N,cudaMemcpyDeviceToHost));
            for(i=0;i<N0N;i++)//every colum of a pic，1层第j神经元 与 0层第i输入元的关系
            {
                node1[j].I+=c[i];//w01[j][i] means 0层第j输入元的关系 与 1层第i神经元
            }
   
           node1[j].O=1.0/(1.0+exp(-node1[j].I));
  }  
  //1->2层训练
  for(j=0;j<N2N;j++)//2nd层神经网络训练，every neuro unit
  {
      node2[j].I=node2[j].bias;
     
   for(i=0;i<N1N;i++)//1层第i神经元与 2层第j输入元的关系
   {
       a[i]=w12[i][j];//w12[i][j]表示1st层第i输入单元到2nd层第j神经单元的权重
       b[i]=node1[i].O;
            }
            HANDLE_ERROR(cudaMemcpy(dev_a,a,sizeof(float)*N1N,cudaMemcpyHostToDevice));
            HANDLE_ERROR(cudaMemcpy(dev_b,b,sizeof(float)*N1N,cudaMemcpyHostToDevice));
            cuda_fun_mul<<<1,N1N>>>(dev_a,dev_b,dev_c);
            HANDLE_ERROR(cudaMemcpy(c,dev_c,sizeof(float)*N1N,cudaMemcpyDeviceToHost));
            for(i=0;i<N1N;i++)
   {
    node2[j].I+=c[i];
   }
   
            node2[j].O=1.0/(1.0+exp(-node2[j].I));
  }  

  //反向过程：
     //cul err
  int targetNum=0;
  for(i=0;i<N2N;i++)//2nd层神经元错误计算
  {
   if(i==targetNum)
       node2[i].Err=node2[i].O*(1-node2[i].O)*(1-node2[i].O);
   else
       node2[i].Err=node2[i].O*(1-node2[i].O)*(0-node2[i].O);
   node2[i].bias+=l*node2[i].Err;
  }
  for(i=0;i<N1N;i++)//1st层神经元错误计算,对于1st层第i个神经元作计算
  {
  node1[i].Err=0;

          for(j=0;j<N2N;j++)
          {
                  a[j]=w12[i][j];
                    b[j]=node2[j].Err;                
          }
          HANDLE_ERROR(cudaMemcpy(dev_a,a,sizeof(float)*N2N,cudaMemcpyHostToDevice));
          HANDLE_ERROR(cudaMemcpy(dev_b,b,sizeof(float)*N2N,cudaMemcpyHostToDevice));
          cuda_fun_mul<<<1,N2N>>>(dev_a,dev_b,dev_c);
          HANDLE_ERROR(cudaMemcpy(c,dev_c,sizeof(float)*N2N,cudaMemcpyDeviceToHost));  
          float temp=node1[i].O*(1-node1[i].O);
          for(j=0;j<N2N;j++)
   {
    node1[i].Err+=temp*c[j];
   }
  
   node1[i].bias+=l*node1[i].Err;
  }
  //change weight
  //w12
     for(i=0;i<N1N;i++)
         a[i]=node1[i].O;
     for(j=0;j<N2N;j++)
         b[j]=node2[j].Err;
     NxN=N2N;
     HANDLE_ERROR(cudaMemcpy(dev_a,a,sizeof(float)*N1N,cudaMemcpyHostToDevice));
     HANDLE_ERROR(cudaMemcpy(dev_b,b,sizeof(float)*N2N,cudaMemcpyHostToDevice));
     HANDLE_ERROR(cudaMemcpy(dev_NxN,&NxN,sizeof(int),cudaMemcpyHostToDevice));
     cuda_fun_mul2<<<N1N,N2N>>>(dev_a,dev_b,dev_cc,dev_NxN,dev_l);
     HANDLE_ERROR(cudaMemcpy(cc,dev_cc,sizeof(float)*N2N*N1N,cudaMemcpyDeviceToHost));  
            
     for(i=0;i<N1N;i++)
      for(j=0;j<N2N;j++)
       w12[i][j]+=*(cc+i*N2N+j); 
  //w01
     for(i=0;i<N0N;i++)
         a[i]=pic[i];
     for(j=0;j<N1N;j++)
         b[j]=node1[j].Err;
    
     NxN=N1N;
     HANDLE_ERROR(cudaMemcpy(dev_a,a,sizeof(float)*N0N,cudaMemcpyHostToDevice));
     HANDLE_ERROR(cudaMemcpy(dev_b,b,sizeof(float)*N1N,cudaMemcpyHostToDevice));
     HANDLE_ERROR(cudaMemcpy(dev_NxN,&NxN,sizeof(int),cudaMemcpyHostToDevice));
     cuda_fun_mul2<<<N0N,N1N>>>(dev_a,dev_b,dev_cc,dev_NxN,dev_l);
     HANDLE_ERROR(cudaMemcpy(cc,dev_cc,sizeof(float)*N0N*N1N,cudaMemcpyDeviceToHost));  
     for(i=0;i<N0N;i++)
      for(j=0;j<N1N;j++)
       w01[i][j]+=*(cc+i*N1N+j); 
 
     saveNet();
}