Homec4science
Diffusion lammps (lammps-icms) Browse

lammps/lib/cuda/cuda_data.cu
bac07b76cef2lammps-icms

Last Change
Raw File

cuda_data.cu
View Options

enum copy_mode {x, xx, xy, yx, xyz, xzy}; // yxz, yzx, zxy, zyx not yet implemented since they were not needed yet
#include "cuda_data_cu.h"
#include "cuda_wrapper_cu.h"
#include "cuda_data_kernel.cu"
#include <cstdio>
void CudaData_Upload_DoubleFloat(void* host_data,void* dev_data, unsigned* n,copy_mode mode,void* buffer)
{
int size=n[0];
if(n[1]>0) size*=n[1];
if(n[2]>0) size*=n[2];
dim3 threads; threads.x=1; threads.y=1; threads.z=1;
dim3 grid; grid.x=1; grid.y=1; grid.z=1;
if(size<=128*30)
threads.x=32;
else if(size<=256*30)
threads.x=64;
else if(size<=512*30)
threads.x=128;
else
threads.x=256;
grid.x=((size-1)+threads.x)/threads.x;
if(grid.x>32000)
grid.x=32000;
while(grid.x*grid.y*threads.x<size) grid.y++;
float debugdata[size];
//int* cu_debug=(int*) CudaWrapper_AllocCudaData(size*sizeof(FLOAT));
size*=sizeof(double);
printf("size: %i (%i %i %i) (%i %i %i) %p\n",size,grid.x,grid.y,threads.x,n[0],n[1],n[2],buffer);
CudaWrapper_UploadCudaData(host_data, buffer, size);
CudaData_Upload_Kernel_DoubleFloat<<<grid,threads>>>((double*)buffer,(float*)dev_data,n[0],n[1],n[2],mode);
cudaThreadSynchronize();
CudaWrapper_DownloadCudaData(debugdata, dev_data, size/2);
double sum=0;
printf("debugdata: ");
for(int i=0;i<size/sizeof(double);i++) sum+=(debugdata[i]-((double*) host_data)[i])*(debugdata[i]-((double*) host_data)[i]);
printf("%lf \n",sum);
}
void CudaData_Upload_DoubleDouble(void* host_data,void* dev_data, unsigned* n,copy_mode mode,void* buffer)
{
int size=n[0];
if(n[1]>0) size*=n[1];
if(n[2]>0) size*=n[2];
dim3 threads; threads.x=1; threads.y=1; threads.z=1;
dim3 grid; grid.x=1; grid.y=1; grid.z=1;
if(size<=128*30)
threads.x=32;
else if(size<=256*30)
threads.x=64;
else if(size<=512*30)
threads.x=128;
else
threads.x=256;
grid.x=((size-1)+threads.x)/threads.x;
if(grid.x>32000)
grid.x=32000;
while(grid.x*grid.y*threads.x<size) grid.y++;
size*=sizeof(double);
CudaWrapper_UploadCudaData(host_data, buffer, size);
CudaData_Upload_Kernel_DoubleDouble<<<grid,threads>>>((double*)buffer,(double*)dev_data,n[0],n[1],n[2],mode);
cudaThreadSynchronize();
}
void CudaData_Upload_FloatDouble(void* host_data,void* dev_data, unsigned* n,copy_mode mode,void* buffer)
{
int size=n[0];
if(n[1]>0) size*=n[1];
if(n[2]>0) size*=n[2];
dim3 threads; threads.x=1; threads.y=1; threads.z=1;
dim3 grid; grid.x=1; grid.y=1; grid.z=1;
if(size<=128*30)
threads.x=32;
else if(size<=256*30)
threads.x=64;
else if(size<=512*30)
threads.x=128;
else
threads.x=256;
grid.x=((size-1)+threads.x)/threads.x;
if(grid.x>32000)
grid.x=32000;
while(grid.x*grid.y*threads.x<size) grid.y++;
size*=sizeof(float);
CudaWrapper_UploadCudaData(host_data, buffer, size);
CudaData_Upload_Kernel_FloatDouble<<<grid,threads>>>((float*)buffer,(double*)dev_data,n[0],n[1],n[2],mode);
cudaThreadSynchronize();
}
void CudaData_Upload_FloatFloat(void* host_data,void* dev_data, unsigned* n,copy_mode mode,void* buffer)
{
int size=n[0];
if(n[1]>0) size*=n[1];
if(n[2]>0) size*=n[2];
dim3 threads; threads.x=1; threads.y=1; threads.z=1;
dim3 grid; grid.x=1; grid.y=1; grid.z=1;
if(size<=128*30)
threads.x=32;
else if(size<=256*30)
threads.x=64;
else if(size<=512*30)
threads.x=128;
else
threads.x=256;
grid.x=((size-1)+threads.x)/threads.x;
if(grid.x>32000)
grid.x=32000;
while(grid.x*grid.y*threads.x<size) grid.y++;
size*=sizeof(float);
CudaWrapper_UploadCudaData(host_data, buffer, size);
CudaData_Upload_Kernel_FloatFloat<<<grid,threads>>>((float*)buffer,(float*)dev_data,n[0],n[1],n[2],mode);
cudaThreadSynchronize();
}
void CudaData_Upload_IntInt(void* host_data,void* dev_data, unsigned* n,copy_mode mode,void* buffer)
{
int size=n[0];
if(n[1]>0) size*=n[1];
if(n[2]>0) size*=n[2];
dim3 threads; threads.x=1; threads.y=1; threads.z=1;
dim3 grid; grid.x=1; grid.y=1; grid.z=1;
if(size<=128*30)
threads.x=32;
else if(size<=256*30)
threads.x=64;
else if(size<=512*30)
threads.x=128;
else
threads.x=256;
grid.x=((size-1)+threads.x)/threads.x;
if(grid.x>32000)
grid.x=32000;
while(grid.x*grid.y*threads.x<size) grid.y++;
size*=sizeof(int);
CudaWrapper_UploadCudaData(host_data, buffer, size);
CudaData_Upload_Kernel_IntInt<<<grid,threads>>>((int*)buffer,(int*)dev_data,n[0],n[1],n[2],mode);
cudaThreadSynchronize();
}
void CudaData_Download(void* host_data,void* dev_data,int host_size, int dev_size, unsigned* n,copy_mode mode,void* buffer)
{
}
c4science · Help