HIRASE CONNECTION WK

[CUDA] CUDA CAPS

T.Hirase — Sat, 03 Oct 2009 01:10:00 GMT

うちのMacBookに載っているGeForce 9400Mで、どの程度、CUDA遊びができるか知りたかったので、コーディング。

#include <stdio.h>

int main(void)
{
  int num_devices;
	cudaGetDeviceCount(&num_devices);
	printf("number of devices : %d\n\n", num_devices);
	
	for (int device_id = 0; device_id < num_devices; ++device_id)
	{
		cudaDeviceProp device_properties;
		cudaGetDeviceProperties(&device_properties, device_id);
		
    printf("name                     : %s\n" , device_properties.name                    ); /* char[256] */
    printf("totalGlobalMem           : %zd\n", device_properties.totalGlobalMem          ); /* size_t    */
    printf("sharedMemPerBlock        : %zd\n", device_properties.sharedMemPerBlock       ); /* size_t    */
    printf("regsPerBlock             : %d\n" , device_properties.regsPerBlock            ); /* int       */
    printf("warpSize                 : %d\n" , device_properties.warpSize                ); /* int       */
    printf("memPitch                 : %zd\n", device_properties.memPitch                ); /* size_t    */
    printf("maxThreadsPerBlock       : %d\n" , device_properties.maxThreadsPerBlock      ); /* int       */
    printf("maxThreadsDim            : {%d,%d,%d}\n", device_properties.maxThreadsDim[0], device_properties.maxThreadsDim[1], device_properties.maxThreadsDim[2]); /* int[3]    */
    printf("maxGridSize              : {%d,%d,%d}\n", device_properties.maxGridSize[0], device_properties.maxGridSize[1], device_properties.maxGridSize[2]); /* int[3]    */
    printf("clockRate                : %d\n" , device_properties.clockRate               ); /* int       */
    printf("totalConstMem            : %zd\n", device_properties.totalConstMem           ); /* size_t    */
    printf("major                    : %d\n" , device_properties.major                   ); /* int       */
    printf("minor                    : %d\n" , device_properties.minor                   ); /* int       */
    printf("textureAlignment         : %zd\n", device_properties.textureAlignment        ); /* size_t    */
    printf("deviceOverlap            : %d\n" , device_properties.deviceOverlap           ); /* int       */
    printf("multiProcessorCount      : %d\n" , device_properties.multiProcessorCount     ); /* int       */
    printf("kernelExecTimeoutEnabled : %d\n" , device_properties.kernelExecTimeoutEnabled); /* int       */
    printf("integrated               : %d\n" , device_properties.integrated              ); /* int       */
    printf("canMapHostMemory         : %d\n" , device_properties.canMapHostMemory        ); /* int       */
    printf("computeMode              : %d\n" , device_properties.computeMode             ); /* int       */
	}

  return 0;
}

結果は、こんな感じ。

number of devices : 1

name                     : GeForce 9400M
totalGlobalMem           : 266010624
sharedMemPerBlock        : 16384
regsPerBlock             : 8192
warpSize                 : 32
memPitch                 : 262144
maxThreadsPerBlock       : 512
maxThreadsDim            : {512,512,64}
maxGridSize              : {65535,65535,1}
clockRate                : 1100000
totalConstMem            : 65536
major                    : 1
minor                    : 1
textureAlignment         : 256
deviceOverlap            : 0
multiProcessorCount      : 2
kernelExecTimeoutEnabled : 1
integrated               : 1
canMapHostMemory         : 1
computeMode              : 0

"Compute Capability" は "1.1" か（major.minor)。ちょっと古いな。"multiProcessorCount" = 2 ってことは、"Scalar Processor" = 2 * 8 = 16 ってことか。まぁ、ぼちぼちかな。Tesla C1060に比べれば少ないけど、まぁ、まぁ、実験くらいはできるなぁ。大学で使ってたXeon×8マシンと数的には一緒か（いや、全然意味が違うけど）。16 ってでも微妙なんだよなぁ。もうちょっと、せめて64SPくらいは欲しいところだぁ。

totalGlobalMem =2GBになってるけど、9400Mってメインメモリと共有だから、これ、メインメモリのサイズだよなぁ。騙されるとこだった。

