OpenCVビルド時のCUDAのccbinの指定方法

リンネの生物階級分類において,「スズキ目」というのは脊椎動物中で最大の規模を誇る「目」なんだそうです.
スズキ目スズキ科スズキはまさに全生物中における最大勢力を束ねる王者と言えましょう.だから何.

さかな前線 » CMakeでCUDAの-ccbinの正しい指定方法の姉妹記事です.

さて,Linux上でOpenCVでCUDAサポートをONにしてビルドするとき,GCCのバージョンを指定したいことがあります.
基本的にはデフォルトGCCなんですが,4.8系などは現在のCUDAのnvccではサポートされていませんので,CUDAビルド時に限りGCC4.4や4.6を使うという設定です.

具体的には,CMakeのときに以下のように書きます

$ pwd
/home/hogehogehogehogeho/opencv-2.4.6.1/build

$ cmake -DWITH_CUDA=ON -DCUDA_HOST_COMPILER=/usr/bin/g++-4.4 ..

$ make -j40 -s; sudo make install

WITH_CUDAとともにCUDA_HOST_COMPILERを正しく設定すればいいのですが,CUDA_HOST_COMPILERに指定するgccはPATHが通っていても絶対パスで指定する必要があるようです.
CUDA_HOST_COMPILER=g++-4.4と書くと,ないよ>< って言われます.

以上です.

参考:OpenCV – Bug #2844: build with CUDA 5.0 on macosx – OpenCV DevZone

root権限の無い哀れな子魚ちゃんたちのためのgcc野良インストール(Linux)

僕の専門は画像パターン認識でして,魚画像からの種類判別とかいつかやっていみたいなぁとか思ってるんですがなかなかこれは問題として難しそうです.ここ(PDF注意)とかにそういう文献が出てるんですがこれは何の本なんでしょう?知ってる方いたら教えて欲しいです.使われてる技術を見ると2000年代以降のものではないなという印象.それより昔のことは体系的には知りません・・・w

さて,常に最新のgccが使いたいという人のためのにgccを野良ビルドしてインストールする手順を以前さかな前線 » GCC 4.6.2が出たので手動インストール方法メモ(Linux)に書きました.

一方で,共用計算機などではroot権限がなく勝手にプログラムをインストールできないことがよくあります.
そんなときでも,自分の領域が割り当てられていればそこに野良ビルドしてインストールして使うことは可能です.
そういう構成を可能にするためのgccのビルド方法をメモっときました.
今回はGCC 4.7.2を64bit版Linuxに入れることを考えています.

基本的にはさかな前線 » GCC 4.6.2が出たので手動インストール方法メモ(Linux)と同じです.
ただし一部,gcc4.6.2の手動インストールメモ – N_Nao’s logを参考にしています.

Continue reading

C++11のstd::discrete_distributionでノンパラメトリックな分布の乱数生成

おさかなおさかなって言ってたらリアルにお魚好きになってしまったって話は前もしたと思うんですが,そのせいでホントに魚が食べたい今日このごろです.調理法としては,お刺身の次点くらいで焼き魚と酒蒸しが好きです.お刺身はそのままスーパーで買ってくればいいし(美味しいかと言われると微妙ですが),焼き魚も買ってきてグリルで焼くだけなので気軽にできるんですが,酒蒸しはちょっと自宅ではできなくて食卓が寂しいです.

さて,C++11では乱数関連のライブラリについてもboost由来の大幅強化がされまして,ホントに強力なものになっています.
まず乱数エンジンでは,よりよい擬似乱数アルゴリズムが豊富に取り揃えられており,さらにはハードウェアを利用した真の乱数も扱えます.
また生成される乱数の分布の点でも,通常の一様分布や各種の確率分布(GaussianやPoissonなど)が利用できます.

今日お話するのは分布の中でも,出力される各値の確率をユーザが指定できるノンパラメトリックな分布を実現するstd::discrete_distributionのお話.

なおboost::randomからstd::randomへは大きな設計変更がありましたけれども,今回の話に関しては普通にstd::discreted_distributionをboost::random::discrete_distributionとそのまま読み替えてもらっても大丈夫です.

Continue reading

SSEとAVXで高次元ベクトルの内積計算を高速化してみた

世界最速のお魚と言えばカジキ類で,泳ぐ速度は時速100km/hを超えるとか.55ノット程になるのでこれはMk-48魚雷にも匹敵するほどです.
一方ちょっとチートな高速お魚としては,お馴染みトビウオが飛行中に最大70km/hほどに達するとか.
今日はそんな若干チートな高速化のお話(?)ということで,SSE組み込み命令について.

