Skip to content
Projects
Groups
Snippets
Help
This project
Loading...
Sign in / Register
Toggle navigation
S
santan-cuda-drr-service
Project
Project
Details
Activity
Cycle Analytics
Repository
Repository
Files
Commits
Branches
Tags
Contributors
Graph
Compare
Charts
Issues
0
Issues
0
List
Board
Labels
Milestones
Merge Requests
0
Merge Requests
0
CI / CD
CI / CD
Pipelines
Jobs
Schedules
Charts
Wiki
Wiki
Snippets
Snippets
Members
Members
Collapse sidebar
Close sidebar
Activity
Graph
Charts
Create a new issue
Jobs
Commits
Issue Boards
Open sidebar
刘劭荣(20软)
santan-cuda-drr-service
Commits
175478db
Commit
175478db
authored
Sep 20, 2020
by
刘劭荣(20软)
Browse files
Options
Browse Files
Download
Email Patches
Plain Diff
增加数组类型参数
parent
c1a78705
Hide whitespace changes
Inline
Side-by-side
Showing
12 changed files
with
366 additions
and
59 deletions
+366
-59
README.md
README.md
+6
-5
CUDAdrr.cu
cuda-drr-socket/CUDAdrr.cu
+95
-16
CUDAdrr.cuh
cuda-drr-socket/CUDAdrr.cuh
+4
-3
drr.cpp
cuda-drr-socket/drr.cpp
+183
-2
drr.h
cuda-drr-socket/drr.h
+2
-1
main.cpp
cuda-drr-socket/main.cpp
+12
-3
socketManager.cpp
cuda-drr-socket/socketManager.cpp
+40
-13
socketManager.h
cuda-drr-socket/socketManager.h
+2
-1
vc141.idb
cuda-drr-socket/x64/Debug/vc141.idb
+0
-0
main.cpp
test/main.cpp
+22
-15
cuda-drr-socket.iobj
x64/Release/cuda-drr-socket.iobj
+0
-0
cuda-drr-socket.ipdb
x64/Release/cuda-drr-socket.ipdb
+0
-0
No files found.
README.md
View file @
175478db
...
...
@@ -4,7 +4,7 @@
-
VS2017
-
ITK5.0.1
-
CUDA
-
CUDA
11.0
# 快速开始
## 配置说明
...
...
@@ -22,18 +22,19 @@
### socket
*
用户向本端口发送
`CT(ct_path_name)`
即可启动该功能,例如发送
`CT(C:/CT/CT001)`
*
待程序读取完毕后,将返回套接字
`ct
upload succeed
`
作为结束
*
待程序读取完毕后,将返回套接字
`ct
imported successfully
`
作为结束
## 生成DRR图像数组
### socket
*
用户必须要在接收到CT序列成功读取的信息后,才能进行该功能
*
用户向本端口发送
`PARA(rx, ry, rz, dx, dy, threshold)`
即可启动该功能
*
用户向本端口发送
`PARA(rx, ry, rz, dx, dy, threshold
, flag
)`
即可启动该功能
*
rx, ry, rz (float): 模型的绕XYZ轴的旋转角度(-180~180)
*
dx, dy (int): 生成图像的大小
*
threshold (float): DRR筛选阈值
*
例如
`PARA(90.0, 0.0, 0.0, 1024, 1024, 0.0)`
*
DRR生成完毕后,用户将接收一个float类型的一维数组,长度为dx
*
dy,用来表示生成的DRR图像
*
flag (int): 选择数组的返回类型,0为float数组,1为unsigned char数组
*
例如
`PARA(90.0, 0.0, 0.0, 1024, 1024, 0.0, 1)`
*
DRR生成完毕后,用户将接收一个一维数组,长度为dx
*
dy,用来表示生成的DRR图像
## 客户端示例
参见test文件夹
cuda-drr-socket/CUDAdrr.cu
View file @
175478db
...
...
@@ -21,6 +21,7 @@ float* d_object3D;
// This variable contains the 2D output from CUDA
float *d_object2D;
unsigned char *c_object2D;
// Constants depending on the DICOM
DEVICE_CONST int d_sizeCT[3];
...
...
@@ -72,12 +73,15 @@ void loadDICOMInGPUMemory(float *cpp_object3D, int *sizeCT, float *pixelSpacing)
}
void loadOuputVariablesInGPUMemory(int dimX, int dimZ)
void loadOuputVariablesInGPUMemory(int dimX, int dimZ
, int flag
)
{
long int vectorSize = dimX * dimZ;
int OutputImageSize[2] = { dimX, dimZ };
cudaMalloc((void**)&d_object2D, vectorSize * sizeof(float));
if (flag == 0)
cudaMalloc((void**)&d_object2D, vectorSize * sizeof(float));
else
cudaMalloc((void**)&c_object2D, vectorSize * sizeof(unsigned char));
cudaMemcpyToSymbol(DRRImageSize, OutputImageSize, 2 * sizeof(int), 0, cudaMemcpyHostToDevice);
}
...
...
@@ -88,9 +92,12 @@ void freeDICOMFromGPUMemory()
cudaFree(d_object3D);
}
void freeAuxiliaryVariablesInGPUMemory()
void freeAuxiliaryVariablesInGPUMemory(
int flag
)
{
cudaFree(d_object2D);
if (flag == 0)
cudaFree(d_object2D);
else
cudaFree(c_object2D);
}
__global__ void hist(float* object2D, int* range) {
...
...
@@ -104,14 +111,26 @@ __global__ void hist(float* object2D, int* range) {
if (idx >= total_dx * total_dz)
return;
object2D[idx] = (object2D[idx] - range[0]) / (range[1] - range[0]) * 255;
}
__global__ void hist_uc(unsigned char* object2D, int* range) {
int total_dx = DRRImageSize[0];
int total_dz = DRRImageSize[1];
__global__ void drrCUDA(float* object2D, cudaTextureObject_t tex_object3D, int* hist)
{
//Every thread calculates its own id number
long int idx = (blockIdx.x*blockDim.x) + threadIdx.x;
// This checks if the thread number is bigger than the amount of pixels
if (idx >= total_dx * total_dz)
return;
float tmp = object2D[idx];
object2D[idx] = (int)((tmp - range[0]) / (range[1] - range[0]) * 255);
}
__device__ float getPixval(cudaTextureObject_t tex_object3D, long int idx) {
float stepInX[3] = { d_DRR_Parameters[0],d_DRR_Parameters[1],d_DRR_Parameters[2] };
float stepInY[3] = { d_DRR_Parameters[3],d_DRR_Parameters[4],d_DRR_Parameters[5] };
float corner00[3] = { d_DRR_Parameters[6],d_DRR_Parameters[7],d_DRR_Parameters[8] };
...
...
@@ -121,13 +140,6 @@ __global__ void drrCUDA(float* object2D, cudaTextureObject_t tex_object3D, int*
int total_dx = DRRImageSize[0];
int total_dz = DRRImageSize[1];
//Every thread calculates its own id number
long int idx = (blockIdx.x*blockDim.x) + threadIdx.x;
// This checks if the thread number is bigger than the amount of pixels
if (idx >= total_dx * total_dz)
return;
// Converting number of pixels to rows and columns
int dz = idx / total_dx;
int dx = idx - dz * total_dx;
...
...
@@ -368,18 +380,49 @@ __global__ void drrCUDA(float* object2D, cudaTextureObject_t tex_object3D, int*
}
} //end of the while-loop
float pixval = 255.0
-
d12;
float pixval = 255.0
-
d12;
if (pixval < 0)
pixval = 255.;
if (pixval > 255)
pixval = 0.;
return pixval;
}
__global__ void drrCUDA(float* object2D, cudaTextureObject_t tex_object3D, int* hist)
{
int total_dx = DRRImageSize[0];
int total_dz = DRRImageSize[1];
//Every thread calculates its own id number
long int idx = (blockIdx.x*blockDim.x) + threadIdx.x;
// This checks if the thread number is bigger than the amount of pixels
if (idx >= total_dx * total_dz)
return;
float pixval = getPixval(tex_object3D, idx);
int tmp = (int)pixval;
atomicAdd(hist + tmp, 1);
//Assign the calculated value for the pixel to its corresponding position in the output array
object2D[idx] = pixval;
}
__global__ void drrCUDA_uc(unsigned char* object2D, cudaTextureObject_t tex_object3D, int* hist) {
int total_dx = DRRImageSize[0];
int total_dz = DRRImageSize[1];
//Every thread calculates its own id number
long int idx = (blockIdx.x*blockDim.x) + threadIdx.x;
// This checks if the thread number is bigger than the amount of pixels
if (idx >= total_dx * total_dz)
return;
float pixval = getPixval(tex_object3D, idx);
int tmp = (int)pixval;
atomicAdd(hist + tmp, 1);
//Assign the calculated value for the pixel to its corresponding position in the output array
object2D[idx] = (int)pixval;
}
__global__ void cal_hist(int* hist, int* range) {
...
...
@@ -399,7 +442,7 @@ __global__ void cal_hist(int* hist, int* range) {
}
}
float* cal
culate
DRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters)
float* cal
Float
DRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters)
{
cudaMemcpyToSymbol(d_DRR_Parameters, DRR_Parameters.stepInX, 15 * sizeof(float), 0, cudaMemcpyHostToDevice);
...
...
@@ -435,6 +478,42 @@ float* calculateDRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_P
return h_object2D;
}
unsigned char* calUCharDRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters)
{
cudaMemcpyToSymbol(d_DRR_Parameters, DRR_Parameters.stepInX, 15 * sizeof(float), 0, cudaMemcpyHostToDevice);
//Block 6
int num_Threads = CUDA_Parameters.numThreads;
int num_Blocks = CUDA_Parameters.numBlocks;
//------------------------------------------------------------
//Launching the threads
int cpu_hist[256] = { 0 };
int* gpu_hist;
cudaMalloc((int**)&gpu_hist, 256 * sizeof(int));
cudaMemcpy(gpu_hist, cpu_hist, 256 * sizeof(int), cudaMemcpyHostToDevice);
drrCUDA_uc << < num_Blocks, num_Threads >> > (c_object2D, tex_object3D, gpu_hist);
cudaMemcpy(cpu_hist, gpu_hist, 256 * sizeof(int), cudaMemcpyDeviceToHost);
int* gpu_range;
cudaMalloc((int**)&gpu_range, 2 * sizeof(int));
cal_hist << <1, 1 >> > (gpu_hist, gpu_range);
hist_uc << <num_Blocks, num_Threads >> > (c_object2D, gpu_range);
//------------------------------------------------------------
cudaFree(gpu_hist);
cudaFree(gpu_range);
//Copying the result from the calculations from device to host
long int vectorSize = (int)DRR_Parameters.size[0] * (int)DRR_Parameters.size[1];
unsigned char* h_object2D = (unsigned char*)malloc(sizeof(unsigned char)*vectorSize);
cudaMemcpy(h_object2D, c_object2D, vectorSize * sizeof(unsigned char), cudaMemcpyDeviceToHost);
return h_object2D;
}
void HandleCudaKernelError(const cudaError_t CudaError, const char* pName /*= ""*/)
{
...
...
cuda-drr-socket/CUDAdrr.cuh
View file @
175478db
...
...
@@ -39,9 +39,10 @@ struct DRRParameters
};
void loadDICOMInGPUMemory(float *cpp_object3D, int *sizeCT, float *pixelSpacing);
void loadOuputVariablesInGPUMemory(int dimX, int dimZ);
void loadOuputVariablesInGPUMemory(int dimX, int dimZ
, int flag=0
);
void freeDICOMFromGPUMemory();
void freeAuxiliaryVariablesInGPUMemory();
void freeAuxiliaryVariablesInGPUMemory(
int flag=0
);
float* calculateDRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters);
float* calFloatDRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters);
unsigned char* calUCharDRRwithCUDA(CUDAParamerters CUDA_Parameters, DRRParameters DRR_Parameters);
cuda-drr-socket/drr.cpp
View file @
175478db
...
...
@@ -63,6 +63,8 @@ drr::drr(std::string input)
PixelSpacingCT
[
1
]
=
ctPixelSpacing
[
1
];
PixelSpacingCT
[
2
]
=
ctPixelSpacing
[
2
];
loadDICOMInGPUMemory
(
cpp_object3D
,
SizeCT
,
PixelSpacingCT
);
free
(
cpp_object3D
);
cpp_object3D
=
NULL
;
std
::
cout
<<
"itk readfile :"
<<
(
double
)
clock
()
/
CLOCKS_PER_SEC
-
time
<<
std
::
endl
;
}
...
...
@@ -72,7 +74,7 @@ drr::~drr()
freeDICOMFromGPUMemory
();
}
float
*
drr
::
cudaDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
float
*
drr
::
cuda
Float
DRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
{
InputImageType
::
PointType
ori
;
ori
[
0
]
=
0.0
;
...
...
@@ -243,7 +245,7 @@ float* drr::cudaDRR(float rx, float ry, float rz, int dx, int dy, float threshol
loadOuputVariablesInGPUMemory
((
int
)
output_size
[
0
],(
int
)
output_size
[
1
]);
float
*
object2d
=
cal
culate
DRRwithCUDA
(
cudaPara
,
para
);
float
*
object2d
=
cal
Float
DRRwithCUDA
(
cudaPara
,
para
);
freeAuxiliaryVariablesInGPUMemory
();
...
...
@@ -298,3 +300,182 @@ float* drr::cudaDRR(float rx, float ry, float rz, int dx, int dy, float threshol
}
unsigned
char
*
drr
::
cudaUCharDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
{
InputImageType
::
PointType
ori
;
ori
[
0
]
=
0.0
;
ori
[
1
]
=
0.0
;
ori
[
2
]
=
0.0
;
image
->
SetOrigin
(
ori
);
typedef
itk
::
Euler3DTransform
<
double
>
TransformType
;
TransformType
::
Pointer
transform
=
TransformType
::
New
();
TransformType
::
OutputVectorType
translation
;
translation
[
0
]
=
0
;
translation
[
1
]
=
0
;
translation
[
2
]
=
0
;
const
double
dtr
=
(
atan
(
1.0
)
*
4.0
)
/
180.0
;
transform
->
SetTranslation
(
translation
);
//transform->SetComputeZYX(true);
transform
->
SetRotation
(
dtr
*
rx
,
dtr
*
ry
,
dtr
*
rz
);
itk
::
Vector
<
double
>
SourceWorld
;
typename
InputImageType
::
SizeType
sizeCT
;
typename
InputImageType
::
RegionType
regionCT
;
typename
InputImageType
::
SpacingType
ctPixelSpacing
;
typename
InputImageType
::
PointType
ctOrigin
;
ctPixelSpacing
=
image
->
GetSpacing
();
ctOrigin
=
image
->
GetOrigin
();
regionCT
=
image
->
GetLargestPossibleRegion
();
sizeCT
=
regionCT
.
GetSize
();
InputImageType
::
SizeType
imSize
=
image
->
GetBufferedRegion
().
GetSize
();
TransformType
::
InputPointType
isocenter
;
isocenter
[
0
]
=
ctOrigin
[
0
]
+
ctPixelSpacing
[
0
]
*
static_cast
<
double
>
(
imSize
[
0
])
/
2.0
;
isocenter
[
1
]
=
ctOrigin
[
1
]
+
ctPixelSpacing
[
1
]
*
static_cast
<
double
>
(
imSize
[
1
])
/
2.0
;
isocenter
[
2
]
=
ctOrigin
[
2
]
+
ctPixelSpacing
[
2
]
*
static_cast
<
double
>
(
imSize
[
2
])
/
2.0
;
transform
->
SetCenter
(
isocenter
);
TransformType
::
Pointer
m_InverseTransform
=
TransformType
::
New
();
m_InverseTransform
->
SetComputeZYX
(
true
);
TransformType
::
Pointer
m_ComposeTransform
=
TransformType
::
New
();
m_ComposeTransform
->
SetComputeZYX
(
true
);
TransformType
::
Pointer
m_GantryRoTransform
=
TransformType
::
New
();
m_GantryRoTransform
->
SetComputeZYX
(
true
);
m_GantryRoTransform
->
SetIdentity
();
TransformType
::
Pointer
m_CamShiftTransform
=
TransformType
::
New
();
m_CamShiftTransform
->
SetComputeZYX
(
true
);
m_CamShiftTransform
->
SetIdentity
();
TransformType
::
Pointer
m_CamRotTransform
=
TransformType
::
New
();
m_CamRotTransform
->
SetComputeZYX
(
true
);
m_CamRotTransform
->
SetIdentity
();
float
dtrf
=
(
atan
(
1.0
)
*
4.0
)
/
180.0
;
m_CamRotTransform
->
SetRotation
(
dtrf
*
(
-
90.0
),
0.0
,
0.0
);
m_ComposeTransform
->
SetIdentity
();
m_ComposeTransform
->
Compose
(
transform
,
0
);
float
rprojection
=
90.
;
float
angle
=
-
dtr
*
rprojection
;
m_GantryRoTransform
->
SetRotation
(
angle
,
2
*
angle
,
0.
);
m_GantryRoTransform
->
SetCenter
(
isocenter
);
m_ComposeTransform
->
Compose
(
m_GantryRoTransform
,
0
);
TransformType
::
OutputVectorType
focalpointtranslation
;
focalpointtranslation
[
0
]
=
-
isocenter
[
0
];
focalpointtranslation
[
1
]
=
400.0
-
isocenter
[
1
];
focalpointtranslation
[
2
]
=
-
isocenter
[
2
];
//std::cout << isocenter[0] << " " << isocenter[1] << " " << isocenter[2] << endl;
//std::cout << focalpointtranslation[0] << " " << focalpointtranslation[1] << " " << focalpointtranslation[2] << endl;
m_CamShiftTransform
->
SetTranslation
(
focalpointtranslation
);
m_ComposeTransform
->
Compose
(
m_CamShiftTransform
,
0
);
m_ComposeTransform
->
Compose
(
m_CamRotTransform
,
0
);
m_ComposeTransform
->
GetInverse
(
m_InverseTransform
);
itk
::
Matrix
<
double
>
m_matrix
=
m_InverseTransform
->
GetMatrix
();
itk
::
Vector
<
double
,
3
>
m_offset
=
m_InverseTransform
->
GetOffset
();
PointType
m_SourcePoint
;
m_SourcePoint
[
0
]
=
0.
;
m_SourcePoint
[
1
]
=
0.
;
m_SourcePoint
[
2
]
=
0.
;
PointType
tmp_SourceWorld
;
m_InverseTransform
->
TransformPoint
(
m_SourcePoint
);
tmp_SourceWorld
=
m_matrix
*
m_SourcePoint
+
m_offset
;
itk
::
Matrix
<
double
>
m_direction
=
image
->
GetDirection
();
//output_2d_image
double
output_spacing
[
3
];
output_spacing
[
0
]
=
0.3125
;
//sx
output_spacing
[
1
]
=
0.3125
;
//sy
output_spacing
[
2
]
=
1
;
double
output_size
[
3
];
output_size
[
0
]
=
dx
;
//dx
output_size
[
1
]
=
dy
;
//dy
output_size
[
2
]
=
1.0
;
itk
::
Vector
<
double
>
origin
;
origin
[
0
]
=
-
output_spacing
[
0
]
*
(
output_size
[
0
]
-
1.
)
/
2.
;
origin
[
1
]
=
-
output_spacing
[
1
]
*
(
output_size
[
1
]
-
1.
)
/
2.
;
origin
[
2
]
=
-
400
;
itk
::
Vector
<
double
>
corner00
;
itk
::
Vector
<
double
>
cornerOrigin01
;
itk
::
Vector
<
double
>
corner01
;
itk
::
Vector
<
double
>
cornerOrigin10
;
itk
::
Vector
<
double
>
corner10
;
cornerOrigin01
[
0
]
=
origin
[
0
]
+
output_spacing
[
0
]
*
(
output_size
[
0
]
-
1.
);
cornerOrigin01
[
1
]
=
origin
[
1
];
cornerOrigin01
[
2
]
=
origin
[
2
];
cornerOrigin10
[
0
]
=
origin
[
0
];
cornerOrigin10
[
1
]
=
origin
[
1
]
+
output_spacing
[
1
]
*
(
output_size
[
1
]
-
1.
);
cornerOrigin10
[
2
]
=
origin
[
2
];
corner00
=
m_matrix
*
origin
+
m_offset
;
corner01
=
m_matrix
*
cornerOrigin01
+
m_offset
;
corner10
=
m_matrix
*
cornerOrigin10
+
m_offset
;
SourceWorld
[
0
]
=
tmp_SourceWorld
[
0
];
SourceWorld
[
1
]
=
tmp_SourceWorld
[
1
];
SourceWorld
[
2
]
=
tmp_SourceWorld
[
2
];
itk
::
Vector
<
double
>
stepInX
;
stepInX
[
0
]
=
(
corner01
[
0
]
-
corner00
[
0
])
/
(
output_size
[
0
]);
stepInX
[
1
]
=
(
corner01
[
1
]
-
corner00
[
1
])
/
(
output_size
[
0
]);
stepInX
[
2
]
=
(
corner01
[
2
]
-
corner00
[
2
])
/
(
output_size
[
0
]);
itk
::
Vector
<
double
>
stepInY
;
stepInY
[
0
]
=
(
corner10
[
0
]
-
corner00
[
0
])
/
(
output_size
[
1
]);
stepInY
[
1
]
=
(
corner10
[
1
]
-
corner00
[
1
])
/
(
output_size
[
1
]);
stepInY
[
2
]
=
(
corner10
[
2
]
-
corner00
[
2
])
/
(
output_size
[
1
]);
DRRParameters
para
;
para
.
stepInX
[
0
]
=
stepInX
[
0
];
para
.
stepInX
[
1
]
=
stepInX
[
1
];
para
.
stepInX
[
2
]
=
stepInX
[
2
];
para
.
stepInY
[
0
]
=
stepInY
[
0
];
para
.
stepInY
[
1
]
=
stepInY
[
1
];
para
.
stepInY
[
2
]
=
stepInY
[
2
];
para
.
corner00
[
0
]
=
corner00
[
0
];
para
.
corner00
[
1
]
=
corner00
[
1
];
para
.
corner00
[
2
]
=
corner00
[
2
];
para
.
SourceWorld
[
0
]
=
SourceWorld
[
0
];
para
.
SourceWorld
[
1
]
=
SourceWorld
[
1
];
para
.
SourceWorld
[
2
]
=
SourceWorld
[
2
];
para
.
size
[
0
]
=
output_size
[
0
];
para
.
size
[
1
]
=
output_size
[
1
];
para
.
threshold
=
threshold
;
Image3D
image3d
;
image3d
.
image
=
cpp_object3D
;
image3d
.
isoCenter
[
0
]
=
isocenter
[
0
];
image3d
.
isoCenter
[
1
]
=
isocenter
[
1
];
image3d
.
isoCenter
[
2
]
=
isocenter
[
2
];
image3d
.
PixelSpacingCT
[
0
]
=
ctPixelSpacing
[
0
];
image3d
.
PixelSpacingCT
[
1
]
=
ctPixelSpacing
[
1
];
image3d
.
PixelSpacingCT
[
2
]
=
ctPixelSpacing
[
2
];
image3d
.
SizeCT
[
0
]
=
sizeCT
[
0
];
image3d
.
SizeCT
[
1
]
=
sizeCT
[
1
];
image3d
.
SizeCT
[
2
]
=
sizeCT
[
2
];
CUDAParamerters
cudaPara
;
cudaDeviceProp
prop
;
cudaGetDeviceProperties
(
&
prop
,
0
);
cudaPara
.
numThreads
=
prop
.
maxThreadsPerBlock
;
cudaPara
.
numBlocks
=
(
int
)
ceil
((
float
)
output_size
[
0
]
*
output_size
[
1
]
/
cudaPara
.
numThreads
);
loadOuputVariablesInGPUMemory
((
int
)
output_size
[
0
],
(
int
)
output_size
[
1
],
1
);
unsigned
char
*
object2d
=
calUCharDRRwithCUDA
(
cudaPara
,
para
);
freeAuxiliaryVariablesInGPUMemory
(
1
);
return
object2d
;
}
cuda-drr-socket/drr.h
View file @
175478db
...
...
@@ -43,7 +43,8 @@ public:
drr
(
std
::
string
path
);
~
drr
();
float
*
cudaDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
float
*
cudaFloatDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
unsigned
char
*
cudaUCharDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
itk
::
ImageSeriesReader
<
itk
::
Image
<
float
,
3
>
>::
Pointer
reader
;
itk
::
Image
<
float
,
3
>::
Pointer
image
;
...
...
cuda-drr-socket/main.cpp
View file @
175478db
...
...
@@ -10,9 +10,19 @@ using namespace std;
int
main
()
{
//_drr_cuda_();
cout
<<
"*******************************************************************************"
<<
endl
;
cout
<<
" Please connect the socket."
<<
endl
;
cout
<<
" Help: "
<<
endl
;
cout
<<
" Send the message
\"
CT(__folder_path__)
\"
for loading a CT sequence."
<<
endl
;
cout
<<
" Client will receive the message to inform if the operation is completed."
<<
endl
;
cout
<<
" Send the message
\"
PARA(rx,ry,rz,dx,dy,threshold,flag)
\"
for generating a DRR image."
<<
endl
;
cout
<<
" rx,ry,rz(float) are the rotation angle for the DRR image."
<<
endl
;
cout
<<
" dx,dy(int) are the width and height of the DRR image."
<<
endl
;
cout
<<
" threshold(float) can filter the required pixels for DRR images."
<<
endl
;
cout
<<
" flag(int) can decide the type of DRR array. 0 for float while 1 for unsigned char."
<<
endl
;
cout
<<
" Client will receive a DRR array if the operation is completed."
<<
endl
;
cout
<<
"*******************************************************************************"
<<
endl
;
socketManager
*
m
=
new
socketManager
;
m
->
open_socket
();
system
(
"pause"
);
return
0
;
}
\ No newline at end of file
cuda-drr-socket/socketManager.cpp
View file @
175478db
...
...
@@ -10,8 +10,8 @@ socketManager::socketManager()
socketManager
::~
socketManager
()
{
delete
CTpath
;
delete
rd
;
free
(
CTpath
)
;
free
(
rd
)
;
}
void
socketManager
::
get_CUDAdrr
(
const
char
*
path
)
...
...
@@ -23,11 +23,18 @@ void socketManager::get_CUDAdrr(const char * path)
rd
=
new
drr
(
path
);
}
float
*
socketManager
::
get_DRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
float
*
socketManager
::
get_
Float
DRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
{
if
(
rd
==
NULL
)
return
NULL
;
return
rd
->
cudaDRR
(
rx
,
ry
,
rz
,
dx
,
dy
,
threshold
);
return
rd
->
cudaFloatDRR
(
rx
,
ry
,
rz
,
dx
,
dy
,
threshold
);
}
unsigned
char
*
socketManager
::
get_UCharDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
{
if
(
rd
==
NULL
)
return
NULL
;
return
rd
->
cudaUCharDRR
(
rx
,
ry
,
rz
,
dx
,
dy
,
threshold
);
}
...
...
@@ -63,11 +70,17 @@ void socketManager::open_socket() {
}
CTpath
[
ret
-
4
]
=
'\0'
;
/////// 导入CT数据
get_CUDAdrr
(
CTpath
);
const
char
*
msg
=
"ct upload succeed"
;
try
{
get_CUDAdrr
(
CTpath
);
}
catch
(
exception
&
e
){
printf
(
"Failed to load the CT sequence
\n
"
);
break
;
}
const
char
*
msg
=
"ct imported successfully"
;
send
(
sClient
,
msg
,
strlen
(
msg
),
0
);
///////
printf
(
"
Received [%d bytes]:
'%s'
\n
"
,
ret
-
3
,
CTpath
);
printf
(
"
CT imported successfully :
'%s'
\n
"
,
ret
-
3
,
CTpath
);
}
else
if
(
szMessage
[
0
]
==
'P'
&&
szMessage
[
1
]
==
'A'
&&
szMessage
[
2
]
==
'R'
&&
szMessage
[
3
]
==
'A'
&&
...
...
@@ -75,11 +88,25 @@ void socketManager::open_socket() {
double
time
=
(
double
)
clock
()
/
CLOCKS_PER_SEC
;
float
*
data
=
handleString
(
szMessage
,
ret
);
//////// 生成DRR
float
*
object2D
=
get_DRR
(
data
[
0
],
data
[
1
],
data
[
2
],(
int
)
data
[
3
],(
int
)
data
[
4
],
data
[
5
]);
send
(
sClient
,
(
char
*
)
object2D
,
sizeof
(
float
)
*
(
int
)
data
[
3
]
*
(
int
)
data
[
4
],
0
);
try
{
if
((
int
)
data
[
6
]
==
0
)
{
float
*
object2D
=
get_FloatDRR
(
data
[
0
],
data
[
1
],
data
[
2
],
(
int
)
data
[
3
],
(
int
)
data
[
4
],
data
[
5
]);
send
(
sClient
,
(
char
*
)
object2D
,
sizeof
(
float
)
*
(
int
)
data
[
3
]
*
(
int
)
data
[
4
],
0
);
free
(
object2D
);
}
else
{
unsigned
char
*
object2D
=
get_UCharDRR
(
data
[
0
],
data
[
1
],
data
[
2
],
(
int
)
data
[
3
],
(
int
)
data
[
4
],
data
[
5
]);
send
(
sClient
,
(
char
*
)
object2D
,
sizeof
(
unsigned
char
)
*
(
int
)
data
[
3
]
*
(
int
)
data
[
4
],
0
);
free
(
object2D
);
}
}
catch
(
exception
&
e
)
{
printf
(
"Failed to generate DRR, please check your parameters.
\n
"
);
break
;
}
////////
printf
(
"Create DRR image succeed: [%f, %f, %f, %d, %d, %f]
\n
"
,
data
[
0
],
data
[
1
],
data
[
2
],
(
int
)
data
[
3
],
(
int
)
data
[
4
],
data
[
5
]);
std
::
cout
<<
"
during drr :"
<<
(
double
)
clock
()
/
CLOCKS_PER_SEC
-
time
<<
std
::
endl
;
printf
(
"Create DRR image succeed: [%f, %f, %f, %d, %d, %f]
"
,
data
[
0
],
data
[
1
],
data
[
2
],
(
int
)
data
[
3
],
(
int
)
data
[
4
],
data
[
5
]);
std
::
cout
<<
"
, time :"
<<
(
double
)
clock
()
/
CLOCKS_PER_SEC
-
time
<<
"s"
<<
std
::
endl
;
}
break
;
}
...
...
@@ -89,12 +116,12 @@ void socketManager::open_socket() {
}
float
*
socketManager
::
handleString
(
char
*
szMessage
,
int
ret
)
{
float
data
[
6
];
float
data
[
7
];
int
k
=
0
;
char
*
buf
=
new
char
[
1024
];
int
j
=
0
;
for
(
int
i
=
5
;
i
<
ret
-
1
;
i
++
)
{
if
(
szMessage
[
i
]
!=
','
&&
k
<
6
)
{
if
(
szMessage
[
i
]
!=
','
&&
k
<
7
)
{
buf
[
j
]
=
szMessage
[
i
];
j
++
;
}
...
...
cuda-drr-socket/socketManager.h
View file @
175478db
...
...
@@ -19,7 +19,8 @@ public:
~
socketManager
();
void
open_socket
();
void
get_CUDAdrr
(
const
char
*
path
);
float
*
get_DRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
float
*
get_FloatDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
unsigned
char
*
get_UCharDRR
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
);
float
*
handleString
(
char
*
szMessage
,
int
ret
);
private
:
...
...
cuda-drr-socket/x64/Debug/vc141.idb
View file @
175478db
No preview for this file type
test/main.cpp
View file @
175478db
...
...
@@ -2,6 +2,8 @@
#include <Windows.h>
#include <stdio.h>
#include <string>
#include <opencv2\opencv.hpp>
#include <opencv2/imgproc/types_c.h>
#pragma warning(disable:4996)
#pragma comment(lib, "WS2_32.lib")
using
namespace
std
;
...
...
@@ -32,7 +34,7 @@ void sendCTPath(const char* path) {
WSACleanup
();
}
float
*
sendPara
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
)
{
void
sendPara
(
float
rx
,
float
ry
,
float
rz
,
int
dx
,
int
dy
,
float
threshold
,
int
flag
)
{
WSADATA
wsaData
;
SOCKET
sClient
;
SOCKADDR_IN
server
;
...
...
@@ -44,28 +46,32 @@ float* sendPara(float rx, float ry, float rz, int dx, int dy, float threshold) {
server
.
sin_port
=
htons
(
PORT
);
server
.
sin_addr
.
s_addr
=
inet_addr
(
SERVER_ADDRESS
);
connect
(
sClient
,
(
struct
sockaddr
*
)
&
server
,
sizeof
(
SOCKADDR_IN
));
string
msg
=
"PARA("
+
to_string
(
rx
)
+
","
+
to_string
(
ry
)
+
","
+
to_string
(
rz
)
+
","
+
to_string
(
dx
)
+
","
+
to_string
(
dy
)
+
","
+
to_string
(
threshold
)
+
")"
;
string
msg
=
"PARA("
+
to_string
(
rx
)
+
","
+
to_string
(
ry
)
+
","
+
to_string
(
rz
)
+
","
+
to_string
(
dx
)
+
","
+
to_string
(
dy
)
+
","
+
to_string
(
threshold
)
+
","
+
to_string
(
flag
)
+
")"
;
const
char
*
szMessage
=
msg
.
c_str
();
send
(
sClient
,
szMessage
,
strlen
(
szMessage
),
0
);
float
*
recvbuf
=
(
float
*
)
malloc
(
sizeof
(
float
)
*
dx
*
dy
);
recv
(
sClient
,
(
char
*
)
recvbuf
,
sizeof
(
float
)
*
dx
*
dy
,
0
);
unsigned
char
*
recvbuf
=
(
unsigned
char
*
)
malloc
(
sizeof
(
unsigned
char
)
*
dx
*
dy
);
recv
(
sClient
,
(
char
*
)
recvbuf
,
sizeof
(
unsigned
char
)
*
dx
*
dy
,
0
);
printf
(
"receive drr(%f,%f,%f,%d,%d,%f)
\n
"
,
rx
,
ry
,
rz
,
dx
,
dy
,
threshold
);
closesocket
(
sClient
);
WSACleanup
();
return
recvbuf
;
cv
::
Mat
gray1_mat
(
dx
,
dy
,
CV_8UC1
,
recvbuf
);
cv
::
imshow
(
"123"
,
gray1_mat
);
cv
::
waitKey
(
0
);
}
int
main
()
{
sendCTPath
(
"D:/
QT/CTxray/
CT1/CT"
);
float
*
object2D
=
sendPara
(
90
,
0
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
90
,
25
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
90
,
0
,
77
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
40
,
4
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
90
,
66
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
30
,
0
,
60
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
45
,
45
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
45
,
70
,
0
,
1024
,
1024
,
0
);
object2D
=
sendPara
(
77
,
60
,
0
,
1024
,
1024
,
0
);
sendCTPath
(
"D:/CT1/CT"
);
sendPara
(
90
,
0
,
0
,
1024
,
1024
,
0
,
1
);
sendPara
(
90
,
25
,
0
,
512
,
512
,
0
,
1
);
sendPara
(
90
,
0
,
77
,
1024
,
1024
,
0
,
1
);
sendPara
(
40
,
4
,
0
,
1024
,
1024
,
0
,
1
);
sendPara
(
90
,
66
,
0
,
1024
,
1024
,
0
,
1
);
sendPara
(
30
,
0
,
60
,
1024
,
1024
,
0
,
1
);
sendPara
(
45
,
45
,
0
,
1024
,
1024
,
0
,
1
);
sendPara
(
45
,
70
,
0
,
512
,
512
,
0
,
1
);
sendPara
(
77
,
60
,
0
,
512
,
512
,
0
,
1
);
system
(
"pause"
);
return
0
;
}
\ No newline at end of file
x64/Release/cuda-drr-socket.iobj
View file @
175478db
No preview for this file type
x64/Release/cuda-drr-socket.ipdb
View file @
175478db
No preview for this file type
Write
Preview
Markdown
is supported
0%
Try again
or
attach a new file
Attach a file
Cancel
You are about to add
0
people
to the discussion. Proceed with caution.
Finish editing this message first!
Cancel
Please
register
or
sign in
to comment