[CUDA] はじめてのCUDA的な。

T.Hirase — Fri, 02 Oct 2009 01:00:00 GMT

CUDA開始。すごく入り口は簡単だけど、並列思考が求められる。

#include <stdio.h>

__global__ void krnlTest(float* dstv, float* srcv)
{
  int idx = blockDim.x * blockIdx.x + threadIdx.x;  // thread index
  dstv[idx] = srcv[idx] * srcv[idx];
}

int main(void)
{
  const int    num     = 512;
  const size_t vecsize = num * sizeof(float);

  // allocate host memory
  float* host_dstv = (float*)malloc(vecsize);
  float* host_srcv = (float*)malloc(vecsize);

  // initialize input data
  for (int i = 0; i < num; ++i) {
    host_srcv[i] = i;
  }

  // allocate kernel memory
  float* krnl_dstv;
  float* krnl_srcv;
  cudaMalloc((void**)&krnl_dstv, vecsize);
  cudaMalloc((void**)&krnl_srcv, vecsize);

  // transfer input data
  cudaMemcpy(krnl_srcv, host_srcv, vecsize, cudaMemcpyHostToDevice);

  // run (asynchronous)
  krnlTest<<<1,512>>>(krnl_dstv, krnl_srcv);

  // wait and transfer output data
  cudaMemcpy(host_dstv, krnl_dstv, vecsize, cudaMemcpyDeviceToHost);

  // free kernel memory
  cudaFree(krnl_dstv);
  cudaFree(krnl_srcv);

  // print
  for (int i = 0; i < num; ++i) {
    printf("%f\n", host_dstv[i]);
  }

  // free host memory
  free(host_dstv);
  free(host_srcv);

  return 0;
}

cuda?関数はCUDA組み込み関数。blockDimなどは組み込み変数。そしてその行はスレッドインデックスを取得するためのおまじない。あと、__global__はCUDA用コンパイルさせるための指定子。ホストPCからCUDA（カーネル）側の関数を呼び出すときは、不思議なおまじない「<<<d,b,n,m>>>」というのが関数名にくっつく。

何か間違えているところあったら、Comments are welcomeです。

[C] 1から10までの総和をマクロでエレファントに解く。

T.Hirase — Wed, 30 Sep 2009 22:30:00 GMT

うすたー：1+2+…+10

うすたー：[C++] １から１０まで足す

#include <stdio.h>

#define SIGMA(X) SIGMA_##X
#define SIGMA_10 (10 + SIGMA_9)
#define SIGMA_9  ( 9 + SIGMA_8)
#define SIGMA_8  ( 8 + SIGMA_7)
#define SIGMA_7  ( 7 + SIGMA_6)
#define SIGMA_6  ( 6 + SIGMA_5)
#define SIGMA_5  ( 5 + SIGMA_4)
#define SIGMA_4  ( 4 + SIGMA_3)
#define SIGMA_3  ( 3 + SIGMA_2)
#define SIGMA_2  ( 2 + SIGMA_1)
#define SIGMA_1  ( 1          )

int main(void)
{
  printf("%d\n", SIGMA(10));
  return 0;
}

もう素直に数字並べて足し算した方がマシなレベル。

[C] C99

T.Hirase — Fri, 25 Sep 2009 16:14:00 GMT

もうすでに10年まえの話ですが、C言語の新しい規格「C99」(ISO/IEC 9899:1999)が制定されました。

それなのに、それなのに、いまだに、こんな（↓）文法を使ったことはない。

static int hoge[100] = {[10]=100, [20]=200,}

int ret = 0;
ret = foo();
int j = bar();

一つ目は、「Designated Initializer」という初期化方法。二つ目は変数宣言位置がブロック先頭になくてもいいという法。

なぜ、使わないって、そりゃ、Visual C++ コンパイラでエラーになるからですよ。痛い。痛いです。某オープンソースプロジェクトでガンガン使われてる。痛い、痛いです。そろそろ宣言位置くらい許してくれません？　何か副作用あります？

※注意：C++の話じゃないですよ。

[C++] new & delete 血の掟

T.Hirase — Sun, 30 Aug 2009 01:35:00 GMT

