2010-01-01から1年間の記事一覧

gccのインラインアセンブラでレジスタに別名をつける

x86

register __m128* tmp asm ("xmm0"); asm volatile ( "pxor %[tmp], %[tmp]" : [tmp] "+x" (tmp) ); なんだかまわりくどいですね…。このコードは、 pxor %xmm0, %xmm0 に展開されます。

MOVDQUの特性について

x86

IntelのOptimization Manualの2.2.5.1 Efficient Handling of Alignment Hazardsによると、Nehalem系CPUからMOVDQUの性能が向上したとあるので、実験してみました。4パターンのmovを実行して速度差を見ます。alignedはアクセス先のアドレスが16バイトアライ…

Boost.Protoぺろぺろ

最初に このエントリは C++ Advent Calendar jp 2010 : ATND 16日目の記事です. Boost.Protoぺろぺろ Boost.ProtoとはBoostに含まれているExpression Template(ET)のためのライブラリです.Xpressiveや,Spirit,Phoenix v3のベースとなっています.このエ…

CURANDライブラリを使う [Device API]

CUDA Toolkit 3.2から,CURANDという疑似乱数ジェネレータライブラリが導入されました.GPUによる乱数の生成を行えるライブラリです.CURANDは,2つのAPIカテゴリから構成されています. Host API ホストからの使用を目的としたAPI群.Device APIの上に構築…

CURANDライブラリを使う [Host API]

CUDA Toolkit 3.2から,CURANDという疑似乱数ジェネレータライブラリが導入されました.GPUによる乱数の生成を行えるライブラリです.CURANDは,2つのAPIカテゴリから構成されています. Host API ホストからの使用を目的としたAPI群.Device APIの上に構築…

fstabをUUID指定にする方法

Blogは一ヶ月overぶりです。生きてます。最近のUbuntuでは、/etc/fstabのパーティションの指定にUUIDを使うようになっています。そのため、従来の/dev/sda1といった指定でたびたび発生していた問題が解消されています。例えば、HDDが増えた時にsdaがsdbにな…

[CUDA] GPUデバイスからprintfする

CUDA Toolkit 3.1からデバイスからprintfを発行できるようになりました.ただし,Device Capability 2.0以降のデバイスでなければサポートされていないので,GTX480といったいわゆるFermiアーキテクチャのカードが必要です. #include <stdio.h> __global__ void test</stdio.h>…

Oven meets Boost.Spirit.Karma

Boost.Spiritは元々が*1がパーサージェネレータであったこともあり、Qiが一番有名であると思いますが、Boost.Spirit一家にはKarmaという次男坊(?)がいます。Spirit.KarmaはSpirit.Qiの逆で、データから文字列を生成するコンビネータ郡です。Performance of N…

Boost.Buildで複数のバージョンのgccを切り替えて使う方法

$ bjam toolset=gcc のようにして実行をすると、システムデフォルトのgcc, g++が使われます。しかしながら、複数のgccのバージョンを切り替えてビルドやテストを行いたいケースもあります。autoconfによるビルドを採用しているソフトウエアならば、一般的に…

tupleのサイズ

ちょっと気になる所があったので, std::tuple (C++0x) boost::tuple boost::fusion::vector について大きさを調べてみました.fusion::vectorはタプルというよりも,タプル+αという感じですが….コンパイラによっても差が出そうだなーと思いつつも,VCがす…

AES-NIについて(頓挫)

最近のIntel CPUにはAES-NIと呼ばれるAES暗号化を加速させるための専用命令群が実装されています。対応CPUは、32nmプロセスで製造されたi7, i5となっています。いろいろややこしいので、詳しくはIntelのサイトかwikipediaなどを参照してください。最終的には…

BrainfuckをJITしてみた

C++

BFだとJITしやすいので作ってみた。簡単な最適化をしています。add/sub [rdi], immの直後にcmp [rdi], 0が来るのは無駄でしかないので、そこの最適化ぐらいはしてみたいですね。http://gist.github.com/518146ソースコードは400行ほどあるのでgist上に置いて…

Parsecで簡単な電卓を作ってみた

LanguageDef と TokenParser と ExpressionParser を使ってみたかったので。。。空白の扱いについて、ごちゃごちゃ書く必要がなくて楽でいいですね。 module Main (main) where import Text.ParserCombinators.Parsec as Parsec import qualified Text.Parse…

CairoとSDLを組合せてつかう

cairoの描画フロントエンドとしてSDLを使ってみた。ちょっと画面に表示させるだけなのにCocoaだのなんだのを使うのが面倒だったので…。見た感じ、cairo image surfaceとSDL surfaceを比較すると、cairoの方がピクセルフォーマットに対する許容度が低かったの…

GF100のキャッシュの性能

Motivation GTX480に代表されるGF100およびGTX460などのGF104にはL1, L2キャッシュが搭載され、メモリアクセスに関する制限が緩和されました。何も考えずに書くとGT200系では性能が出なかった場合でも、十分な性能が発揮されることがあります。そこで、定数…

Boost.Rangeでdropを実装してみた

なぜかBoost.Range.Adaptorsにdropがなかったので作ってみた。slicedでdropを書けなくはないんですが、長さがわからないrangeに適用できないのが問題です。実際はtakeを先に作ったんですが、dropの方が短かったので…。 boost::copy(input | taco::drop(5), o…

std::chronoによる処理時間測定

C++

処理時間の測定は、clockやgettimeofdayを使う事が多いと思いますが、あまりC++っぽくありません。そこで、C++0xからstd名前空間に入ったstd::chronoを使ってみようと思います。std::chrono::system_clock::now()で現在の時刻(time_point)が取れるので、それ…

Boost.GILでブレセンハムのアルゴリズム

Boost.GILで線を引きたくなったので作ってみた。ある点 a から点 b まで線を引くパターンで、b.x >= a.x かつ b.y >= a.y かつ b.x - a.x >= b.y - a.yのパターンのみ実装しています。それに合わないケースはrotated_viewやtransposed_viewで視点を変更して…

Boost.GILによるpng生成

libpngってC++ friendlyじゃないよね、って呟いていたら、id:faith_and_braveさんに「そこでBoost.GILですよ」と言われたので、作ってみた。 #include <utility> #include <boost/gil/gil_all.hpp> #include <boost/gil/extension/io/png_io.hpp> int main() { namespace gil = boost::gil; gil::point2<ptrdiff_t> dim(400, 400); gil::gray</ptrdiff_t></boost/gil/extension/io/png_io.hpp></boost/gil/gil_all.hpp></utility>…

std::asyncによる並列for_each

C++

C++ではOpenMPよりも、std algorithmが並列に動いた方が便利だよね。ということで作りました。元ネタはhttps://twitter.com/cpp_akira/status/16195866927です。 #ifdef _OPENMP inline size_t concurrency_omp() { size_t n; #pragma omp parallel { n = om…

CUDAでFunctorを使う 捕捉

CUDAでFunctorを使う - fjnlの生存記録のような何かに対する捕捉です。当初、CUDAはデバイスコードへのパラメータをshared memory経由で渡していました。しかし、Compute Capability 2.0*1からはconstant memory経由で渡されるように変更されています。share…

CUDAでFunctorを使う

CUDAの.cuファイルはC++として解釈されるので、templateといった記法やfunctorといったデザインをデバイスコード上でも使うことができます。nvcc 3.0で動作を確認しました。もう2.xの環境が手元になかったので、2.xで使えるかはわかりません。 struct add { …