C#通過OpenCL調用顯卡GPU做高效並行運算

GPU的並行運算能力遠超CPU,有時候我們會需要用到超大數據並行運算,可以考慮用GPU實現,這是一篇C#調用GPU進行運算的入門教程.

1: 下載相關的庫:

https://sourceforge.net/projects/openclnet/
看起來已經N久沒更新了, 不過沒關系,這只是API聲明和參數,opencl本身是有在更新的.

里面有源碼也有DLL,可以引用DLL,也可以直接把源碼添加到工程使用.(建議直接添加代碼...)
*** 需要注意的是 ***:自己建立的工程有個默認的Program類,要改成別的名字,不然會和這里面一個同名的類沖突....

2:建立工程

打開VS建立一個C#控制台工程,Program類改名為MainProgram,添加OpenCL.Net源碼引用

項目屬性里改為[允許不安全代碼]:

3:在MainProgram里聲明引用:

using OpenCLNet;
using CL = OpenCLNet;

4:在項目里添加一個Extend類,內容如下

public static class Extend
{
	/// <summary>
	/// 取指針
	/// </summary>
	/// <param name="obj"></param>
	/// <returns></returns>
	public static unsafe IntPtr ToIntPtr(this int[] obj)
	{
		IntPtr PtrA = IntPtr.Zero;
		fixed (int* Ap = obj) return new IntPtr(Ap);
	}
}//End Class

5:在MainProgram把一段運行在GPU的代碼放在C#的字符串里:

#region OpenCL代碼
private static string CLCode = @"
__kernel void vector_add_gpu(__global int* src_a, __global int* src_b, __global int* res)
{
	const int idx = get_global_id(0);
	res[idx] =src_a[idx] + src_b[idx];
}

__kernel void vector_inc_gpu(__global int* src_a, __global int* res)
{
	const int idx = get_global_id(0);
	res[idx] =src_a[idx] + 1;
}
";
#endregion

6:選中一個設備

在大多數電腦上有1個CPU和2個GPU(集顯,獨顯),有的電腦會有更多或者更少,這里需要選中一個

//獲取平台數量
OpenCL.GetPlatformIDs(32, new IntPtr[32], out uint num_platforms);
var devs = new List<Device>();
//枚舉所有平台下面的設備(CPU和GPU)
for (int i = 0; i < num_platforms; i++)
{
	//這里后面有個參數,是Enum,這里選擇GPU,表示只枚舉GPU,在沒有GPU的電腦上可以選CPU,也可以傳ALL,會把所有設備枚舉出來供選擇
	devs.AddRange(OpenCL.GetPlatform(i).QueryDevices(DeviceType.GPU));
}

//選中運算設備,這里選第一個其它的釋放掉
var oclDevice = devs[0];

7:配置上下文

上下文用來描述CPU與運算設備之間的主從關系.

//根據配置建立上下文
var oclContext = oclDevice.Platform.CreateContext(
	new[] { (IntPtr)ContextProperties.PLATFORM, oclDevice.Platform.PlatformID, IntPtr.Zero, IntPtr.Zero },
	new[] { oclDevice },
	(errInfo, privateInfo, cb, userData) => { },
	IntPtr.Zero
);

8:創建命令隊列

opencl的命令要放到隊列里,然后一次性調用執行方法等待執行完畢,它可以亂序執行,也可以順序執行.如果你想等某命令執行完再繼續,可以在中間加上柵欄(下面會講)

//創建命令隊列
var oclCQ = oclContext.CreateCommandQueue(oclDevice, CommandQueueProperties.PROFILING_ENABLE);

9:編譯OpenCL代碼,並引出兩個Kernel

//定義一個字典用來存放所有核
var Kernels = new Dictionary<string, Kernel>();
#region 編譯代碼並導出核
var oclProgram = oclContext.CreateProgramWithSource(CLCode);
try
{
	oclProgram.Build();
}
catch (OpenCLBuildException EEE)
{
	Console.WriteLine(EEE.BuildLogs[0]);
	Console.ReadKey(true);
	throw EEE;
	//return null;
}
foreach (var item in new[] { "vector_add_gpu", "vector_inc_gpu" })
{
	Kernels.Add(item, oclProgram.CreateKernel(item));
}
oclProgram.Dispose();
#endregion

10:調用Kernel示例:

#region 調用vector_add_gpu核
{
	var A = new int[] { 1, 2, 3, 1722 };
	var B = new int[] { 456, 2, 1, 56 };
	var C = new int[4];
	//在顯存創建緩沖區並把HOST的數據拷貝過去
	var n1 = oclContext.CreateBuffer(MemFlags.READ_WRITE | MemFlags.COPY_HOST_PTR, A.Length * sizeof(int), A.ToIntPtr());
	var n2 = oclContext.CreateBuffer(MemFlags.READ_WRITE | MemFlags.COPY_HOST_PTR, B.Length * sizeof(int), B.ToIntPtr());
	//還有一個緩沖區用來接收回參
	var n3 = oclContext.CreateBuffer(MemFlags.READ_WRITE, B.Length * sizeof(int), IntPtr.Zero);
	//把參數填進Kernel里
	Kernels["vector_add_gpu"].SetArg(0, n1);
	Kernels["vector_add_gpu"].SetArg(1, n2);
	Kernels["vector_add_gpu"].SetArg(2, n3);
	//把調用請求添加到隊列里,參數分別是:Kernel,數據的維度,每個維度的全局工作項ID偏移,每個維度工作項數量(我們這里有4個數據,所以設為4),每個維度的工作組長度(這里設為每4個一組)
	oclCQ.EnqueueNDRangeKernel(Kernels["vector_add_gpu"], 1, new[] { 0 }, new[] { 4 }, new[] { 4 });
	//設置柵欄強制要求上面的命令執行完才繼續下面的命令.
	oclCQ.EnqueueBarrier();
	//添加一個讀取數據命令到隊列里,用來讀取運算結果
	oclCQ.EnqueueReadBuffer(n3, true, 0, C.Length * sizeof(int), C.ToIntPtr());
	//開始執行
	oclCQ.Finish();

	n1.Dispose();
	n2.Dispose();
	n3.Dispose();
	C = C;//在這里打斷點,查看返回值
}
// */
#endregion

