syghの新フラグメント置き場

プログラミングTipsやコード断片の保管場所です。お絵描きもときどき載せます。

GPU

Vulkanグラフィックスの簡単なレンダリングサンプル

このあいだ書いた記事「VulkanシェーダーでSub-group命令を使う」では、グラフィックスをすっ飛ばしていきなりコンピュートシェーダーで拡張命令を試すという変態的なことをしましたが、GPGPUに興味のない人には退屈な内容だったかもしれません。以前、Direc…

VulkanシェーダーでSub-group命令を使う

NVIDIAはKepler (Compute Capability 3.0) 世代のハードウェアにおいて、Warp Shuffle命令を実装しました。WarpシャッフルはCUDAから組み込み関数 (intrinsic function) の形で利用できるSIMD命令の一種で、Warpと呼ばれるスレッドグループ内での並列データ…

GeForceドライバー380系列のDirect3D 11バグ

3DグラフィックスとC++の研究目的で、DirectX 11 (Direct3D 11) を使った自前FBXビューアーを開発しているのですが、とある自作FBXファイル(約18,000ポリゴン程度)を開いて、カメラを回転させながら描画すると、レンダリングが停止する現象に遭遇しました…

OpenCL/OpenGL/OpenCVのバイナリキャッシュ機能は使ってはいけない

OpenCL/OpenGLには当初、カーネルおよびシェーダープログラムに関してSPIR/SPIR-Vのような中間表現(バイトコード)規格が用意されておらず、それゆえオフラインコンパイルがサポートされていませんでしたが、コンパイル済みバイナリ(ベンダー依存)のキャ…

Vulkan SDK付属のGLSLコンパイラー

2016年2月にVulkanの正式仕様とSDKがリリースされ、その後NVIDIAやAMDなど大手GPUベンダー各社からもPC向け正式ドライバーがリリースされ始めているので、そろそろ試してみることにしました。まずはシェーダーのコンパイラーまわりから入ります。ちなみにVul…

Direct3D 12 (DirectX 12) の簡単なサンプル

(本記事の内容は執筆時点の暫定仕様に基づくものです。実際の最終製品版とは一部異なる部分があります)GDCの開催された2015年3月にはすでにDirect3D 12 (DirectX 12) の暫定ドキュメントとVisual Studio 2015 CTPが公開されていたのですが、肝心の実行環境…

GPGPU戦争の歴史を紐解く ―勃発から現在に至るまで―

GDC 2015でついにOpenCL 2.1が発表されました。SPIR-V中間表現(中間言語、バイトコード)でプログラミング基盤をVulkan API (OpenGL Next Generation, glNext) と共有することができるそうです。これでカーネルをオフラインコンパイルできるようになります…

NVIDIA/AMDのDirectX 11.x/12対応に関して

NVIDIAの場合 GeForce GTX 970のメモリ3.5GB問題で、嘘スペック(なんちゃって4GB他)を公表していたがために訴訟を起こされてしまったNVIDIAですが、実はDirectX (Direct3D) 対応レベルに関してもわりとあこぎなことをやっているというか実際詐欺です。Ferm…

NVIDIAのOpenCL対応状況

NVIDIAのOpenCLドライバーはIntelやAMDと比べて規格への対応が遅く、2015年2月時点でもOpenCL 1.1どまりです。NVIDIAのOpenCL SDKとなるCUDA Toolkitも、7.0時点でOpenCL 1.1までしか対応していません(cl.hに定義されているのはCL_VERSION_1_1まで)。例え…

CUDA Warpシャッフル命令のエミュレーション

今更ですがせっかくCompute Capability 3.0対応のKepler世代グラフィックスカードを手に入れたので、CUDAのWarpシャッフル命令の動作テストを兼ねて、代替機能をエミュレートする関数を書いてみました。 Visual Studio 2012、CUDA 6.5、GeForce GTX 770で動…

OpenGLコンピュートシェーダー

(これは2013-01-30に書いた故OCNブログの記事を移植したものです)2013-01-05に公開されたNVIDIAのリテールドライバー310.90で、ようやくOpenGL 4.3が正式に使えるようになりました。最新世代のKeplerのほか、旧世代のFermiアーキテクチャのグラフィックス…

GLSLが使いにくい件について

(これは2011-09-20に書いた故OCNブログの記事を加筆修正したものです)個人的にグラフィックス プログラミングが好きで、仕事でも趣味でもDirectXやOpenGLを使っているんですが、OpenGL/OpenGL ESの不甲斐なさにそろそろ愛想が尽きてきました。 具体的にDir…

コンピュートシェーダーの性能

(これは2011-10-08に書いた故OCNブログの記事を加筆修正したものです)DirectX 9世代のシェーダーモデル3.0までは、画面座標系でのポスト エフェクト処理(たとえばガウスぼかしやブルーム、トーン マッピング、SSAOなど)はレンダーターゲットを切り替えて…

コンピュートシェーダーでGPUパーティクル

DirectX (Direct3D) のコンピュートシェーダー (Compute Shader, DirectCompute) には4バイト (uint) 単位で要素アクセスが可能なByteAddressBuffer/RWByteAddressBuffer (BAB) というデータ型があるんですが、DirectXにおいてGPUで頂点バッファの内容を直接…

HLSLの行列乗算がmul()関数な理由

GLSLでは行列の乗算に*演算子を使います。 // GLSL mat4 mTransformA; mat4 mTransformB; vec4 vIn; vec4 vOut = mTransformB * mTransformA * vIn; 一方HLSLではmul()関数を使います。 // HLSL float4x4 mTransformA; float4x4 mTransformB; float4 vIn; fl…

BGRA/RGBAに関してそろそろ一言いっておくか

世には4つのカラーチャンネル(Red, Green, Blue, Alpha)をサポートする画像フォーマットは数多あるのですが、整理も兼ねて Direct3D や OpenGL のテクスチャ フォーマットとの関連に言及しながらまとめてみようと思います。早速ですがいきなりサンプルコー…

CUDAでテクスチャ

前回の記事では CUDA の線形メモリ(Linear Memory, cudaMalloc() で生成)を使って FP16 浮動小数点数値を格納しました。 今回は CUDA の Array と Texture を使って、0x0000 ~ 0xFFFF の範囲の 16bit 整数値(ushort)を正規化して 0.0 ~ 1.0 の範囲の 32b…

CUDAでhalf型

VRAM 使用量や帯域の節約目的で、CUDA や OpenCL のカーネルに FP16 半精度浮動小数点数(half)型のデータを渡す場合の話です。 通例 GPGPU で使われる実数は FP32 単精度浮動小数点数(float)型なんですが、half だとその半分のデータ量で済むため、帯域…

OpenCLデバイスの列挙

1年くらい前に書いたコードですが、OpenCLプログラムの入門、まずはデバイスの列挙です。 VC++ 2010、AMD APP SDK v2、NVIDIA GPU Computing SDK 4.0で検証。 #define __CL_ENABLE_EXCEPTIONS #pragma warning(push) #pragma warning(disable:4290) #include <CL/cl.hpp></cl/cl.hpp>…