今日のって、もう昨日の横浜勉強会から時既に12時間ほどが経過していますが、忘れないうちにεπιστημηさんのセッションで出てきたnew / delete 血の掟を簡単にまとめておきます。最後のケース６は、勝手に追加。やっちゃダメですよね？？

class Hoge { public: void func(){} };

int main()
{
	{ // 正しい new/delete の使い方
		Hoge* p = new Hoge();
		p->func();
		delete p;
	}
	
	{ // 正しい new[]/delete[] の使い方
		Hoge* p = new Hoge[10];
		p[0].func();
		delete[] p;
	}
	
	{ // 間違いケース１：お別れの挨拶を
		Hoge* p = new Hoge();
		p->func();
	}
	
	{ // 間違いケース２：なりすましに注意
		Hoge* p = new Hoge();
		p->func();
		p = new Hoge();
		p->func();
		delete p;
	}
	
	{ // 間違いケース３：お別れし過ぎ
		Hoge* p = new Hoge();
		p->func();
		delete p;
		// ... いろいろあって ...
		delete p;
	}
	
	{ // 間違いケース４：未練を残すな
		Hoge* p = new Hoge();
		delete p;
		p->func();
	}
	
	{ // 間違いケース５：強欲
		Hoge* p = new Hoge[10];
		p[10].func();
		delete[] p;
	}
	
	{ // 間違いケース６：すべて清算せよ
		Hoge* p = new Hoge[10];
		delete p;
	}
	
	return 0;
}

間違えたときは、核ミサイルが発射され、家賃の振込が物凄い額で振り込まれ、消したはずの恥ずかしいアレが消えなかったり、いやむしろ、まぁくんのお母さんに見つかる場所に自動で移動したり、もう何が起きても不思議ではありません。えぇ、ですから、掟をやぶったら、何がオキテもしょうがありません。えぇ、仕様がありませんから。

おやじデイ

T.Hirase — Sat, 29 Aug 2009 22:13:00 GMT

おやじデイ、というわけで、Gears of War 2 バリにイカすオヤジさんたちがスピーカーのわんくま横浜勉強会に初乗り込みしてきました（←乗り込んだのは私ですよ。）。スピーカーの皆さん、スタッフの皆さんお疲れさまでした。カンパは気持ち程度しかできませんでしたけど、また横浜で勉強会＆お食事会があれば参加します・・・・って、年１回になりそう。。

さっそく、うちに帰って、COMCOMとかOLEOLEしようとか思ったのですが、、、、、うち、メインマシンMacだった・・・・。や、Vistaマシンもあるにはあるんですが、最近起動してない・・・・。や、会社じゃちゃんとXP使って仕事してるんですけど、、うちMac。昨日発売されたばかりのSnowLeopard堪能中でした・・・・・。や、VirtualBoxすりゃいいんですけどね。や、XPのディスクどこいったんだっけ。。。

ま、今流行のまあくんを堪能でもしておきますか。Objective-Cで。

や、今日は、それより、Ninety Nine Nightsを完全実績解除するぞ。

thanks!

[VC] デバッグあり実行で、コンソールウィンドウが勝手に閉じる件

T.Hirase — Fri, 21 Aug 2009 19:12:00 GMT

#include <windows.h>
int main()
{
  if ( IsDebuggerPresent() ) system("pause");
  return 0;
}

これで、デバッグ実行でも、デバッグなし実行でも、プログラム終了時に止まってくれる。