SSEやAVXといえばお馴染みSIMD命令で,それをプログラムから構造体と関数の形式で高移殖に記述する方法がSIMD組み込み関数(SIMD Intrinsic)なわけですが,これを使ってごく典型的なベクトルの内積計算を高速化してみました.

ベクトルの内積の高速化と言えば星の数ほどもされてる話なわけで,いまさら魚の情報なんか役に立つ気は全くしないのですが,純粋に自分でやらないとわかんない>< ということで,

  • とにかく書いてみよう
  • 効果の程はいかに?

を調べてみたくて,やってみました.
もちろん完全に最適化されたコードでもなく,この記事の内容は無保証です.

参考は主に
SSE.浮動小数点演算手動最適化は本当に効果的なのか – デー
インテル(R) Advanced Vector Extensions (インテル(R) AVX) の組み込み関数
です.

Continue reading

google-glogで最低限使いたい,ただ1つの機能 〜自動スタックトレース出力〜

もう10年以上も前になりますけども,近場の漁港によく連れて行ってもらって魚釣りしていました.
ちょうどアジは西日本の日本海側が本場ということもあって,僕の地元では岸壁に立ってサビキをすると小アジがたくさん釣れました.

さて,google-glogと言えばgoogle謹製のログライブラリ.
Google Japan Blog: C++ のプログラムのデバッグを楽にする方法にて概要が紹介されていて,ドキュメントには簡単なチュートリアルがあります.基本的には使い方に難しいところは何も無いです.

お魚さん的には学部卒論のときから使っててもはや手放せないんですが,作業が切羽詰ってくるとあんまり律儀にログを出す余裕がなくなってきますw
そんな時でもただ一ついつでも使ってたのが,「落ちた時にスタックトレースを吐く機能」です.
ということで今回はそれに絞って紹介.まとまった記事を書くには今時間的余裕がない><

Continue reading

BundlerによるStructure from MotionでKAZE局所特徴量を使ってみた

一応お魚キャラとしてやっているわけですが,私のアイコンの魚を前からみるとどうなるの?とときどき聞かれます.
回答としては,このお魚は”幅を持たない”形状をしています.二次元です。

さて,大量の未整列な多視点画像からの三次元形状復元ついでにカメラの内外部パラメータの推定まで一気にしちゃう手法のことをStructure from Motion(SfM)と言い,結構長いこと研究されているホットな話題であります.

そんなSfMを行うためのツールにBundlerがあります.r○byのbundlerとは全く関係ありません.
Bundlerは局所特徴量を用いたsparseな画像間対応からSfMをするもので,一般的にはSIFTが使われていますが,対応付けは独立プログラムに委譲されてるので,うまくやればどんな局所特徴量でもSfMをお楽しみいただける設計になっています.

ということで,Computer Vision Advent Calendar 2012で書いたさかな前線 » ECCV2012で発表されたKAZE局所特徴量を試してみたの続きという事で,KAZE特徴量を用いたBundlerでのSfMを試してみました.実用的な例がほしかったのでw
CVAdventCalendarの@yasutomo57jpさんの記事の二番煎じです…

