スポンサーリンク

| キーワード:

CUDAで直線を書くことを考える(2)(書いてみる)

前回のC++のプログラムをCUDAで書き直してみる。

流石に無駄が多すぎることは自分でもわかるので次回はもっとましな方法を考える。

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>
#include <stdlib.h>

///////////////////////////////////////////////
// GPU側 ////////////////////////////////////// 
__global__ void line(
  unsigned char* pimage,
  const int width,
  const int height,
  const int sx,
  const int sy,
  const int ex,
  const int ey) {

  //アクセス法
  //このスレッドが担当する画素の位置を二次元座標で求める
  int xpos = blockIdx.x * blockDim.x + threadIdx.x;
  int ypos = blockIdx.y * blockDim.y + threadIdx.y;


  int pos = ypos * width + xpos;
  if (xpos >= width || ypos >= height)return;
  if (pos < 0 || pos >= width * height)return;

  double a = (ey - sy) / double(ex - sx);//傾き
  double b = sy - a * sx;//切片
  double y = int(a * xpos + b);

  if (ypos < min(sy, ey))return;
  if (xpos < min(sx, ex))return;
  if (ypos > max(sy, ey))return;
  if (xpos > max(sx, ex))return;
  
  if (y == ypos) {
    pimage[pos] = 1;
  }
  else {
    pimage[pos] = 0;
  }

}
/////
void pbmP1_Write(const char* const fname, const int width, const int height, const unsigned char* const p);

int main()
{
  int width = 100;
  int height = 50;

  int sx = 98; // 線分の始点・終点を設定
  int sy = 48;
  int ex = 1;
  int ey = 3;

  // 最小単位 合計で512未満
  //10×10の領域に分けて計算する
  dim3 block(10, 10);

  //グリッド数
  // スレッド数はblock数×grid数なので(10x100)x(10x100)=1000x1000
  // 多すぎるが今回はそこまで考えない
  dim3 grid(100, 100);

  unsigned char* p_cpu = new unsigned char[width*height];
  for (int i = 0; i < width*height; i++) {
    p_cpu[i] = 1;
  }

  unsigned char* p_gpu;//GPU側メモリ確保
  cudaError_t ce = cudaMalloc((void**)&p_gpu, width*height);
  printf("error:%d\n", ce);
  line << <grid, block >> > (p_gpu, width, height,sx,sy,ex,ey);

  //GPU→CPU側へメモリコピー
  cudaMemcpy(p_cpu, p_gpu, width*height, cudaMemcpyDeviceToHost);
  cudaFree(p_gpu);


  ///////////////////////////////////////////////
  // 結果出力

  pbmP1_Write("C:\\dev\\cudaline.pbm", width, height, p_cpu);

  delete[] p_cpu;

  return 0;
}

//! @brief PBM(1byte,テキスト)を書き込む
//! @param [in] fname ファイル名
//! @param [in] width 画像の幅
//! @param [in] height 画像の高さ
//! @param [in] p 画像のメモリへのアドレス
//! @details 1画素1Byteのメモリを渡すと、0,1テキストでファイル名fnameで書き込む
void pbmP1_Write(const char* const fname, const int width, const int height, const unsigned char* const p) { // PPM ASCII

  FILE* fp = fopen(fname, "wb");
  fprintf(fp, "P1\n%d\n%d\n", width, height);

  size_t k = 0;
  for (size_t i = 0; i < (size_t)height; i++) {
    for (size_t j = 0; j < (size_t)width; j++) {
      fprintf(fp, "%d ", p[k] ? 1 : 0);
      k++;
    }
    fprintf(fp, "\n");
  }

  fclose(fp);
}

コメントを残す

メールアドレスが公開されることはありません。

日本語が含まれない投稿は無視されますのでご注意ください。(スパム対策)


この記事のトラックバックURL: