DPCT1008
目次
DPCT1008#
メッセージ#
<function name> 関数は SYCL* では定義されていません。これは、ハードウェア固有の機能です。代替関数については、ハードウェア・ベンダーにお問い合わせください。
詳細な説明#
clock 関数呼び出しは、SYCL* では定義されていないため、置き換えられませんでした。このハードウェア固有の機能を、ハードウェア・ベンダーが提供するものに置き換えてください。
修正方法の提案#
代替関数については、ハードウェア・ベンダーにお問い合わせください。
例えば、以下のオリジナル CUDA* コードについて考えてみます。
1 __global__ void k(clock_t *timer) {
2 int tid = threadIdx.x;
3 if (tid == 0)
4 timer[0] = clock();
5
6 // workload
7
8 __syncthreads();
9 if (tid == 0)
10 timer[1] = clock();
11 }
12
13 int main() {
14 clock_t *dtimer = NULL;
15 cudaMalloc((void **)&dtimer, sizeof(clock_t) * 2);
16 clock_t timer[2];
17 k<<<1, 128>>>(dtimer);
18 cudaMemcpy(timer, dtimer, sizeof(clock_t) * 2, cudaMemcpyDeviceToHost);
19 cudaFree(dtimer);
20 long double time = timer[1] - timer[0];
21 return 0;
22 }このコードは、以下の SYCL* コードに移行されます。
1 #include <sycl/sycl.hpp>
2 #include <dpct/dpct.hpp>
3 #include <time.h>
4 void k(clock_t *timer, sycl::nd_item<3> item_ct1) {
5 int tid = item_ct1.get_local_id(2);
6 if (tid == 0)
7 /*
8 DPCT1008:1: clock function is not defined in SYCL.This is a
9 hardware-specific feature.Consult with your hardware vendor to find a
10 replacement.
11 */
12 timer[0] = clock();// clock() is used to measure the kernel runtime
13
14 // workload
15
16 item_ct1.barrier();
17 if (tid == 0)
18 /*
19 DPCT1008:2: clock function is not defined in SYCL.This is a
20 hardware-specific feature.Consult with your hardware vendor to find a
21 replacement.
22 */
23 timer[1] = clock();
24 }
25
26 int main() {
27 dpct::device_ext &dev_ct1 = dpct::get_current_device();
28 sycl::queue &q_ct1 = dev_ct1.default_queue();
29 clock_t *dtimer = NULL;
30 dtimer = sycl::malloc_device<clock_t>(2, q_ct1);
31 clock_t timer[2];
32 q_ct1.parallel_for(
33 sycl::nd_range<3>(sycl::range<3>(1, 1, 128), sycl::range<3>(1, 1, 128)),
34 [=](sycl::nd_item<3> item_ct1) {
35 k(dtimer, item_ct1);
36 });
37 q_ct1.memcpy(timer, dtimer, sizeof(clock_t) * 2).wait();
38 sycl::free(dtimer, q_ct1);
39 long double time = timer[1] - timer[0];
40 return 0;
41 }このコードは次のように書き換えられます。
1 #define DPCT_PROFILING_ENABLED
2 #include <sycl/sycl.hpp>
3 #include <dpct/dpct.hpp>
4 #include <time.h>
5 void k(sycl::nd_item<3> item_ct1) {
6 // workload
7 }
8
9 int main() {
10 dpct::device_ext &dev_ct1 = dpct::get_current_device();
11 sycl::queue &q_ct1 = dev_ct1.default_queue();
12 dpct::event_ptr start;
13 dpct::event_ptr end;
14 start = new sycl::event();
15 end = new sycl::event();
16 *start = q_ct1.ext_oneapi_submit_barrier();
17 q_ct1.parallel_for(
18 sycl::nd_range<3>(sycl::range<3>(1, 1, 128), sycl::range<3>(1, 1, 128)),
19 [=](sycl::nd_item<3> item_ct1) {
20 k(item_ct1);
21 });
22 *end = q_ct1.ext_oneapi_submit_barrier();
23 end->wait_and_throw();
24 long double time =
25 (end->get_profiling_info<sycl::info::event_profiling::command_end>() -
26 start
27 ->get_profiling_info<sycl::info::event_profiling::command_start>()) /
28 1000000.0f; return 0;
29 }