あかすくぱるふぇ

同人サークル「あかすくぱるふぇ」のブログです。

exeから呼び出す関数には、__declspec(dllexport)をつける。
ただし、defファイルを使えば__declspec(dllexport)の記述を省略できる。

名前装飾を無効化するために、extern "C"もつける。

DLLのリンクには、明示的リンクと暗黙的リンクとがある。
明示的リンクは、LoadLibrary()などで実現する。
暗黙的リンクは、関数宣言に、__declspec(import)をつけ、コンパイル時にlibファイルをリンクすることで実現する。

暗黙的リンクで利用するヘッダファイルは、
#ifdef MYDLL
#define DLLAPI extern "C" __declspec(dllexport)
#else
#define DLLAPI extern "C" __declspec(dllimport)
#endif
などとしておくと、DLL側とEXE側で共用できる。

以下は参考ページ。
http://exlight.net/devel/windows/dll/windll.html
http://akasuku.blog.jp/archives/43379993.html
http://fa11enprince.hatenablog.com/entry/2014/06/20/015808

以上です。

CUDA Sampleでよく行われている処理。
ただ、CUDA Sampleはごちゃごちゃしていてよくわからないので、最小構成を抜き出してみました。
int main(int argc, char *argv[])
{
// 画像ファイル入力
cvImg = cv::imread("Lenna.bmp", 0);

// 初期化
init(argc, argv);

glutMainLoop();

// 終了処理
finalize();

return 0;
}
main関数。
画像ファイルの入力と初期化を行った後に、glutMainLoop()を実行します。

void init(int argc, char *argv[])
{
size_t dataSize = cvImg.cols * cvImg.rows * sizeof(unsigned char);

glInit(argc, argv); // GLの初期化
GenTexture(); // テクスチャの生成
GenAndRegisterPBO(dataSize); // PBOの生成とCUDAへの登録
   
// 入力画像のGPU側メモリ確保
  checkCudaErrors(cudaMalloc((unsigned char**)&dImgIn, dataSize));
}
初期化関数です。
以下に各関数を載せます。

void glInit(int argc, char *argv[])
{
glutInit(&argc, argv);
glutInitWindowSize(cvImg.cols, cvImg.rows);
glutInitDisplayMode(GLUT_RGBA | GLUT_DOUBLE);
glutCreateWindow(argv[0]);
glewInit();
glutDisplayFunc(display);
}

void GenTexture()
{
glGenTextures(1, &tex);
glBindTexture(GL_TEXTURE_2D, tex);
glTexImage2D(GL_TEXTURE_2D, 0, GL_LUMINANCE, cvImg.cols, cvImg.rows,
      0, GL_LUMINANCE, GL_UNSIGNED_BYTE, NULL);
glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
}

void GenAndRegisterPBO(const size_t dataSize)
{
glGenBuffers(1, &pbo);
glBindBuffer(GL_PIXEL_UNPACK_BUFFER, pbo);
glBufferData(GL_PIXEL_UNPACK_BUFFER, dataSize, 0, GL_STREAM_DRAW);
glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);
cudaGraphicsGLRegisterBuffer(&cuda_pbo_resource, pbo,
      cudaGraphicsMapFlagsWriteDiscard);
}
glInit()はGLUT, GLEWのおなじみの処理。
GenTexture()はテクスチャ生成のおなじみの処理。
GenAndRegisterPBO()はPBOを生成し、そのPBOをCUDAに登録しています。
CUDAへの登録はcudaGraphicsGLRegisterBuffer()で行います。
cudaGraphicsMapFlagsWriteDiscardはPBOがCUDAからWriteOnlyであるというフラグです。

void display()
{
// カーネル関数で書き換える画像(PBO)のポインタを取得
unsigned char *dImgOut;
size_t dataSizeTmp;
checkCudaErrors(cudaGraphicsMapResources(1, &cuda_pbo_resource, 0));
checkCudaErrors(cudaGraphicsResourceGetMappedPointer(
(void **)&dImgOut, &dataSizeTmp, cuda_pbo_resource));

// 入力画像をGPUのグローバルメモリに転送
size_t dataSize = cvImg.cols * cvImg.rows * sizeof(unsigned char);
checkCudaErrors(cudaMemcpy(dImgIn, cvImg.data, dataSize, cudaMemcpyHostToDevice));

// カーネル関数の実行
dim3 grid(cvImg.rows);
dim3 block(1024);
invertKernel << <grid, block >> >(dImgOut, dImgIn, cvImg.cols);

checkCudaErrors(cudaGraphicsUnmapResources(1, &cuda_pbo_resource, 0));

glBindTexture(GL_TEXTURE_2D, tex);
glBindBuffer(GL_PIXEL_UNPACK_BUFFER, pbo);
glTexSubImage2D(GL_TEXTURE_2D, 0, 0, 0, cvImg.cols, cvImg.rows,
GL_LUMINANCE, GL_UNSIGNED_BYTE, 0);
glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);

glEnable(GL_TEXTURE_2D);
glBegin(GL_QUADS);
{
glTexCoord2f(0, 1);
glVertex2f(-1, -1);
glTexCoord2f(1, 1);
glVertex2f(1, -1);
glTexCoord2f(1, 0);
glVertex2f(1, 1);
glTexCoord2f(0, 0);
glVertex2f(-1, 1);
}
glEnd();

glBindTexture(GL_TEXTURE_2D, 0);
glutSwapBuffers();
}
display関数です。
PBOのポインタ取得、入力画像のGPU転送、カーネル実行、テクスチャマッピングを行っています。
処理結果をCPUに戻さずに、PBOからテクスチャにコピー(GPU内で閉じてる)して描画しているのがポイントです。

念のため、カーネル関数の一例も載せておきます。
__global__ void invertKernel(unsigned char *out, const unsigned char *in, const int width)
{
for (int i = threadIdx.x; i < width; i += blockDim.x) {
int index = blockIdx.x * width + i;
out[index] = 255 - in[index];
}
}

以上です。

・分析、設計、実装、テストといった開発工程がどれも、途切れることなく続く連続的な取り組みになる。
・イテレーションは全力疾走する短距離走で、この期間を通じて、ユーザーストーリーをリリース可能な動くソフトウェアに変換する。
・リリースはMinimalかつMarketableであるべき。システムの価値の80%が機能の20%からもたらされている。
・コードはいつもばっちり結合した状態にしておく。最後まで全体として動作するかわからない状況に甘んじてはいけない。
・メンバーの役割は気にしない。気にかけるのは、その役割によって果たされる仕事が、チームの誰かによってちゃんとこなされるようになっているかどうかだけ。
・顧客に積極的にかかわってもらう。
・「自分がいる理由」、「目的」、「顧客」、「主要課題」、「判断者」は少なくとも明確にする。
・見積もりを行うのは、ターゲットが達成可能な程度に現実的なものかどうか判断するためである。
・見積もりは理想日単位ではなく、相対的なポイントで表現した方がよい。
・リファクタリングは常に細かくやり続ける。
・カバレッジ100%を目指すべきではない。網羅性ではなく、リリース準備が整っていると確信を持つことが大事。
・ビルド時間は10分以内が望ましい。
・参考書:レガシーコード改善ガイド、テスト駆動開発入門

↑このページのトップヘ