Customize Moving Your CUDA* Code to SYCL* with User-Defined Migration Rules

Migration to SYCL Done Your Way

The Intel® DPC++ Compatibility Tool, designed to facilitate the transition from CUDA* to SYCL*, offers a comprehensive set of built-in rules for automatic code migration. However, developers often encounter scenarios where the default migration rules may not align perfectly with their specific codebase or optimization requirements. To address this, the Compatibility Tool provides the capability to define custom migration rules, empowering developers to tailor the migration process to their unique needs.

Whether you have a common function API you prefer over the oneAPI library specification or rely on a target hardware-specific solution not covered by the Unified Acceleration Foundation’s (UXL) oneAPI approach, user-defined migration rules let you customize how CUDA APIs are translated.    

This article discusses user-defined migration rules with the Intel DPC++ Compatibility Tool and how they can create a controlled, customizable, and efficient code porting experience. Everything described here also applies to its open-source counterpart, SYCLomatic.

Understanding User-Defined Migration Rules

User-defined migration rules in the Compatibility Tool allow developers to specify how certain CUDA constructs should be translated into SYCL. These rules are essential when dealing with proprietary or non-standard CUDA extensions, project-specific coding patterns, or optimizing the migrated code for performance on a target platform.

The User-defined migration rules exist in a set of YAML files, which programmers specify on the dpct/c2s command line. When you invoke the compatibility tool, simply provide the name and relative path for the migration rules file using the --rule-file=<filename> option.

The general command syntax is:

dpct [options] [<source0>... <sourceN>]

Thus, a typical command line with user-defined migration rules looks like this:

dpct sample.cu --rule-file=rule_file1.YAML --rule-file=rule_file2.YAML

The –rule-file command line option with your migration command invokes your desired migration rules in the specified YAML file. The –rule-file option can be used multiple times with a single command to specify multiple migration rule files. 

Creating Custom Migration Rules

To create custom migration rules, developers must understand the structure and syntax of the rule definition language. This involves specifying the CUDA source patterns to match and the corresponding SYCL code replacements. A rule is defined by using a <key>: <value> pair. Refer to the Intel DPC++ Compatibility Tool Developer Guide for a complete list of valid keys and values.

Key Components of Custom Rules

1. Pattern Matching: Define the CUDA code patterns that need to be identified for migration. This can include function calls, variable declarations, or any code snippet that requires a custom translation. This is typically defined by the "Kind" and "In" fields of the rule file.
2. Replacement Specification: Articulate the SYCL code that should replace the matched CUDA patterns. Depending on context, this can involve simple one-to-one substitutions or more complex transformations. This is typically defined by the "Out" field and also "Includes", "Subrules" "Prefix" and "Postfix" fields depending on the complexity of the code. 
3. Context Awareness: Incorporate the surrounding code context to ensure the custom rules are applied accurately and do not disrupt the program's overall logic.
4. Rule Priority: Assign priorities to custom rules to manage situations where multiple rules could apply to the same code segment. This ensures a deterministic and expected migration outcome. This is defined by the "Priority" field of the rule file. 

Examples of Migrating Common Elements

User-defined rules are most useful in providing workflow-specific migration. Some common elements in customized migration are API calls, macro, header file, type, class, enum, disable API migration, and pattern rewriter. Following are examples of how their rule files can be written:

1. API Call

Here is a CUDA code with cudaMalloc API

int main(){
  int *ptr;
  cudaMalloc(&ptr, 50);
}

The YAML rule file to migrate this code and the corresponding migrated code using the command

dpct main.cu -rule-file=rule_example.yaml

is as shown below:

Now let us walk through the arguments in the "Out" field of the rule file. 

1. Definition of variables
   The API cudaMalloc(void **(&ptr), size_t 50) is abstracted as  cudaMalloc($1, $2), and 
   $1 is defined as 1st argument: &ptr
   $2 is defined as 2nd argument: 50