[C/C++/CLI/C#] 二値Enum型を作りたい。

T.Hirase — Thu, 30 Jul 2009 23:39:00 GMT

うすたー：[C#] Enumを二値扱いたかったりする。

うすたー：Enum で bool 型を使いたい時って？ (R.Tanaka.Ichiro's Blog)

前回の記事があまりにひどいできだったので、やり直します。

戻り値はtrue/falseでいいですか？

たとえば、データの書き込み関数「TryWritingBytes」を考えます。バッファリングやハードウェアの都合上、書き込みには失敗する可能性があります。

以下に、TryWritingBytes関数の使用例を示します。同関数の実装は省略。

extern bool TryWritingBytes(void * data, size_t size);

int main()
{
	// something to do.
	
	if (TryWritingBytes(data, size)) {
		// 成功時の処理
	} else {
		// 失敗時の処理
	}
	
	// something to do.
}

このTryWritingBytesの戻り値は、書き込み成功時にtrueを返す仕様になります。が、こういう「成功=Trueだな」みたいに、経験十分プログラマ的な感覚で決められた関数って、危ういと思います。みんながみんな経験十分プログラマとは限らないからです。

成功不成功という値を作るべし。

そんなわけで、trueがどっちを表すかなんて、不毛な議論やコメントを書くくらいなら、さっさと別の二値型を作って、読む人も書く人も間違えないようにすべきです。

そんなこんなで、このコード：

enum IsSuccess : bool {
	Success,
	Failed
};

extern IsSuccess TryWritingBytes(void * data, size_t size);

int main()
{
	// something to do.
	
	if (TryWritingBytes(data, size) == Success) {
		// 成功時の処理
	} else {	// Failed
		// 失敗時の処理
	}
	
	// something to do.
}

こんな言語仕様あった方がいいと思いません？

補足

先の記事で、私、「true/falseを別名で扱いたいだけです」と書いていますが、考え直したところ、全然ダメです。true/falseを別名で扱うこともやはり手段であって、目的ではないですし、仮にできたとしても、所詮別名、成功不成功を表す別名にtrue/falseを代入できてしまいますから。

なので、やっぱり、二値Enum。

また先の記事のコメントやトラックバックで、いくつか代替方法をいただきましたが、やはり私を含む経験不足プログラマが見てもわかりやすい方式（二値Enumに限らず）が言語仕様に組み込まれる方が良いと思います。

ちなみに、別に今回の二値Enumがベストな解だとは思っていません。

あとあと、最後に。他にも言語仕様に不満ってありません？　っていうか何でテキストエディタなんだよ？　とか。クラスのメソッドを編集するのに、どのファイルを開けばいいかなんて、覚えるのなんて意味ないよ？　とか。もっと自由にやらせろ、的な。

追記＠2009-07-30T23:40

Success/False以外の例をいくつか。

enum OpenClose : bool { Open, Close };
OpenClose CheckOpenClose();
OpenClose OpenTray();

enum YesNo : bool { Yes, No };
YesNo CheckResponse();
YesNo WaitResponse();

enum GoodBad : bool { Good, Bad };
GoodBad CheckFile(const char * file_name);

enum MaleFemale : bool { Male, Female };
MaleFemale CheckSexuality();

などなど

[C#] Enumを二値扱いたかったりする。

T.Hirase — Fri, 17 Jul 2009 11:34:00 GMT

以下、妄想コード。

enum ProcResult : bool  /* 2値であることを宣言 */
{
    Success,
    Failure,
}

class Program
{
    static void Main(string[] args)
    {
        ProcResult v = ProcResult.Success;
        v = !v; /* 2値なんで、「!」演算したいよぅ。 */

        System.Console.WriteLine(v);
    }
}

学生時代に学んでおきたいこと

T.Hirase — Wed, 08 Jul 2009 01:45:00 GMT

WEB+DB PRESS vol.51で約４０ページにわたって、「巧いメソッド設計」という特集がくんであります。

中身を見なくても、職業プログラマや、自分で開発してメンテナンスしたことがある人なら納得の内容ばかりです。

第1章：品質を実現する“巧い”プログラミング
プログラミングこそ，真の設計

第2章：クラス設計／パブリックメソッド設計の考え方
処理分担をどう実現するか

第3章：プログラムの意図を読み手にいかに伝えるか
コメント，ネーミング，コーディング標準

第4章：メソッド構造化の実践知識
実装者が心得ておくべき品質向上のための手法

第5章：メソッドの副作用を考える
同じ引数で呼び出しても結果が違うメソッドの扱い方

第6章：メソッドのエラー処理を考える
「契約による設計」を使った潜在的バグへの対処

第7章：Javadocをコード品質の向上に活用しよう
ドキュメントツールJavadoc再入門

大学のとき、こういう授業ってなかったなぁと思います。またデバッグの授業とかもあればいいなと思います。

どうやってバグを回避するか、バグが出たときにどう対処するか、それを学生さんに学んでほしいです。というか、自分が学びたかったです。

学問としての面白みは皆無なような気も、のめり込めば、楽し過ぎる気もしますが、職業プログラマとしては必要なことなので、きっちり学んでから就職したいものです。