質問編集履歴
4
コードの修正
title
CHANGED
|
File without changes
|
body
CHANGED
|
@@ -50,9 +50,12 @@
|
|
|
50
50
|
float* a, * b, * c, * d_a, * d_b, * d_c;
|
|
51
51
|
|
|
52
52
|
//ホスト側の配列を用意
|
|
53
|
-
a = (float*)malloc(N * sizeof(float));
|
|
53
|
+
//a = (float*)malloc(N * sizeof(float));
|
|
54
|
-
b = (float*)malloc(N * sizeof(float));
|
|
54
|
+
//b = (float*)malloc(N * sizeof(float));
|
|
55
|
-
c = (float*)malloc(N * sizeof(float));
|
|
55
|
+
//c = (float*)malloc(N * sizeof(float));
|
|
56
|
+
cudaMallocHost(&a, N * sizeof(float));
|
|
57
|
+
cudaMallocHost(&b, N * sizeof(float));
|
|
58
|
+
cudaMallocHost(&c, N * sizeof(float));
|
|
56
59
|
|
|
57
60
|
//デバイス側の配列を用意
|
|
58
61
|
cudaMalloc(&d_a, N * sizeof(float));
|
|
@@ -68,40 +71,57 @@
|
|
|
68
71
|
|
|
69
72
|
//時間計測開始
|
|
70
73
|
std::chrono::system_clock::time_point start, memcpyh2d, memcpyd2h, gpus, gpue, cpu;
|
|
71
|
-
start = std::chrono::system_clock::now();
|
|
72
74
|
std::cout << "-----TimeStart-----" << std::endl;
|
|
73
75
|
|
|
74
|
-
|
|
76
|
+
for (int i = 0; i < 10; i++) {
|
|
75
|
-
cudaMemcpy(d_a, a, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
76
|
-
cudaMemcpy(d_b, b, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
77
|
-
cudaMemcpy(d_c, c, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
78
77
|
|
|
79
|
-
//ホストの配列をデバイス側にコピーするのにかかった時間
|
|
80
|
-
|
|
78
|
+
start = std::chrono::system_clock::now();
|
|
81
79
|
|
|
82
|
-
|
|
80
|
+
//ホスト側の配列の内容をデバイス側にコピー
|
|
81
|
+
cudaMemcpy(d_a, a, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
83
|
-
|
|
82
|
+
cudaMemcpy(d_b, b, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
83
|
+
cudaMemcpy(d_c, c, N * sizeof(float), cudaMemcpyHostToDevice);
|
|
84
84
|
|
|
85
|
-
|
|
85
|
+
//ホストの配列をデバイス側にコピーするのにかかった時間
|
|
86
|
-
//ブロックの総数（gridsize）をN/512用意する
|
|
87
|
-
//したがって総スレッド数は blocksize × gridsize = N 個
|
|
88
|
-
dim3 block(blocksize, 1, 1);
|
|
89
|
-
|
|
86
|
+
memcpyh2d = std::chrono::system_clock::now();
|
|
90
87
|
|
|
91
|
-
|
|
88
|
+
//スレッドの設定
|
|
92
|
-
|
|
89
|
+
int blocksize = 256;
|
|
93
90
|
|
|
94
|
-
|
|
91
|
+
//ブロックあたりのスレッド数（blocksize)を512、
|
|
92
|
+
//ブロックの総数（gridsize）をN/512用意する
|
|
95
|
-
|
|
93
|
+
//したがって総スレッド数は blocksize × gridsize = N 個
|
|
96
|
-
|
|
94
|
+
dim3 block(blocksize, 1, 1);
|
|
95
|
+
dim3 grid(N / block.x, 1, 1);
|
|
97
96
|
|
|
98
|
-
|
|
97
|
+
//GPUの計算時間計測開始
|
|
99
|
-
|
|
98
|
+
gpus = std::chrono::system_clock::now();
|
|
100
99
|
|
|
101
|
-
|
|
100
|
+
// カーネル関数の呼び出し
|
|
102
|
-
|
|
101
|
+
vec_sum_k << <grid, block >> > (2.0f, d_a, d_b, d_c);
|
|
102
|
+
cudaDeviceSynchronize();
|
|
103
103
|
|
|
104
|
+
//GPUの計算時間計測終了
|
|
105
|
+
gpue = std::chrono::system_clock::now();
|
|
106
|
+
|
|
107
|
+
//計算結果をホストへコピー
|
|
108
|
+
cudaMemcpy(c, d_c, N * sizeof(float), cudaMemcpyDeviceToHost);
|
|
109
|
+
|
|
110
|
+
//計算結果をホストにコピー。GPUの計算終わり。
|
|
111
|
+
memcpyd2h = std::chrono::system_clock::now();
|
|
112
|
+
|
|
113
|
+
// 計測時間の表示
|
|
114
|
+
std::cout << "-------GPULoop: " << i << " -------" << std::endl;
|
|
115
|
+
double HtDtime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(memcpyh2d - start).count());
|
|
116
|
+
std::cout << "memcpy(HtD): " << HtDtime << " microsec" << std::endl;
|
|
117
|
+
double gputime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(gpue - gpus).count());
|
|
118
|
+
std::cout << "gpu: " << gputime << " microsec" << std::endl;
|
|
119
|
+
double DtHtime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(memcpyd2h - gpue).count());
|
|
120
|
+
std::cout << "memcpy(DtH): " << DtHtime << " microsec" << std::endl;
|
|
121
|
+
|
|
122
|
+
}
|
|
123
|
+
|
|
104
|
-
//計算結果をホストにコピー。GPUの計算終わり。CPUの計算
|
|
124
|
+
//計算結果をホストにコピー。GPUの計算終わり。CPUの計算開始
|
|
105
125
|
memcpyd2h = std::chrono::system_clock::now();
|
|
106
126
|
|
|
107
127
|
vec_sum_c(2.0f, c, a, b);
|
|
@@ -116,12 +136,7 @@
|
|
|
116
136
|
std::cout << "MaxError: : " << maxError << std::endl;
|
|
117
137
|
|
|
118
138
|
//計測時間の表示
|
|
119
|
-
double HtDtime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(memcpyh2d - start).count());
|
|
120
|
-
std::cout << "memcpy(HtD): " << HtDtime << " microsec" << std::endl;
|
|
121
|
-
double gputime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(gpue - gpus).count());
|
|
122
|
-
std::cout << "
|
|
139
|
+
std::cout << "-------CPU-------" << std::endl;
|
|
123
|
-
double DtHtime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(memcpyd2h - gpue).count());
|
|
124
|
-
std::cout << "memcpy(DtH): " << DtHtime << " microsec" << std::endl;
|
|
125
140
|
double cputime = static_cast<double>(std::chrono::duration_cast<std::chrono::microseconds>(cpu - memcpyd2h).count());
|
|
126
141
|
std::cout << "cpu: " << cputime << " microsec" << std::endl;
|
|
127
142
|
|
3
タイトルの修正
title
CHANGED
|
@@ -1,1 +1,1 @@
|
|
|
1
|
-
【CUDA】cudaMemcpyに時間がかかり
|
|
1
|
+
【CUDA】cudaMemcpyに時間がかかり速度が出ないときの対処法
|
body
CHANGED
|
File without changes
|
2
コードの修正
title
CHANGED
|
File without changes
|
body
CHANGED
|
@@ -32,7 +32,7 @@
|
|
|
32
32
|
void vec_sum_c(float k, float c[], float a[], float b[]) {
|
|
33
33
|
|
|
34
34
|
for (int i = 0; i < N; i++) {
|
|
35
|
-
c[
|
|
35
|
+
c[i] = k * a[i] + b[i];
|
|
36
36
|
}
|
|
37
37
|
}
|
|
38
38
|
|
1
「実現したいこと」を一部修正
title
CHANGED
|
File without changes
|
body
CHANGED
|
@@ -141,7 +141,7 @@
|
|
|
141
141
|
### 実現したいこと
|
|
142
142
|
|
|
143
143
|
メモリコピーの時間を含めてCPUより高速で動作するようなCUDAプログラムを作成するためにはどのような方法があるでしょうか？
|
|
144
|
-
また計算自体は高速化されているとはいえ、期待したほど(十倍~数百倍以上)ではないのも気になります。計算の高速化のためにはどのような工夫ができるでしょうか？
|
|
144
|
+
また計算自体は高速化されているとはいえ、期待したほど(十倍~数百倍以上)ではないのも気になります。上記サイトの結果では少なくとも3000倍の高速化に成功していたようです。私のプログラムがそれほど高速化されていない理由は何でしょうか？ また計算の高速化のためにはどのような工夫ができるでしょうか？
|
|
145
145
|
|
|
146
146
|
### 補足情報（FW/ツールのバージョンなど）
|
|
147
147
|
|