4

シミュレーションを pyOpenCL に移動していますが、データ アクセスが機能しません。ベクトルの 1D 配列を提供しようとしています (まあ、実際にはいくつかありますが、ここに示した例では 1 つだけを使用しています)。

現在、いくつかのベクトルが正常にコピーされていますが、データは私が提供したものではありません。

以前にここに投稿したことがないと思うので、フォーマット/プレゼンテーションのいずれかが間違っている場合はお詫び申し上げます. また、すべてのシミュレーション コードを削除したので、このコードは現在実際には何もしていないことに気付きました。バッファの受け渡しを正しくしたいだけです。

前もって感謝します。

カーネル (kertest.py):

step1 = """
#pragma OPENCL EXTENSION cl_amd_printf: enable
#define X xdim
#define Y ydim
__kernel void k1(__global float3 *spins,
                 __local float3 *tile)
{        
    ushort lid = 2 * get_local_id(0);
    ushort group = 2 * get_group_id(0);
    ushort num = get_num_groups(0);
    int lim = X*Y*3;

    for (ushort i = 0; i < lim; i++)
        {
            if (lid == 0 && group == 0)
            {
                printf("%f :: %d\\n", spins[i].x, i);
            }
         }
}"""

コード自体 (gputest.py):

import kertest as k2D
import numpy as np
import pyopencl as cl

class GPU_MC2DSim():
    def __init__(self, x, y):
        self.x = x
        self.y = y

        if x >= y:
            self.xdim = int(self.x)
            self.ydim = int(self.y)
        else:
            self.xdim = int(self.y)
            self.ydim = int(self.x)

        if self.xdim % 2 != 0: self.xdim += 1

        if self.ydim % 2 != 0: self.ydim += 1

        self.M = np.ones((self.xdim*self.ydim, 3)).astype(np.float32)
        self.M[:, 1] += 1.0
        self.M[:, 2] += 2.0

        print self.M

    def simulate(self):
        ctx = cl.create_some_context()
        q = cl.CommandQueue(ctx)
        mf = cl.mem_flags

        #Pass buffer:
        M_buf = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf = self.M)

        #Insert kernel parameters:
        params = {"xdim" : "%d" % (self.xdim),
                  "ydim" : "%d" % (self.ydim),
                  }
        for name in params:
            k2D.step1 = k2D.step1.replace(name, params[name])

        #Compile kernel:
        step1 = cl.Program(ctx, k2D.step1).build()

        locmem = cl.LocalMemory(self.xdim*4*4)

        step1.k1(q, ((self.xdim*self.ydim)/4,), (self.xdim/2,), M_buf, locmem).wait()
        return None

xdim = 4
ydim = 4
sim = GPU_MC2DSim(xdim, ydim)
sim.simulate()
4

1 に答える 1

4

データをデバイスにコピーするためのコードは問題ありません。ただし、カーネルには少なくとも2つの問題があります。

  1. float3OpenCL 1.2仕様、6.1.5に従って、値は16バイトに整列されると予想されます。

    3成分ベクトルデータ型の場合、データ型のサイズは4 * sizeof（成分）です。これは、3コンポーネントのベクトルデータ型が4 * sizeof（コンポーネント）の境界に揃えられることを意味します。vload3およびvstore3の組み込み関数を使用して、パックされたスカラーデータ型の配列から3成分のベクトルデータ型をそれぞれ読み取りおよび書き込みできます。

    float3デバイスにアップロードする値は、カーネルが値を直接読み取るために適切に調整されていません。

  2. 制限の計算int lim = X*Y*3;が少しずれています。の配列から既に読み取ろうとしているfloat3ので、*3は不要です。

両方の問題の解決策は簡単です。仕様に記載されているように、vload3floatを使用してsの配列からロードする必要があります。

#pragma OPENCL EXTENSION cl_amd_printf: enable
#define X xdim
#define Y ydim
__kernel void k1(__global float *spins,
                 __local float3 *tile)
{
    ushort lid = 2 * get_local_id(0);
    ushort group = 2 * get_group_id(0);
    ushort num = get_num_groups(0);
    int lim = X*Y;

    for (ushort i = 0; i < lim; i++)
        {
            if (lid == 0 && group == 0)
            {
                float3 vec = vload3(i, spins);
                printf("(%f, %f, %f) :: %d\\n", vec.x, vec.y, vec.z, i);
            }
         }
}
于 2012-07-31T19:43:55.543 に答える