Merge 'master' into 'amd-master'
Change-Id: I2bb8d2d98bb90185d663d0dc3bad1803acd857cc
Этот коммит содержится в:
Jenkins
@@ -1,61 +0,0 @@
|
||||
#!/bin/bash
|
||||
|
||||
function die {
|
||||
echo "${1-Died}." >&2
|
||||
exit 1
|
||||
}
|
||||
|
||||
if [ $# = 0 ]; then
|
||||
die "$(basename $0): Invalid number of arguments"
|
||||
fi
|
||||
|
||||
: ${ROCM_PATH:=/opt/rocm}
|
||||
: ${ROCM_TARGET:=gfx803}
|
||||
|
||||
INPUT_FILES=""
|
||||
OUTPUT_FILE=""
|
||||
while [[ $# -gt 1 ]]; do
|
||||
key="$1"
|
||||
case $key in
|
||||
-o)
|
||||
OUTPUT_FILE="$2"
|
||||
shift
|
||||
;;
|
||||
*)
|
||||
INPUT_FILES="$INPUT_FILES $key"
|
||||
esac
|
||||
shift
|
||||
done
|
||||
|
||||
[ INPUT_FILES != "" ] || die "No source files specified"
|
||||
[ OUTPUT_FILE != "" ] || die "Output file not specified"
|
||||
|
||||
SOURCE="${BASH_SOURCE[0]}"
|
||||
HIP_PATH="$( command cd -P "$( dirname "$SOURCE" )/.." && pwd )"
|
||||
|
||||
export KMDUMPISA=1
|
||||
export KMDUMPLLVM=1
|
||||
hipgenisa_dir=`mktemp -d --tmpdir=/tmp hip.XXXXXXXX`
|
||||
hipgenisa_main=`mktemp src.XXXXXXXX.cpp`
|
||||
hipgenisa_files="$hipgenisa_main"
|
||||
|
||||
for inputfile in $INPUT_FILES; do
|
||||
sed 's/extern \+"C" \+//g' $inputfile > $inputfile.kernel.tmp.cpp
|
||||
hipgenisa_files="$hipgenisa_files $inputfile.kernel.tmp.cpp"
|
||||
done
|
||||
printf "\nint main(){}\n" >> $hipgenisa_main
|
||||
|
||||
$HIP_PATH/bin/hipcc -DGENERIC_GRID_LAUNCH=0 $hipgenisa_files -o $hipgenisa_dir/a.out
|
||||
mv dump* $hipgenisa_dir
|
||||
|
||||
hsaco_file="dump-$ROCM_TARGET.hsaco"
|
||||
map_sym=""
|
||||
kernels=$(objdump -t $hipgenisa_dir/$hsaco_file | grep grid_launch_parm | sed 's/ \+/ /g; s/\t/ /g' | cut -d" " -f6)
|
||||
for mangled_sym in $kernels; do
|
||||
real_sym=$(c++filt $(c++filt $mangled_sym | cut -d: -f3 | sed 's/_functor//g') | cut -d\( -f1)
|
||||
map_sym="--redefine-sym $mangled_sym=$real_sym $map_sym"
|
||||
done
|
||||
objcopy -F elf64-little $map_sym $hipgenisa_dir/$hsaco_file $OUTPUT_FILE
|
||||
|
||||
rm $hipgenisa_files
|
||||
rm -r $hipgenisa_dir
|
||||
@@ -1,207 +0,0 @@
|
||||
hsa_amd_coherency_get_type
|
||||
hsa_amd_coherency_set_type
|
||||
hsa_amd_profiling_get_dispatch_time
|
||||
hsa_amd_profiling_get_async_copy_time
|
||||
hsa_amd_image_get_info_max_dim
|
||||
hsa_amd_queue_cu_set_mask
|
||||
hsa_amd_agent_iterate_memory_pools
|
||||
hsa_amd_memory_pool_free
|
||||
hsa_amd_agent_memory_pool_get_info
|
||||
hsa_amd_memory_migrate
|
||||
hsa_amd_memory_lock
|
||||
hsa_amd_memory_unlock
|
||||
hsa_amd_interop_map_buffer
|
||||
hsa_amd_interop_unmap_buffer
|
||||
hsa_amd_image_create
|
||||
hsa_ext_program_create
|
||||
hsa_ext_program_destroy
|
||||
hsa_ext_program_add_module
|
||||
hsa_ext_program_iterate_modules
|
||||
hsa_ext_program_get_info
|
||||
hsa_ext_program_finalize
|
||||
hsa_ext_image_data_get_info
|
||||
hsa_ext_sampler_create
|
||||
hsa_status_string
|
||||
hsa_init
|
||||
hsa_shut_down
|
||||
hsa_system_get_info
|
||||
hsa_extension_get_name
|
||||
hsa_system_extension_supported
|
||||
hsa_system_major_extension_supported
|
||||
hsa_system_get_extension_table
|
||||
hsa_system_get_major_extension_table
|
||||
hsa_agent_get_info
|
||||
hsa_iterate_agents
|
||||
hsa_agent_get_info_size
|
||||
hsa_agent_set_info
|
||||
hsa_agent_get_exception_policies
|
||||
hsa_cache_get_info
|
||||
hsa_agent_iterate_caches
|
||||
hsa_agent_extension_supported
|
||||
hsa_agent_major_extension_supported
|
||||
hsa_signal_create
|
||||
hsa_signal_destroy
|
||||
hsa_signal_load_scacquire
|
||||
hsa_signal_load_relaxed
|
||||
hsa_signal_load_acquire
|
||||
hsa_signal_store_relaxed
|
||||
hsa_signal_store_screlease
|
||||
hsa_signal_store_release
|
||||
hsa_signal_silent_store_relaxed
|
||||
hsa_signal_silent_store_screlease
|
||||
hsa_signal_exchange_scacq_screl
|
||||
hsa_signal_exchange_acq_rel
|
||||
hsa_signal_exchange_scacquire
|
||||
hsa_signal_exchange_acquire
|
||||
hsa_signal_exchange_relaxed
|
||||
hsa_signal_exchange_screlease
|
||||
hsa_signal_exchange_release
|
||||
hsa_signal_cas_scacq_screl
|
||||
hsa_signal_cas_acq_rel
|
||||
hsa_signal_cas_scacquire
|
||||
hsa_signal_cas_acquire
|
||||
hsa_signal_cas_relaxed
|
||||
hsa_signal_cas_screlease
|
||||
hsa_signal_cas_release
|
||||
hsa_signal_add_scacq_screl
|
||||
hsa_signal_add_acq_rel
|
||||
hsa_signal_add_scacquire
|
||||
hsa_signal_add_acquire
|
||||
hsa_signal_add_relaxed
|
||||
hsa_signal_add_screlease
|
||||
hsa_signal_add_release
|
||||
hsa_signal_subtract_scacq_screl
|
||||
hsa_signal_subtract_acq_rel
|
||||
hsa_signal_subtract_scacquire
|
||||
hsa_signal_subtract_acquire
|
||||
hsa_signal_subtract_relaxed
|
||||
hsa_signal_subtract_screlease
|
||||
hsa_signal_subtract_release
|
||||
hsa_signal_and_scacq_screl
|
||||
hsa_signal_and_acq_rel
|
||||
hsa_signal_and_scacquire
|
||||
hsa_signal_and_acquire
|
||||
hsa_signal_and_relaxed
|
||||
hsa_signal_and_screlease
|
||||
hsa_signal_and_release
|
||||
hsa_signal_or_scacq_screl
|
||||
hsa_signal_or_acq_rel
|
||||
hsa_signal_or_scacquire
|
||||
hsa_signal_or_acquire
|
||||
hsa_signal_or_relaxed
|
||||
hsa_signal_or_screlease
|
||||
hsa_signal_or_release
|
||||
hsa_signal_xor_scacq_screl
|
||||
hsa_signal_xor_acq_rel
|
||||
hsa_signal_xor_scacquire
|
||||
hsa_signal_xor_acquire
|
||||
hsa_signal_xor_relaxed
|
||||
hsa_signal_xor_screlease
|
||||
hsa_signal_xor_release
|
||||
hsa_signal_wait_scacquire
|
||||
hsa_signal_wait_relaxed
|
||||
hsa_signal_wait_acquire
|
||||
hsa_signal_group_create
|
||||
hsa_signal_group_destroy
|
||||
hsa_signal_group_wait_any_scacquire
|
||||
hsa_signal_group_wait_any_relaxed
|
||||
hsa_queue_create
|
||||
hsa_soft_queue_create
|
||||
hsa_queue_destroy
|
||||
hsa_queue_inactivate
|
||||
hsa_queue_load_read_index_acquire
|
||||
hsa_queue_load_read_index_scacquire
|
||||
hsa_queue_load_read_index_relaxed
|
||||
hsa_queue_load_write_index_acquire
|
||||
hsa_queue_load_write_index_scacquire
|
||||
hsa_queue_load_write_index_relaxed
|
||||
hsa_queue_store_write_index_relaxed
|
||||
hsa_queue_store_write_index_release
|
||||
hsa_queue_store_write_index_screlease
|
||||
hsa_queue_cas_write_index_acq_rel
|
||||
hsa_queue_cas_write_index_scacq_screl
|
||||
hsa_queue_cas_write_index_acquire
|
||||
hsa_queue_cas_write_index_scacquire
|
||||
hsa_queue_cas_write_index_relaxed
|
||||
hsa_queue_cas_write_index_release
|
||||
hsa_queue_cas_write_index_screlease
|
||||
hsa_queue_add_write_index_acq_rel
|
||||
hsa_queue_add_write_index_scacq_screl
|
||||
hsa_queue_add_write_index_acquire
|
||||
hsa_queue_add_write_index_scacquire
|
||||
hsa_queue_add_write_index_relaxed
|
||||
hsa_queue_add_write_index_release
|
||||
hsa_queue_add_write_index_screlease
|
||||
hsa_queue_store_read_index_relaxed
|
||||
hsa_queue_store_read_index_release
|
||||
hsa_queue_store_read_index_screlease
|
||||
hsa_region_get_info
|
||||
hsa_agent_iterate_regions
|
||||
hsa_memory_allocate
|
||||
hsa_memory_free
|
||||
hsa_memory_copy
|
||||
hsa_memory_assign_agent
|
||||
hsa_memory_register
|
||||
hsa_memory_deregister
|
||||
hsa_isa_from_name
|
||||
hsa_agent_iterate_isas
|
||||
hsa_isa_get_info
|
||||
hsa_isa_get_info_alt
|
||||
hsa_isa_get_exception_policies
|
||||
hsa_isa_get_round_method
|
||||
hsa_wavefront_get_info
|
||||
hsa_isa_iterate_wavefronts
|
||||
hsa_isa_compatible
|
||||
hsa_code_object_reader_create_from_file
|
||||
hsa_code_object_reader_create_from_memory
|
||||
hsa_code_object_reader_destroy
|
||||
hsa_executable_create
|
||||
hsa_executable_create_alt
|
||||
hsa_executable_destroy
|
||||
hsa_executable_load_program_code_object
|
||||
hsa_executable_load_agent_code_object
|
||||
hsa_executable_freeze
|
||||
hsa_executable_get_info
|
||||
hsa_executable_global_variable_define
|
||||
hsa_executable_agent_global_variable_define
|
||||
hsa_executable_readonly_variable_define
|
||||
hsa_executable_validate
|
||||
hsa_executable_validate_alt
|
||||
hsa_executable_get_symbol
|
||||
hsa_executable_get_symbol_by_name
|
||||
hsa_executable_symbol_get_info
|
||||
hsa_executable_iterate_symbols
|
||||
hsa_executable_iterate_agent_symbols
|
||||
hsa_executable_iterate_program_symbols
|
||||
hsa_code_object_serialize
|
||||
hsa_code_object_deserialize
|
||||
hsa_code_object_destroy
|
||||
hsa_code_object_get_info
|
||||
hsa_executable_load_code_object
|
||||
hsa_code_object_get_symbol
|
||||
hsa_code_object_get_symbol_from_name
|
||||
hsa_code_symbol_get_info
|
||||
hsa_code_object_iterate_symbols
|
||||
hsa_ven_amd_loader_query_host_address
|
||||
hsa_ven_amd_loader_query_segment_descriptors
|
||||
hsa_amd_profiling_set_profiler_enabled
|
||||
hsa_amd_profiling_async_copy_enable
|
||||
hsa_amd_profiling_convert_tick_to_system_domain
|
||||
hsa_amd_signal_async_handler
|
||||
hsa_amd_async_function
|
||||
hsa_amd_signal_wait_any
|
||||
hsa_amd_memory_pool_get_info
|
||||
hsa_amd_memory_pool_allocate
|
||||
hsa_amd_memory_async_copy
|
||||
hsa_amd_agents_allow_access
|
||||
hsa_amd_memory_pool_can_migrate
|
||||
hsa_amd_memory_fill
|
||||
|
||||
hsa_ext_image_get_capability
|
||||
hsa_ext_image_create
|
||||
hsa_ext_image_destroy
|
||||
hsa_ext_image_copy
|
||||
hsa_ext_image_import
|
||||
hsa_ext_image_export
|
||||
hsa_ext_image_clear
|
||||
hsa_ext_sampler_destroy
|
||||
@@ -36,7 +36,7 @@ union TData {
|
||||
#define __TEXTURE_FUNCTIONS_DECL__ static __inline__ __device__
|
||||
|
||||
|
||||
#if (__hcc_workweek__ >= 18115)
|
||||
#if (__hcc_workweek__ >= 18114)
|
||||
#define ADDRESS_SPACE_CONSTANT __attribute__((address_space(4)))
|
||||
#else
|
||||
#define ADDRESS_SPACE_CONSTANT __attribute__((address_space(2)))
|
||||
|
||||
Ссылка в новой задаче
Block a user