2. Getting reference to the variables and types of arguments in API
   ptr is presented as: $deref($1)
   type of ptr int * is presented as: $deref_type($1)
   size_t is presented as: $type_name_of($2)
   Address of (&ptr) is presented as: $addr_of($1)
   $deref()/ $deref_type()/ $type_name_of()/$addr_of() are util functions provided by user-defined migration rules.
3. Predefined variables for SYCL context
   $queue is generally defined as dpct::get_out_of_order_queue() 
   $context is generally defined as dpct::get_default_context()
   $device is generally defined as dpct::get_current_device()

2. Macro

The YAML rule file to migrate a macro can be written as shown below:

- Rule: rule_FOO
  Kind: Macro
  Priority: Takeover
  In: FOO
  Out: GOO
  Includes: []

The migration result of using this rule looks as follows:

dpct main.cu -rule-file=rule_example.yaml

3. Header File

To custom migrate a header file, the rule file can be written as follows:

- Rule: rule_cmath
  Kind: Header
  Priority: Takeover
  In: cmath
  Out: cmath2
  Prefix: "#ifdef MACRO_A\n"
  Includes: ["<cmath3>"]
  Postfix: "#endif"

The migration result using this rule file is as shown below:

4. Type

To customize the migration of a data type, the rule file can be written as follows:

- Rule: type_rule
  Kind: Type
  Priority: Takeover
  In: OldType
  Out: NewType
  Includes: []

The migration result using the above rule file is as shown below:

5. Class

Here is an example of a rule file that can be used to custom-migrate a class:

- Rule: rule_classA
  Kind: Class
  Priority: Takeover
  In: ClassA
  Out: ClassB
  Includes: []
  Fields:
    - In: fieldA
      OutGetter: get_a
      OutSetter: set_a
    - In: fieldC
      Out: fieldD

The migration result of using this file is:

6. Enum

The YAML rule file to migrate an enum can be written as

- Rule: rule_Fruit
  Kind: Enum
  Priority: Takeover
  EnumName: Fruit
  In: apple
  Out: pineapple
  Includes: []

The corresponding migration result using this rule file is

7. Pattern Rewriter

Pattern Rewriter rules are applied to the migrated SYCL code instead of the original CUDA code. Below is a rule file to re-write a pattern of migrated code:

- Rule: rule_post
  Kind: PatternRewriter
  Priority: Takeover
  In: my_max(${args});
  Out: my_min(${args});
  Includes: []
  Subrules:
    args:
      In: a
      Out: b

Now, let’s see how this user-defined rule is applied to a snippet of migrated code.

Benefits of Custom Migration Rules

1. Enhanced Control: Developers gain granular control over migrating specific code constructs, leading to a more predictable and desirable outcome.
2. Optimization Opportunities: Custom rules can be crafted to optimize the migrated code for specific hardware architectures or performance goals.
3. Code Consistency: Maintain coding standards and practices consistent with the rest of the project or organization.

SYCL Migration: Easy, Complete, User-Defined

User-defined migration rules in the Intel DPC++ Compatibility Tool are powerful features that enable developers to customize the code migration process to fit their specific requirements. By leveraging these custom rules, developers can ensure that the migrated SYCL code adheres to project-specific conventions, achieves desired performance characteristics, and integrates seamlessly with existing codebases. As heterogeneous computing environments become more prevalent, tools like the Intel DPC++ Compatibility Tool and features like user-defined migration rules will be instrumental in streamlining the development process and unlocking the full potential of diverse computing platforms.

Start Your SYCL Migration Now

Get started and download the Intel DPC++ Compatibility Tool. It is available standalone or as part of the Intel® oneAPI  Base Toolkit. In addition to these download locations, it is also available through partner repositories and as SYCLomatic in source code.

Additional Resources

• Migrate from CUDA* to C++ with SYCL*
• CUDA* to SYCL* Catalog of Ready-to-Use Applications
• Intel DPC++ Compatibility Tool
• Intel DPC++ Compatibility Tool Developer Guide and Reference
• SYCLomatic