Bundler(やPMVS)のアルゴリズムについては他に譲ります(←

20121209_bundler_example_pmvs

※結論から言うとやり方がまずいのでto be continuedです…

Continue reading

cv::Matとcv::Mat_<T>の画素アクセスの速度

近くのスーパーに秋は国内産の鮭が売っててよく買ってたんですが最近は輸入のサーモンばっかしです少し寂しいです.

OpenCVでちょっと気になって,画素単位のアクセスをするときcv::Matのat()による方法とcv::Mat_のoperator()による方法でどっちが速いかなと思って調べてみました.
星の数ほどされてる議論のようなするけどぱっと見た範囲で見当たらなかったのでやってみた.
参考までに画像ピクセル値へのアクセスと計算速度 | OpenCV.jpもあります.

結論から言うと,全く同じでした.

コードはこんなかんじ.時間計測はオレオレライブラリで高精度な計測ができます.

実行結果はこんなかんじ.

% ./a.out
cv::Mat::at<T>() -> 9.82353ms/回
cv::Mat_<unsigned char>::operator() -> 9.69412ms/回

大きさ(w,h)によらず同じくらいの模様.

あと気になる生配列とのアクセス速度の差ですが,上述プログラムと同じ処理を書いちゃうと最適化で全部消えちゃうしそうならないくらい複雑なプログラムは面倒だったので断念.
最適化しなければ,cv::Matの方が3割くらい遅いくらいでした.最適化するとどうなるやら.

  • CPU: i7 870 (2.93GHz)
  • OS: Ubuntu 11.10 x86
  • gcc: 4.7.0
  • option: -std=c++0x
  • OpenCV: 2.4.3

追記:
Ivy Bridge速すぎワロタw

  • CPU: i5 3570 (3.40 GHz)
  • OS: Linux Mint 13
  • gcc: 4.6.3
  • OpenCV: 2.4.3
% ./a.out
cv::Mat::at<T>() -> 2.34902ms/回
cv::Mat_<unsigned char>::operator() -> 2.34902ms/回

ECCV2012で発表されたKAZE局所特徴量を試してみた

先日ちょっと膝をケガしてしまいまして,ギプスに松葉杖です.けっこうリアルさかなに近づけたんじゃないかと思うのですがいかがでしょう?
折しも12/6で私も23歳を迎えました.できれば無事健康体で迎えたかったこの日.

この記事はComputer Vision Advent Calendar 2012の12/7担当分として書かれたものです.
主催者でありながら華麗に遅刻をキメました.これはひどい.
高専時代の後輩でもある@blue_jamくん,お疲れ様でした.そして誕生日メッセージありがとうございます.

さて,コンピュータビジョンにおける極めて重要な要素のひとつに画像特徴量があります.
これまでにたくさんの特徴抽出手法が提案されてきているわけですが,この度は最近の国際会議で発表されたばかりの新しい手法の実装が公開されていたので,それを試して見ることにしました.

Continue reading

OpenCVの二値画像からの画素値の取得でハマった

OpenCVの画像形式cv::Matはいろんな格納方式ができて,二値画像ももちろんいけるんですが,画素値を個別にアクセスするときのやり方でうっかりしたミスでハマってしまいました.

ということで自虐を兼ねてメモ.(逆では

//画像を読み込む
cv::Mat img = cv::imread("fugafuga.png", CV_LOAD_IMAGE_GRAYSCALE);

//大津の方法で二値化.こうすると二値画像のフォーマットでMatに格納してくれます
//srcとdstが一緒ですが一応問題ないです
cv::threshold(img, img, 0, 255, CV_THRESH_BINARY | CV_THRESH_OTSU);

//全画素をPGMフォーマットで出してみる
std::cout << "P1" << std::endl;
std::cout << img.cols << " " << img.rows << std::endl;
for(int y = 0; y < img.rows; y++)
{
    for(int x = 0; x < img.cols; x++)
    {
        //std::cout << img.at<int>(y, x) << " ";    //まちがい!!
        std::cout << static_cast<int>(img.at<unsigned char>(y, x)) << " ";
    }
    std::cout << std::endl;
}

at<int>()はカラー画像のBGRAをごそっと取り出すとき用だったりするので,二値画像の画素値を出すときはat<unsigned char>()を使いましょうというお話でした.static_castはべつにせんでもええです.
ちなみにat<int>()でやると事実上のゴミが出てきます.
何をいまさらこんなところでつまづいてるのか小一時間問い詰めたい.

nvccとgccでのオブジェクトファイルのリンク

ねむいですね。
最近ブログのネタがないので、しょーもない内容で記事ひとつ。すぐ↓の説明さえ読んでわかるならそれ以下は見る必要ないですww

さて、CUDAコンパイラnvccの吐くオブジェクトファイルは普通にgccの吐くのと同じELF形式ですので、デバイスコードを記述したソース(*.cu)をnvccで-cオプション付きでコンパイルして出力したオブジェクトファイルとgccで出したオブジェクトファイルをリンクすることができます。

mainから直接nvccを通さないために、カーネルを呼び出すフロントエンド関数を用意

void a_cuda_kernerl_frontend();

デバイスコードがこちら

#include "a_cuda_wrapper.h"

__global__ void cuda_kernel()
{
}

void a_cuda_kernerl_frontend()
{
	cuda_kernel<<<1, 1>>>();
}

そしてnvccを通さないmainがこちら

#include <iostream>
using namespace std;

#include "a_cuda_wrapper.h"

int main()
{
	a_cuda_kernerl_frontend();
}

これで

% g++ -c main.cpp
% nvcc -c kernel.cu
% nvcc main.o kernel.o
% ./a.out

リンクするときにg++じゃなくてnvccを使いましょう。当然ですね。

これで、C++ベースのでかいシステムの一部としてCUDAを用いることもできますね!!
これができることくらい誰でも知ってると思うんだけど、実際にやってるサンプルて見ないなぁと思って(mainまで*.cuに書いてある)
(そうですサンプルにするほどのこともないからです