11:釋放資源

//按順序釋放之前構造的對象
oclCQ.Dispose();
oclContext.Dispose();
oclDevice.Dispose();

MainProgram所有代碼:

class MainProgram
	{

		#region OpenCL代碼
		private static string CLCode = @"
__kernel void vector_add_gpu(__global int* src_a, __global int* src_b, __global int* res)
{
	const int idx = get_global_id(0);
	res[idx] =src_a[idx] + src_b[idx];
}

__kernel void vector_inc_gpu(__global int* src_a, __global int* res)
{
	const int idx = get_global_id(0);
	res[idx] =src_a[idx] + 1;
}
";
		#endregion


		static void Main(string[] args)
		{
			//獲取平台數量
			OpenCL.GetPlatformIDs(32, new IntPtr[32], out uint num_platforms);
			var devs = new List<Device>();
			//枚舉所有平台下面的設備(CPU和GPU)
			for (int i = 0; i < num_platforms; i++)
			{
				//這里后面有個參數,是Enum,這里選擇GPU,表示只枚舉GPU,在沒有GPU的電腦上可以選CPU,也可以傳ALL,會把所有設備枚舉出來供選擇
				devs.AddRange(OpenCL.GetPlatform(i).QueryDevices(DeviceType.GPU));
			}

			//選中運算設備,這里選第一個其它的釋放掉
			var oclDevice = devs[0];
			for (int i = 1; i < devs.Count; i++) devs[i].Dispose();

			//根據配置建立上下文
			var oclContext = oclDevice.Platform.CreateContext(
				new[] { (IntPtr)ContextProperties.PLATFORM, oclDevice.Platform.PlatformID, IntPtr.Zero, IntPtr.Zero },
				new[] { oclDevice },
				(errInfo, privateInfo, cb, userData) => { },
				IntPtr.Zero
			);

			//創建命令隊列
			var oclCQ = oclContext.CreateCommandQueue(oclDevice, CommandQueueProperties.PROFILING_ENABLE);

			//定義一個字典用來存放所有核
			var Kernels = new Dictionary<string, Kernel>();
			#region 編譯代碼並導出核
			var oclProgram = oclContext.CreateProgramWithSource(CLCode);
			try
			{
				oclProgram.Build();
			}
			catch (OpenCLBuildException EEE)
			{
				Console.WriteLine(EEE.BuildLogs[0]);
				Console.ReadKey(true);
				throw EEE;
				//return null;
			}
			foreach (var item in new[] { "vector_add_gpu", "vector_inc_gpu" })
			{
				Kernels.Add(item, oclProgram.CreateKernel(item));
			}
			oclProgram.Dispose();
			#endregion

			#region 調用vector_add_gpu核
			{
				var A = new int[] { 1, 2, 3, 1722 };
				var B = new int[] { 456, 2, 1, 56 };
				var C = new int[4];
				//在顯存創建緩沖區並把HOST的數據拷貝過去
				var n1 = oclContext.CreateBuffer(MemFlags.READ_WRITE | MemFlags.COPY_HOST_PTR, A.Length * sizeof(int), A.ToIntPtr());
				var n2 = oclContext.CreateBuffer(MemFlags.READ_WRITE | MemFlags.COPY_HOST_PTR, B.Length * sizeof(int), B.ToIntPtr());
				//還有一個緩沖區用來接收回參
				var n3 = oclContext.CreateBuffer(MemFlags.READ_WRITE, B.Length * sizeof(int), IntPtr.Zero);
				//把參數填進Kernel里
				Kernels["vector_add_gpu"].SetArg(0, n1);
				Kernels["vector_add_gpu"].SetArg(1, n2);
				Kernels["vector_add_gpu"].SetArg(2, n3);
				//把調用請求添加到隊列里,參數分別是:Kernel,數據的維度,每個維度的全局工作項ID偏移,每個維度工作項數量(我們這里有4個數據,所以設為4),每個維度的工作組長度(這里設為每4個一組)
				oclCQ.EnqueueNDRangeKernel(Kernels["vector_add_gpu"], 1, new[] { 0 }, new[] { 4 }, new[] { 4 });
				//設置柵欄強制要求上面的命令執行完才繼續下面的命令.
				oclCQ.EnqueueBarrier();
				//添加一個讀取數據命令到隊列里,用來讀取運算結果
				oclCQ.EnqueueReadBuffer(n3, true, 0, C.Length * sizeof(int), C.ToIntPtr());
				//開始執行
				oclCQ.Finish();

				n1.Dispose();
				n2.Dispose();
				n3.Dispose();
				C = C;//在這里打斷點,查看返回值
			}
			// */
			#endregion

			//按順序釋放之前構造的對象
			oclCQ.Dispose();
			oclContext.Dispose();
			oclDevice.Dispose();
		}
	}//End Class

運行效果:

至此,操作完成~
我在文中留了一個Kernel,你可以嘗試調用看看.
相關代碼git:
https://gitee.com/ASMTeam/CSharpOpenCLDemo